From 3fcdce08ab494f78d691ab5c7dc26b0891ee82a0 Mon Sep 17 00:00:00 2001
From: Bill Hollings <bill.hollings@brenwill.com>
Date: Tue, 26 Dec 2017 13:39:07 -0500
Subject: [PATCH] CompilerMSL support platform semantics.

Support customizing MSL based on iOS or macOS platform.
Support SPIV-V containing multiple memory semantics.
---
 spirv_msl.cpp | 30 +++++++-----------------------
 spirv_msl.hpp | 15 +++++++++++++++
 2 files changed, 22 insertions(+), 23 deletions(-)

diff --git a/spirv_msl.cpp b/spirv_msl.cpp
index 087c4150..dd58121c 100644
--- a/spirv_msl.cpp
+++ b/spirv_msl.cpp
@@ -1559,34 +1559,18 @@ void CompilerMSL::emit_barrier(uint32_t id_exe_scope, uint32_t id_mem_scope, uin
 	string bar_stmt = "threadgroup_barrier(mem_flags::";
 
 	uint32_t mem_sem = id_mem_sem ? get<SPIRConstant>(id_mem_sem).scalar() : uint32_t(MemorySemanticsMaskNone);
-	switch (mem_sem)
-	{
-	case MemorySemanticsCrossWorkgroupMemoryMask:
+
+	if (mem_sem & MemorySemanticsCrossWorkgroupMemoryMask)
 		bar_stmt += "mem_device";
-		break;
-
-	case MemorySemanticsSubgroupMemoryMask:
-	case MemorySemanticsWorkgroupMemoryMask:
-	case MemorySemanticsAtomicCounterMemoryMask:
+	else if (mem_sem & (MemorySemanticsSubgroupMemoryMask | MemorySemanticsWorkgroupMemoryMask |
+	                    MemorySemanticsAtomicCounterMemoryMask))
 		bar_stmt += "mem_threadgroup";
-		break;
-
-	case MemorySemanticsImageMemoryMask:
+	else if (mem_sem & MemorySemanticsImageMemoryMask)
 		bar_stmt += "mem_texture";
-		break;
-
-	case MemorySemanticsAcquireMask:
-	case MemorySemanticsReleaseMask:
-	case MemorySemanticsAcquireReleaseMask:
-	case MemorySemanticsSequentiallyConsistentMask:
-	case MemorySemanticsUniformMemoryMask:
-	case MemorySemanticsMaskNone:
-	default:
+	else
 		bar_stmt += "mem_none";
-		break;
-	}
 
-	if (options.supports_msl_version(2))
+	if (options.is_ios() && options.supports_msl_version(2))
 	{
 		bar_stmt += ", ";
 
diff --git a/spirv_msl.hpp b/spirv_msl.hpp
index a7082ae3..3ec98397 100644
--- a/spirv_msl.hpp
+++ b/spirv_msl.hpp
@@ -77,10 +77,25 @@ public:
 	// Options for compiling to Metal Shading Language
 	struct Options
 	{
+		typedef enum {
+			iOS,
+			macOS,
+		} Platform;
+
+		Platform platform = macOS;
 		uint32_t msl_version = make_msl_version(1, 2);
 		bool enable_point_size_builtin = true;
 		bool resolve_specialized_array_lengths = true;
 
+		bool is_ios()
+		{
+			return platform == iOS;
+		}
+		bool is_macos()
+		{
+			return platform == macOS;
+		}
+
 		void set_msl_version(uint32_t major, uint32_t minor = 0, uint32_t patch = 0)
 		{
 			msl_version = make_msl_version(major, minor, patch);