Updated spirv-cross.

3rdparty/spirv-cross/CMakeLists.txt

@@ -195,7 +195,7 @@ endif()
 
 if (SPIRV_CROSS_SHARED)
 	set(spirv-cross-abi-major 0)
-	set(spirv-cross-abi-minor 1)
+	set(spirv-cross-abi-minor 2)
 	set(spirv-cross-abi-patch 0)
 	set(SPIRV_CROSS_VERSION ${spirv-cross-abi-major}.${spirv-cross-abi-minor}.${spirv-cross-abi-patch})
 	set(SPIRV_CROSS_INSTALL_LIB_DIR ${CMAKE_INSTALL_PREFIX}/lib)

3rdparty/spirv-cross/main.cpp

@@ -495,6 +495,8 @@ struct CLIArguments
 	bool msl_ios = false;
 	bool msl_pad_fragment_output = false;
 	bool msl_domain_lower_left = false;
+	bool msl_argument_buffers = false;
+	vector<uint32_t> msl_discrete_descriptor_sets;
 	vector<PLSArg> pls_in;
 	vector<PLSArg> pls_out;
 	vector<Remap> remaps;
@@ -552,6 +554,8 @@ static void print_help()
 	                "\t[--msl-ios]\n"
 	                "\t[--msl-pad-fragment-output]\n"
 	                "\t[--msl-domain-lower-left]\n"
+	                "\t[--msl-argument-buffers]\n"
+	                "\t[--msl-discrete-descriptor-set <index>]\n"
 	                "\t[--hlsl]\n"
 	                "\t[--reflect]\n"
 	                "\t[--shader-model]\n"
@@ -723,6 +727,9 @@ static int main_inner(int argc, char *argv[])
 	cbs.add("--msl-ios", [&args](CLIParser &) { args.msl_ios = true; });
 	cbs.add("--msl-pad-fragment-output", [&args](CLIParser &) { args.msl_pad_fragment_output = true; });
 	cbs.add("--msl-domain-lower-left", [&args](CLIParser &) { args.msl_domain_lower_left = true; });
+	cbs.add("--msl-argument-buffers", [&args](CLIParser &) { args.msl_argument_buffers = true; });
+	cbs.add("--msl-discrete-descriptor-set",
+	        [&args](CLIParser &parser) { args.msl_discrete_descriptor_sets.push_back(parser.next_uint()); });
 	cbs.add("--extension", [&args](CLIParser &parser) { args.extensions.push_back(parser.next_string()); });
 	cbs.add("--rename-entry-point", [&args](CLIParser &parser) {
 		auto old_name = parser.next_string();
@@ -855,7 +862,10 @@ static int main_inner(int argc, char *argv[])
 			msl_opts.platform = CompilerMSL::Options::iOS;
 		msl_opts.pad_fragment_output_components = args.msl_pad_fragment_output;
 		msl_opts.tess_domain_origin_lower_left = args.msl_domain_lower_left;
+		msl_opts.argument_buffers = args.msl_argument_buffers;
 		msl_comp->set_msl_options(msl_opts);
+		for (auto &v : args.msl_discrete_descriptor_sets)
+			msl_comp->add_discrete_descriptor_set(v);
 	}
 	else if (args.hlsl)
 		compiler.reset(new CompilerHLSL(move(spirv_parser.get_parsed_ir())));

3rdparty/spirv-cross/reference/opt/shaders-msl/comp/argument-buffers-discrete.msl2.argument.discrete.comp

@@ -0,0 +1,40 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct SSBO3
+{
+    float4 v;
+};
+
+struct SSBO0
+{
+    float4 v;
+};
+
+struct SSBO1
+{
+    float4 v;
+};
+
+struct SSBO2
+{
+    float4 v;
+};
+
+struct spvDescriptorSetBuffer0
+{
+    const device SSBO0* ssbo0 [[id(0)]];
+};
+
+struct spvDescriptorSetBuffer1
+{
+    const device SSBO1* ssbo1 [[id(0)]];
+};
+
+kernel void main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], constant spvDescriptorSetBuffer1& spvDescriptorSet1 [[buffer(1)]], const device SSBO2& ssbo2 [[buffer(5)]], device SSBO3& ssbo3 [[buffer(6)]])
+{
+    ssbo3.v = ((*spvDescriptorSet0.ssbo0).v + (*spvDescriptorSet1.ssbo1).v) + ssbo2.v;
+}
+

3rdparty/spirv-cross/reference/opt/shaders-msl/comp/argument-buffers-image-load-store.msl2.argument.comp

@@ -0,0 +1,17 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct spvDescriptorSetBuffer0
+{
+    texture2d<float, access::write> uImage [[id(1)]];
+    texture2d<float> uImageRead [[id(2)]];
+};
+
+kernel void main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
+{
+    int2 _17 = int2(gl_GlobalInvocationID.xy);
+    spvDescriptorSet0.uImage.write(spvDescriptorSet0.uImageRead.read(uint2(_17)), uint2(_17));
+}
+

3rdparty/spirv-cross/reference/opt/shaders-msl/frag/argument-buffers.msl2.argument.frag

@@ -0,0 +1,73 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct SSBO
+{
+    float4 ssbo;
+};
+
+struct SSBOs
+{
+    float4 ssbo;
+};
+
+struct Push
+{
+    float4 push;
+};
+
+struct UBO
+{
+    float4 ubo;
+};
+
+struct UBOs
+{
+    float4 ubo;
+};
+
+struct spvDescriptorSetBuffer0
+{
+    texture2d<float> uTexture [[id(2)]];
+    sampler uTextureSmplr [[id(3)]];
+    constant UBO* m_90 [[id(5)]];
+    array<texture2d<float>, 2> uTextures [[id(6)]];
+    array<sampler, 2> uTexturesSmplr [[id(8)]];
+};
+
+struct spvDescriptorSetBuffer1
+{
+    array<texture2d<float>, 4> uTexture2 [[id(3)]];
+    device SSBO* m_60 [[id(7)]];
+    const device SSBOs* ssbos [[id(8)]][2];
+    array<sampler, 2> uSampler [[id(10)]];
+};
+
+struct spvDescriptorSetBuffer2
+{
+    constant UBOs* ubos [[id(4)]][4];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+struct main0_in
+{
+    float2 vUV [[user(locn0)]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], constant spvDescriptorSetBuffer1& spvDescriptorSet1 [[buffer(1)]], constant spvDescriptorSetBuffer2& spvDescriptorSet2 [[buffer(2)]], constant Push& registers [[buffer(3)]])
+{
+    main0_out out = {};
+    out.FragColor = ((((((spvDescriptorSet0.uTexture.sample(spvDescriptorSet0.uTextureSmplr, in.vUV) + spvDescriptorSet1.uTexture2[2].sample(spvDescriptorSet1.uSampler[1], in.vUV)) + spvDescriptorSet0.uTextures[1].sample(spvDescriptorSet0.uTexturesSmplr[1], in.vUV)) + (*spvDescriptorSet1.m_60).ssbo) + spvDescriptorSet1.ssbos[0]->ssbo) + registers.push) + (*spvDescriptorSet0.m_90).ubo) + spvDescriptorSet2.ubos[0]->ubo;
+    out.FragColor += (*spvDescriptorSet0.m_90).ubo;
+    out.FragColor += (*spvDescriptorSet1.m_60).ssbo;
+    out.FragColor += spvDescriptorSet2.ubos[1]->ubo;
+    out.FragColor += registers.push;
+    return out;
+}
+

3rdparty/spirv-cross/reference/shaders-msl/comp/argument-buffers-discrete.msl2.argument.discrete.comp

@@ -0,0 +1,40 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct SSBO3
+{
+    float4 v;
+};
+
+struct SSBO0
+{
+    float4 v;
+};
+
+struct SSBO1
+{
+    float4 v;
+};
+
+struct SSBO2
+{
+    float4 v;
+};
+
+struct spvDescriptorSetBuffer0
+{
+    const device SSBO0* ssbo0 [[id(0)]];
+};
+
+struct spvDescriptorSetBuffer1
+{
+    const device SSBO1* ssbo1 [[id(0)]];
+};
+
+kernel void main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], constant spvDescriptorSetBuffer1& spvDescriptorSet1 [[buffer(1)]], const device SSBO2& ssbo2 [[buffer(5)]], device SSBO3& ssbo3 [[buffer(6)]])
+{
+    ssbo3.v = ((*spvDescriptorSet0.ssbo0).v + (*spvDescriptorSet1.ssbo1).v) + ssbo2.v;
+}
+

3rdparty/spirv-cross/reference/shaders-msl/comp/argument-buffers-image-load-store.msl2.argument.comp

@@ -0,0 +1,17 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct spvDescriptorSetBuffer0
+{
+    texture2d<float, access::write> uImage [[id(1)]];
+    texture2d<float> uImageRead [[id(2)]];
+};
+
+kernel void main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
+{
+    int2 coord = int2(gl_GlobalInvocationID.xy);
+    spvDescriptorSet0.uImage.write(spvDescriptorSet0.uImageRead.read(uint2(coord)), uint2(coord));
+}
+

3rdparty/spirv-cross/reference/shaders-msl/frag/argument-buffers.msl2.argument.frag

@@ -0,0 +1,94 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct SSBO
+{
+    float4 ssbo;
+};
+
+struct SSBOs
+{
+    float4 ssbo;
+};
+
+struct Push
+{
+    float4 push;
+};
+
+struct UBO
+{
+    float4 ubo;
+};
+
+struct UBOs
+{
+    float4 ubo;
+};
+
+struct spvDescriptorSetBuffer0
+{
+    texture2d<float> uTexture [[id(2)]];
+    sampler uTextureSmplr [[id(3)]];
+    constant UBO* v_90 [[id(5)]];
+    array<texture2d<float>, 2> uTextures [[id(6)]];
+    array<sampler, 2> uTexturesSmplr [[id(8)]];
+};
+
+struct spvDescriptorSetBuffer1
+{
+    array<texture2d<float>, 4> uTexture2 [[id(3)]];
+    device SSBO* v_60 [[id(7)]];
+    const device SSBOs* ssbos [[id(8)]][2];
+    array<sampler, 2> uSampler [[id(10)]];
+};
+
+struct spvDescriptorSetBuffer2
+{
+    constant UBOs* ubos [[id(4)]][4];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+struct main0_in
+{
+    float2 vUV [[user(locn0)]];
+};
+
+float4 sample_in_function2(thread texture2d<float> uTexture, thread const sampler uTextureSmplr, thread float2& vUV, thread const array<texture2d<float>, 4> uTexture2, thread const array<sampler, 2> uSampler, thread const array<texture2d<float>, 2> uTextures, thread const array<sampler, 2> uTexturesSmplr, device SSBO& v_60, const device SSBOs* constant (&ssbos)[2], constant Push& registers)
+{
+    float4 ret = uTexture.sample(uTextureSmplr, vUV);
+    ret += uTexture2[2].sample(uSampler[1], vUV);
+    ret += uTextures[1].sample(uTexturesSmplr[1], vUV);
+    ret += v_60.ssbo;
+    ret += ssbos[0]->ssbo;
+    ret += registers.push;
+    return ret;
+}
+
+float4 sample_in_function(thread texture2d<float> uTexture, thread const sampler uTextureSmplr, thread float2& vUV, thread const array<texture2d<float>, 4> uTexture2, thread const array<sampler, 2> uSampler, thread const array<texture2d<float>, 2> uTextures, thread const array<sampler, 2> uTexturesSmplr, device SSBO& v_60, const device SSBOs* constant (&ssbos)[2], constant Push& registers, constant UBO& v_90, constant UBOs* constant (&ubos)[4])
+{
+    float4 ret = sample_in_function2(uTexture, uTextureSmplr, vUV, uTexture2, uSampler, uTextures, uTexturesSmplr, v_60, ssbos, registers);
+    ret += v_90.ubo;
+    ret += ubos[0]->ubo;
+    return ret;
+}
+
+fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], constant spvDescriptorSetBuffer1& spvDescriptorSet1 [[buffer(1)]], constant spvDescriptorSetBuffer2& spvDescriptorSet2 [[buffer(2)]], constant Push& registers [[buffer(3)]])
+{
+    main0_out out = {};
+    out.FragColor = sample_in_function(spvDescriptorSet0.uTexture, spvDescriptorSet0.uTextureSmplr, in.vUV, spvDescriptorSet1.uTexture2, spvDescriptorSet1.uSampler, spvDescriptorSet0.uTextures, spvDescriptorSet0.uTexturesSmplr, (*spvDescriptorSet1.v_60), spvDescriptorSet1.ssbos, registers, (*spvDescriptorSet0.v_90), spvDescriptorSet2.ubos);
+    out.FragColor += (*spvDescriptorSet0.v_90).ubo;
+    out.FragColor += (*spvDescriptorSet1.v_60).ssbo;
+    out.FragColor += spvDescriptorSet2.ubos[1]->ubo;
+    out.FragColor += registers.push;
+    return out;
+}
+

3rdparty/spirv-cross/shaders-msl/comp/argument-buffers-discrete.msl2.argument.discrete.comp

@@ -0,0 +1,27 @@
+#version 450
+layout(local_size_x = 1) in;
+
+layout(set = 0, binding = 0) readonly buffer SSBO0
+{
+   vec4 v;
+} ssbo0;
+
+layout(set = 1, binding = 0) readonly buffer SSBO1
+{
+   vec4 v;
+} ssbo1;
+
+layout(set = 2, binding = 5) readonly buffer SSBO2
+{
+   vec4 v;
+} ssbo2;
+
+layout(set = 3, binding = 6) writeonly buffer SSBO3
+{
+   vec4 v;
+} ssbo3;
+
+void main()
+{
+   ssbo3.v = ssbo0.v + ssbo1.v + ssbo2.v;
+}

3rdparty/spirv-cross/shaders-msl/comp/argument-buffers-image-load-store.msl2.argument.comp

@@ -0,0 +1,10 @@
+#version 450
+
+layout(set = 0, binding = 1, r32f) writeonly uniform image2D uImage;
+layout(set = 0, binding = 2, r32f) readonly uniform image2D uImageRead;
+
+void main()
+{
+   ivec2 coord = ivec2(gl_GlobalInvocationID.xy);
+   imageStore(uImage, coord, imageLoad(uImageRead, coord));
+}

3rdparty/spirv-cross/shaders-msl/frag/argument-buffers.msl2.argument.frag

@@ -0,0 +1,61 @@
+#version 450
+
+layout(std430, push_constant) uniform Push
+{
+   vec4 push;
+} registers;
+
+layout(std140, set = 0, binding = 5) uniform UBO
+{
+   vec4 ubo;
+};
+
+layout(std430, set = 1, binding = 7) buffer SSBO
+{
+   vec4 ssbo;
+};
+
+layout(std430, set = 1, binding = 8) readonly buffer SSBOs
+{
+   vec4 ssbo;
+} ssbos[2];
+
+layout(std140, set = 2, binding = 4) uniform UBOs
+{
+   vec4 ubo;
+} ubos[4];
+
+layout(set = 0, binding = 2) uniform sampler2D uTexture;
+layout(set = 0, binding = 6) uniform sampler2D uTextures[2];
+layout(set = 1, binding = 3) uniform texture2D uTexture2[4];
+layout(set = 1, binding = 10) uniform sampler uSampler[2];
+layout(location = 0) in vec2 vUV;
+layout(location = 0) out vec4 FragColor;
+
+vec4 sample_in_function2()
+{
+   vec4 ret = texture(uTexture, vUV);
+   ret += texture(sampler2D(uTexture2[2], uSampler[1]), vUV);
+   ret += texture(uTextures[1], vUV);
+   ret += ssbo;
+   ret += ssbos[0].ssbo;
+   ret += registers.push;
+   return ret;
+}
+
+vec4 sample_in_function()
+{
+   vec4 ret = sample_in_function2();
+   ret += ubo;
+   ret += ubos[0].ubo;
+   return ret;
+}
+
+void main()
+{
+   FragColor = sample_in_function();
+   FragColor += ubo;
+   FragColor += ssbo;
+   FragColor += ubos[1].ubo;
+   FragColor += registers.push;
+}

3rdparty/spirv-cross/spirv_common.hpp

@@ -1404,8 +1404,9 @@ struct Meta
 		{
 			uint32_t packed_type = 0;
 			bool packed = false;
-			uint32_t ib_member_index = static_cast<uint32_t>(-1);
+			uint32_t ib_member_index = ~(0u);
 			uint32_t ib_orig_id = 0;
+			uint32_t argument_buffer_id = ~(0u);
 		} extended;
 	};
 

3rdparty/spirv-cross/spirv_cross.cpp

@@ -1091,6 +1091,11 @@ const std::string &Compiler::get_member_name(uint32_t id, uint32_t index) const
 	return ir.get_member_name(id, index);
 }
 
+void Compiler::set_qualified_name(uint32_t id, const string &name)
+{
+	ir.meta[id].decoration.qualified_alias = name;
+}
+
 void Compiler::set_member_qualified_name(uint32_t type_id, uint32_t index, const std::string &name)
 {
 	ir.meta[type_id].members.resize(max(ir.meta[type_id].members.size(), size_t(index) + 1));
@@ -1156,6 +1161,10 @@ void Compiler::set_extended_decoration(uint32_t id, ExtendedDecorations decorati
 	case SPIRVCrossDecorationInterfaceOrigID:
 		dec.extended.ib_orig_id = value;
 		break;
+
+	case SPIRVCrossDecorationArgumentBufferID:
+		dec.extended.argument_buffer_id = value;
+		break;
 	}
 }
 
@@ -1182,6 +1191,10 @@ void Compiler::set_extended_member_decoration(uint32_t type, uint32_t index, Ext
 	case SPIRVCrossDecorationInterfaceOrigID:
 		dec.extended.ib_orig_id = value;
 		break;
+
+	case SPIRVCrossDecorationArgumentBufferID:
+		dec.extended.argument_buffer_id = value;
+		break;
 	}
 }
 
@@ -1205,6 +1218,9 @@ uint32_t Compiler::get_extended_decoration(uint32_t id, ExtendedDecorations deco
 
 	case SPIRVCrossDecorationInterfaceOrigID:
 		return dec.extended.ib_orig_id;
+
+	case SPIRVCrossDecorationArgumentBufferID:
+		return dec.extended.argument_buffer_id;
 	}
 
 	return 0;
@@ -1233,6 +1249,9 @@ uint32_t Compiler::get_extended_member_decoration(uint32_t type, uint32_t index,
 
 	case SPIRVCrossDecorationInterfaceOrigID:
 		return dec.extended.ib_orig_id;
+
+	case SPIRVCrossDecorationArgumentBufferID:
+		return dec.extended.argument_buffer_id;
 	}
 
 	return 0;
@@ -1258,6 +1277,9 @@ bool Compiler::has_extended_decoration(uint32_t id, ExtendedDecorations decorati
 
 	case SPIRVCrossDecorationInterfaceOrigID:
 		return dec.extended.ib_orig_id != 0;
+
+	case SPIRVCrossDecorationArgumentBufferID:
+		return dec.extended.argument_buffer_id != 0;
 	}
 
 	return false;
@@ -1286,6 +1308,9 @@ bool Compiler::has_extended_member_decoration(uint32_t type, uint32_t index, Ext
 
 	case SPIRVCrossDecorationInterfaceOrigID:
 		return dec.extended.ib_orig_id != 0;
+
+	case SPIRVCrossDecorationArgumentBufferID:
+		return dec.extended.argument_buffer_id != uint32_t(-1);
 	}
 
 	return false;
@@ -1305,12 +1330,16 @@ void Compiler::unset_extended_decoration(uint32_t id, ExtendedDecorations decora
 		break;
 
 	case SPIRVCrossDecorationInterfaceMemberIndex:
-		dec.extended.ib_member_index = -1;
+		dec.extended.ib_member_index = ~(0u);
 		break;
 
 	case SPIRVCrossDecorationInterfaceOrigID:
 		dec.extended.ib_orig_id = 0;
 		break;
+
+	case SPIRVCrossDecorationArgumentBufferID:
+		dec.extended.argument_buffer_id = 0;
+		break;
 	}
 }
 
@@ -1330,12 +1359,16 @@ void Compiler::unset_extended_member_decoration(uint32_t type, uint32_t index, E
 		break;
 
 	case SPIRVCrossDecorationInterfaceMemberIndex:
-		dec.extended.ib_member_index = -1;
+		dec.extended.ib_member_index = ~(0u);
 		break;
 
 	case SPIRVCrossDecorationInterfaceOrigID:
 		dec.extended.ib_orig_id = 0;
 		break;
+
+	case SPIRVCrossDecorationArgumentBufferID:
+		dec.extended.argument_buffer_id = 0;
+		break;
 	}
 }
 

3rdparty/spirv-cross/spirv_cross.hpp

@@ -120,6 +120,7 @@ enum ExtendedDecorations
 	SPIRVCrossDecorationPackedType,
 	SPIRVCrossDecorationInterfaceMemberIndex,
 	SPIRVCrossDecorationInterfaceOrigID,
+	SPIRVCrossDecorationArgumentBufferID
 };
 
 class Compiler
@@ -209,9 +210,6 @@ public:
 	// or an empty string if no qualified alias exists
 	const std::string &get_member_qualified_name(uint32_t type_id, uint32_t index) const;
 
-	// Sets the qualified member identifier for OpTypeStruct ID, member number "index".
-	void set_member_qualified_name(uint32_t type_id, uint32_t index, const std::string &name);
-
 	// Gets the decoration mask for a member of a struct, similar to get_decoration_mask.
 	const Bitset &get_member_decoration_bitset(uint32_t id, uint32_t index) const;
 
@@ -587,6 +585,10 @@ protected:
 	// Gets the SPIR-V element type underlying an array variable.
 	const SPIRType &get_variable_element_type(const SPIRVariable &var) const;
 
+	// Sets the qualified member identifier for OpTypeStruct ID, member number "index".
+	void set_member_qualified_name(uint32_t type_id, uint32_t index, const std::string &name);
+	void set_qualified_name(uint32_t id, const std::string &name);
+
 	// Returns if the given type refers to a sampled image.
 	bool is_sampled_image_type(const SPIRType &type);
 

3rdparty/spirv-cross/spirv_cross_c.cpp

@@ -470,6 +470,10 @@ spvc_result spvc_compiler_options_set_uint(spvc_compiler_options options, spvc_c
 		options->msl.platform = static_cast<CompilerMSL::Options::Platform>(value);
 		break;
 
+	case SPVC_COMPILER_OPTION_MSL_ARGUMENT_BUFFERS:
+		options->msl.argument_buffers = value != 0;
+		break;
+
 	default:
 		options->context->report_error("Unknown option.");
 		return SPVC_ERROR_INVALID_ARGUMENT;
@@ -699,6 +703,19 @@ spvc_result spvc_compiler_msl_add_resource_binding(spvc_compiler compiler,
 	return SPVC_SUCCESS;
 }
 
+spvc_result spvc_compiler_msl_add_discrete_descriptor_set(spvc_compiler compiler, unsigned desc_set)
+{
+	if (compiler->backend != SPVC_BACKEND_MSL)
+	{
+		compiler->context->report_error("MSL function used on a non-MSL backend.");
+		return SPVC_ERROR_INVALID_ARGUMENT;
+	}
+
+	auto &msl = *static_cast<CompilerMSL *>(compiler->compiler.get());
+	msl.add_discrete_descriptor_set(desc_set);
+	return SPVC_SUCCESS;
+}
+
 spvc_bool spvc_compiler_msl_is_vertex_attribute_used(spvc_compiler compiler, unsigned location)
 {
 	if (compiler->backend != SPVC_BACKEND_MSL)

3rdparty/spirv-cross/spirv_cross_c.h

@@ -33,7 +33,7 @@ extern "C" {
 /* Bumped if ABI or API breaks backwards compatibility. */
 #define SPVC_C_API_VERSION_MAJOR 0
 /* Bumped if APIs or enumerations are added in a backwards compatible way. */
-#define SPVC_C_API_VERSION_MINOR 1
+#define SPVC_C_API_VERSION_MINOR 2
 /* Bumped if internal implementation details change. */
 #define SPVC_C_API_VERSION_PATCH 0
 
@@ -418,6 +418,7 @@ typedef enum spvc_compiler_option
 	SPVC_COMPILER_OPTION_MSL_PAD_FRAGMENT_OUTPUT_COMPONENTS = 29 | SPVC_COMPILER_OPTION_MSL_BIT,
 	SPVC_COMPILER_OPTION_MSL_TESS_DOMAIN_ORIGIN_LOWER_LEFT = 30 | SPVC_COMPILER_OPTION_MSL_BIT,
 	SPVC_COMPILER_OPTION_MSL_PLATFORM = 31 | SPVC_COMPILER_OPTION_MSL_BIT,
+	SPVC_COMPILER_OPTION_MSL_ARGUMENT_BUFFERS = 32 | SPVC_COMPILER_OPTION_MSL_BIT,
 
 	SPVC_COMPILER_OPTION_INT_MAX = 0x7fffffff
 } spvc_compiler_option;
@@ -504,6 +505,7 @@ SPVC_PUBLIC_API spvc_result spvc_compiler_msl_add_vertex_attribute(spvc_compiler
                                                                    const spvc_msl_vertex_attribute *attrs);
 SPVC_PUBLIC_API spvc_result spvc_compiler_msl_add_resource_binding(spvc_compiler compiler,
                                                                    const spvc_msl_resource_binding *binding);
+SPVC_PUBLIC_API spvc_result spvc_compiler_msl_add_discrete_descriptor_set(spvc_compiler compiler, unsigned desc_set);
 SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_is_vertex_attribute_used(spvc_compiler compiler, unsigned location);
 SPVC_PUBLIC_API spvc_bool spvc_compiler_msl_is_resource_used(spvc_compiler compiler,
                                                              SpvExecutionModel model,

3rdparty/spirv-cross/spirv_msl.cpp

@@ -62,6 +62,12 @@ void CompilerMSL::add_msl_resource_binding(const MSLResourceBinding &binding)
 	resource_bindings.push_back({ binding, false });
 }
 
+void CompilerMSL::add_discrete_descriptor_set(uint32_t desc_set)
+{
+	if (desc_set < kMaxArgumentBuffers)
+		argument_buffer_discrete_mask |= 1u << desc_set;
+}
+
 bool CompilerMSL::is_msl_vertex_attribute_used(uint32_t location)
 {
 	return vtx_attrs_in_use.count(location) != 0;
@@ -630,6 +636,15 @@ string CompilerMSL::compile()
 	// the loop, so the hooks aren't added multiple times.
 	fix_up_shader_inputs_outputs();
 
+	// If we are using argument buffers, we create argument buffer structures for them here.
+	// These buffers will be used in the entry point, not the individual resources.
+	if (msl_options.argument_buffers)
+	{
+		if (!msl_options.supports_msl_version(2, 0))
+			SPIRV_CROSS_THROW("Argument buffers can only be used with MSL 2.0 and up.");
+		analyze_argument_buffers();
+	}
+
 	uint32_t pass_count = 0;
 	do
 	{
@@ -4271,7 +4286,10 @@ void CompilerMSL::emit_function_prototype(SPIRFunction &func, const Bitset &)
 
 	if (processing_entry_point)
 	{
-		decl += entry_point_args(!func.arguments.empty());
+		if (msl_options.argument_buffers)
+			decl += entry_point_args_argument_buffer(!func.arguments.empty());
+		else
+			decl += entry_point_args_classic(!func.arguments.empty());
 
 		// If entry point function has variables that require early declaration,
 		// ensure they each have an empty initializer, creating one if needed.
@@ -4778,7 +4796,15 @@ string CompilerMSL::to_func_call_arg(uint32_t id)
 	// Manufacture automatic sampler arg if the arg is a SampledImage texture.
 	auto &type = expression_type(id);
 	if (type.basetype == SPIRType::SampledImage && type.image.dim != DimBuffer)
-		arg_str += ", " + to_sampler_expression(id);
+	{
+		// Need to check the base variable in case we need to apply a qualified alias.
+		uint32_t var_id = 0;
+		auto *sampler_var = maybe_get<SPIRVariable>(id);
+		if (sampler_var)
+			var_id = sampler_var->basevariable;
+
+		arg_str += ", " + to_sampler_expression(var_id ? var_id : id);
+	}
 	if (msl_options.swizzle_texture_samples && has_sampled_images && is_sampled_image_type(type))
 		arg_str += ", " + to_swizzle_expression(id);
 
@@ -4963,6 +4989,10 @@ string CompilerMSL::to_struct_member(const SPIRType &type, uint32_t member_type_
 	const SPIRType *effective_membertype = &membertype;
 	SPIRType override_type;
 
+	uint32_t orig_id = 0;
+	if (has_extended_member_decoration(type.self, index, SPIRVCrossDecorationInterfaceOrigID))
+		orig_id = get_extended_member_decoration(type.self, index, SPIRVCrossDecorationInterfaceOrigID);
+
 	if (member_is_packed_type(type, index))
 	{
 		// If we're packing a matrix, output an appropriate typedef
@@ -4992,8 +5022,23 @@ string CompilerMSL::to_struct_member(const SPIRType &type, uint32_t member_type_
 			pack_pfx = "packed_";
 	}
 
-	return join(pack_pfx, type_to_glsl(*effective_membertype), " ", qualifier, to_member_name(type, index),
-	            member_attribute_qualifier(type, index), type_to_array_glsl(membertype), ";");
+	// Very specifically, image load-store in argument buffers are disallowed on MSL on iOS.
+	if (msl_options.is_ios() && membertype.basetype == SPIRType::Image && membertype.image.sampled == 2)
+	{
+		if (!has_decoration(orig_id, DecorationNonWritable))
+			SPIRV_CROSS_THROW("Writable images are not allowed in argument buffers on iOS.");
+	}
+
+	// Array information is baked into these types.
+	string array_type;
+	if (membertype.basetype != SPIRType::Image && membertype.basetype != SPIRType::Sampler &&
+	    membertype.basetype != SPIRType::SampledImage)
+	{
+		array_type = type_to_array_glsl(membertype);
+	}
+
+	return join(pack_pfx, type_to_glsl(*effective_membertype, orig_id), " ", qualifier, to_member_name(type, index),
+	            member_attribute_qualifier(type, index), array_type, ";");
 }
 
 // Emit a structure member, padding and packing to maintain the correct memeber alignments.
@@ -5014,6 +5059,10 @@ string CompilerMSL::member_attribute_qualifier(const SPIRType &type, uint32_t in
 	BuiltIn builtin = BuiltInMax;
 	bool is_builtin = is_member_builtin(type, index, &builtin);
 
+	if (has_extended_member_decoration(type.self, index, SPIRVCrossDecorationArgumentBufferID))
+		return join(" [[id(", get_extended_member_decoration(type.self, index, SPIRVCrossDecorationArgumentBufferID),
+		            ")]]");
+
 	// Vertex function inputs
 	if (execution.model == ExecutionModelVertex && type.storage == StorageClassInput)
 	{
@@ -5393,7 +5442,7 @@ string CompilerMSL::get_argument_address_space(const SPIRVariable &argument)
 	return "thread";
 }
 
-string CompilerMSL::get_type_address_space(const SPIRType &type)
+string CompilerMSL::get_type_address_space(const SPIRType &type, uint32_t id)
 {
 	switch (type.storage)
 	{
@@ -5401,8 +5450,16 @@ string CompilerMSL::get_type_address_space(const SPIRType &type)
 		return "threadgroup";
 
 	case StorageClassStorageBuffer:
-		// FIXME: Need to use 'const device' for pointers into non-writable SSBOs
-		return "device";
+	{
+		// This can be called for variable pointer contexts as well, so be very careful about which method we choose.
+		Bitset flags;
+		if (ir.ids[id].get_type() == TypeVariable && has_decoration(type.self, DecorationBlock))
+			flags = get_buffer_block_flags(id);
+		else
+			flags = get_decoration_bitset(id);
+
+		return flags.get(DecorationNonWritable) ? "const device" : "device";
+	}
 
 	case StorageClassUniform:
 	case StorageClassUniformConstant:
@@ -5410,9 +5467,17 @@ string CompilerMSL::get_type_address_space(const SPIRType &type)
 		if (type.basetype == SPIRType::Struct)
 		{
 			bool ssbo = has_decoration(type.self, DecorationBufferBlock);
-			// FIXME: Need to use 'const device' for pointers into non-writable SSBOs
 			if (ssbo)
-				return "device";
+			{
+				// This can be called for variable pointer contexts as well, so be very careful about which method we choose.
+				Bitset flags;
+				if (ir.ids[id].get_type() == TypeVariable && has_decoration(type.self, DecorationBlock))
+					flags = get_buffer_block_flags(id);
+				else
+					flags = get_decoration_bitset(id);
+
+				return flags.get(DecorationNonWritable) ? "const device" : "device";
+			}
 			else
 				return "constant";
 		}
@@ -5435,10 +5500,9 @@ string CompilerMSL::get_type_address_space(const SPIRType &type)
 	return "thread";
 }
 
-// Returns a string containing a comma-delimited list of args for the entry point function
-string CompilerMSL::entry_point_args(bool append_comma)
+string CompilerMSL::entry_point_arg_stage_in()
 {
-	string ep_args;
+	string decl;
 
 	// Stage-in structure
 	uint32_t stage_in_id;
@@ -5452,19 +5516,137 @@ string CompilerMSL::entry_point_args(bool append_comma)
 		auto &var = get<SPIRVariable>(stage_in_id);
 		auto &type = get_variable_data_type(var);
 
+		add_resource_name(var.self);
+		decl = join(type_to_glsl(type), " ", to_name(var.self), " [[stage_in]]");
+	}
+
+	return decl;
+}
+
+void CompilerMSL::entry_point_args_builtin(string &ep_args)
+{
+	// Builtin variables
+	ir.for_each_typed_id<SPIRVariable>([&](uint32_t var_id, SPIRVariable &var) {
+		BuiltIn bi_type = ir.meta[var_id].decoration.builtin_type;
+
+		// Don't emit SamplePosition as a separate parameter. In the entry
+		// point, we get that by calling get_sample_position() on the sample ID.
+		if (var.storage == StorageClassInput && is_builtin_variable(var) &&
+		    get_variable_data_type(var).basetype != SPIRType::Struct &&
+		    get_variable_data_type(var).basetype != SPIRType::ControlPointArray)
+		{
+			if (bi_type != BuiltInSamplePosition && bi_type != BuiltInHelperInvocation &&
+			    bi_type != BuiltInPatchVertices && bi_type != BuiltInTessLevelInner &&
+			    bi_type != BuiltInTessLevelOuter && bi_type != BuiltInPosition && bi_type != BuiltInPointSize &&
+			    bi_type != BuiltInClipDistance && bi_type != BuiltInCullDistance)
+			{
+				if (!ep_args.empty())
+					ep_args += ", ";
+
+				ep_args += builtin_type_decl(bi_type) + " " + to_expression(var_id);
+				ep_args += " [[" + builtin_qualifier(bi_type) + "]]";
+			}
+		}
+	});
+
+	// Vertex and instance index built-ins
+	if (needs_vertex_idx_arg)
+		ep_args += built_in_func_arg(BuiltInVertexIndex, !ep_args.empty());
+
+	if (needs_instance_idx_arg)
+		ep_args += built_in_func_arg(BuiltInInstanceIndex, !ep_args.empty());
+
+	if (capture_output_to_buffer)
+	{
+		// Add parameters to hold the indirect draw parameters and the shader output. This has to be handled
+		// specially because it needs to be a pointer, not a reference.
+		if (stage_out_var_id)
+		{
+			if (!ep_args.empty())
+				ep_args += ", ";
+			ep_args += join("device ", type_to_glsl(get_stage_out_struct_type()), "* ", output_buffer_var_name,
+			                " [[buffer(", msl_options.shader_output_buffer_index, ")]]");
+		}
+
+		if (stage_out_var_id || get_execution_model() == ExecutionModelTessellationControl)
+		{
+			if (!ep_args.empty())
+				ep_args += ", ";
+			ep_args +=
+			    join("device uint* spvIndirectParams [[buffer(", msl_options.indirect_params_buffer_index, ")]]");
+		}
+
+		// Tessellation control shaders get three additional parameters:
+		// a buffer to hold the per-patch data, a buffer to hold the per-patch
+		// tessellation levels, and a block of workgroup memory to hold the
+		// input control point data.
+		if (get_execution_model() == ExecutionModelTessellationControl)
+		{
+			if (patch_stage_out_var_id)
+			{
+				if (!ep_args.empty())
+					ep_args += ", ";
+				ep_args +=
+				    join("device ", type_to_glsl(get_patch_stage_out_struct_type()), "* ", patch_output_buffer_var_name,
+				         " [[buffer(", convert_to_string(msl_options.shader_patch_output_buffer_index), ")]]");
+			}
+			if (!ep_args.empty())
+				ep_args += ", ";
+			ep_args += join("device ", get_tess_factor_struct_name(), "* ", tess_factor_buffer_var_name, " [[buffer(",
+			                convert_to_string(msl_options.shader_tess_factor_buffer_index), ")]]");
+			if (stage_in_var_id)
+			{
+				if (!ep_args.empty())
+					ep_args += ", ";
+				ep_args += join("threadgroup ", type_to_glsl(get_stage_in_struct_type()), "* ", input_wg_var_name,
+				                " [[threadgroup(", convert_to_string(msl_options.shader_input_wg_index), ")]]");
+			}
+		}
+	}
+}
+
+string CompilerMSL::entry_point_args_argument_buffer(bool append_comma)
+{
+	string ep_args = entry_point_arg_stage_in();
+
+	for (uint32_t i = 0; i < kMaxArgumentBuffers; i++)
+	{
+		uint32_t id = argument_buffer_ids[i];
+		if (id == 0)
+			continue;
+
+		add_resource_name(id);
+		auto &var = get<SPIRVariable>(id);
+		auto &type = get_variable_data_type(var);
+
 		if (!ep_args.empty())
 			ep_args += ", ";
 
-		add_resource_name(var.self);
-		ep_args += join(type_to_glsl(type), " ", to_name(var.self), " [[stage_in]]");
+		ep_args += get_argument_address_space(var) + " " + type_to_glsl(type) + "& " + to_name(id);
+		ep_args += " [[buffer(" + convert_to_string(i) + ")]]";
+
+		// Makes it more practical for testing, since the push constant block can occupy the first available
+		// buffer slot if it's not bound explicitly.
+		next_metal_resource_index_buffer = i + 1;
 	}
 
+	entry_point_args_discrete_descriptors(ep_args);
+	entry_point_args_builtin(ep_args);
+
+	if (!ep_args.empty() && append_comma)
+		ep_args += ", ";
+
+	return ep_args;
+}
+
+void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args)
+{
 	// Output resources, sorted by resource index & type
 	// We need to sort to work around a bug on macOS 10.13 with NVidia drivers where switching between shaders
 	// with different order of buffers can result in issues with buffer assignments inside the driver.
 	struct Resource
 	{
-		Variant *id;
+		SPIRVariable *var;
 		string name;
 		SPIRType::BaseType basetype;
 		uint32_t index;
@@ -5472,25 +5654,30 @@ string CompilerMSL::entry_point_args(bool append_comma)
 
 	vector<Resource> resources;
 
-	ir.for_each_typed_id<SPIRVariable>([&](uint32_t self, SPIRVariable &var) {
-		auto &id = ir.ids[self];
-		auto &type = get_variable_data_type(var);
-
-		uint32_t var_id = var.self;
-
+	ir.for_each_typed_id<SPIRVariable>([&](uint32_t, SPIRVariable &var) {
 		if ((var.storage == StorageClassUniform || var.storage == StorageClassUniformConstant ||
 		     var.storage == StorageClassPushConstant || var.storage == StorageClassStorageBuffer) &&
 		    !is_hidden_variable(var))
 		{
+			auto &type = get_variable_data_type(var);
+			uint32_t var_id = var.self;
+
+			if (var.storage != StorageClassPushConstant)
+			{
+				uint32_t desc_set = get_decoration(var_id, DecorationDescriptorSet);
+				if (descriptor_set_is_argument_buffer(desc_set))
+					return;
+			}
+
 			if (type.basetype == SPIRType::SampledImage)
 			{
 				add_resource_name(var_id);
 				resources.push_back(
-				    { &id, to_name(var_id), SPIRType::Image, get_metal_resource_index(var, SPIRType::Image) });
+				    { &var, to_name(var_id), SPIRType::Image, get_metal_resource_index(var, SPIRType::Image) });
 
 				if (type.image.dim != DimBuffer && constexpr_samplers.count(var_id) == 0)
 				{
-					resources.push_back({ &id, to_sampler_expression(var_id), SPIRType::Sampler,
+					resources.push_back({ &var, to_sampler_expression(var_id), SPIRType::Sampler,
 					                      get_metal_resource_index(var, SPIRType::Sampler) });
 				}
 			}
@@ -5499,18 +5686,18 @@ string CompilerMSL::entry_point_args(bool append_comma)
 				// constexpr samplers are not declared as resources.
 				add_resource_name(var_id);
 				resources.push_back(
-				    { &id, to_name(var_id), type.basetype, get_metal_resource_index(var, type.basetype) });
+				    { &var, to_name(var_id), type.basetype, get_metal_resource_index(var, type.basetype) });
 			}
 		}
 	});
 
-	std::sort(resources.begin(), resources.end(), [](const Resource &lhs, const Resource &rhs) {
+	sort(resources.begin(), resources.end(), [](const Resource &lhs, const Resource &rhs) {
 		return tie(lhs.basetype, lhs.index) < tie(rhs.basetype, rhs.index);
 	});
 
 	for (auto &r : resources)
 	{
-		auto &var = r.id->get<SPIRVariable>();
+		auto &var = *r.var;
 		auto &type = get_variable_data_type(var);
 
 		uint32_t var_id = var.self;
@@ -5571,85 +5758,15 @@ string CompilerMSL::entry_point_args(bool append_comma)
 			break;
 		}
 	}
+}
 
-	// Builtin variables
-	ir.for_each_typed_id<SPIRVariable>([&](uint32_t var_id, SPIRVariable &var) {
-		BuiltIn bi_type = ir.meta[var_id].decoration.builtin_type;
-
-		// Don't emit SamplePosition as a separate parameter. In the entry
-		// point, we get that by calling get_sample_position() on the sample ID.
-		if (var.storage == StorageClassInput && is_builtin_variable(var) &&
-		    get_variable_data_type(var).basetype != SPIRType::Struct &&
-		    get_variable_data_type(var).basetype != SPIRType::ControlPointArray)
-		{
-			if (bi_type != BuiltInSamplePosition && bi_type != BuiltInHelperInvocation &&
-			    bi_type != BuiltInPatchVertices && bi_type != BuiltInTessLevelInner &&
-			    bi_type != BuiltInTessLevelOuter && bi_type != BuiltInPosition && bi_type != BuiltInPointSize &&
-			    bi_type != BuiltInClipDistance && bi_type != BuiltInCullDistance)
-			{
-				if (!ep_args.empty())
-					ep_args += ", ";
-
-				ep_args += builtin_type_decl(bi_type) + " " + to_expression(var_id);
-				ep_args += " [[" + builtin_qualifier(bi_type) + "]]";
-			}
-		}
-	});
-
-	// Vertex and instance index built-ins
-	if (needs_vertex_idx_arg)
-		ep_args += built_in_func_arg(BuiltInVertexIndex, !ep_args.empty());
-
-	if (needs_instance_idx_arg)
-		ep_args += built_in_func_arg(BuiltInInstanceIndex, !ep_args.empty());
-
-	if (capture_output_to_buffer)
-	{
-		// Add parameters to hold the indirect draw parameters and the shader output. This has to be handled
-		// specially because it needs to be a pointer, not a reference.
-		if (stage_out_var_id)
-		{
-			if (!ep_args.empty())
-				ep_args += ", ";
-			ep_args += join("device ", type_to_glsl(get_stage_out_struct_type()), "* ", output_buffer_var_name,
-			                " [[buffer(", msl_options.shader_output_buffer_index, ")]]");
-		}
-
-		if (stage_out_var_id || get_execution_model() == ExecutionModelTessellationControl)
-		{
-			if (!ep_args.empty())
-				ep_args += ", ";
-			ep_args +=
-			    join("device uint* spvIndirectParams [[buffer(", msl_options.indirect_params_buffer_index, ")]]");
-		}
-
-		// Tessellation control shaders get three additional parameters:
-		// a buffer to hold the per-patch data, a buffer to hold the per-patch
-		// tessellation levels, and a block of workgroup memory to hold the
-		// input control point data.
-		if (get_execution_model() == ExecutionModelTessellationControl)
-		{
-			if (patch_stage_out_var_id)
-			{
-				if (!ep_args.empty())
-					ep_args += ", ";
-				ep_args +=
-				    join("device ", type_to_glsl(get_patch_stage_out_struct_type()), "* ", patch_output_buffer_var_name,
-				         " [[buffer(", convert_to_string(msl_options.shader_patch_output_buffer_index), ")]]");
-			}
-			if (!ep_args.empty())
-				ep_args += ", ";
-			ep_args += join("device ", get_tess_factor_struct_name(), "* ", tess_factor_buffer_var_name, " [[buffer(",
-			                convert_to_string(msl_options.shader_tess_factor_buffer_index), ")]]");
-			if (stage_in_var_id)
-			{
-				if (!ep_args.empty())
-					ep_args += ", ";
-				ep_args += join("threadgroup ", type_to_glsl(get_stage_in_struct_type()), "* ", input_wg_var_name,
-				                " [[threadgroup(", convert_to_string(msl_options.shader_input_wg_index), ")]]");
-			}
-		}
-	}
+// Returns a string containing a comma-delimited list of args for the entry point function
+// This is the "classic" method of MSL 1 when we don't have argument buffer support.
+string CompilerMSL::entry_point_args_classic(bool append_comma)
+{
+	string ep_args = entry_point_arg_stage_in();
+	entry_point_args_discrete_descriptors(ep_args);
+	entry_point_args_builtin(ep_args);
 
 	if (!ep_args.empty() && append_comma)
 		ep_args += ", ";
@@ -5871,6 +5988,22 @@ string CompilerMSL::argument_decl(const SPIRFunction::Parameter &arg)
 		// Arrays of images and samplers are special cased.
 		if (!address_space.empty())
 			decl = join(address_space, " ", decl);
+
+		if (msl_options.argument_buffers)
+		{
+			// An awkward case where we need to emit *more* address space declarations (yay!).
+			// An example is where we pass down an array of buffer pointers to leaf functions.
+			// It's a constant array containing pointers to constants.
+			// The pointer array is always constant however. E.g.
+			// device SSBO * constant (&array)[N].
+			// const device SSBO * constant (&array)[N].
+			// constant SSBO * constant (&array)[N].
+			// However, this only matters for argument buffers, since for MSL 1.0 style codegen,
+			// we emit the buffer array on stack instead, and that seems to work just fine apparently.
+			if (storage == StorageClassUniform || storage == StorageClassStorageBuffer)
+				decl += " constant";
+		}
+
 		decl += " (&";
 		decl += to_expression(name_id);
 		decl += ")";
@@ -6230,9 +6363,16 @@ string CompilerMSL::to_member_reference(uint32_t base, const SPIRType &type, uin
 	auto *var = maybe_get<SPIRVariable>(base);
 	// If this is a buffer array, we have to dereference the buffer pointers.
 	// Otherwise, if this is a pointer expression, dereference it.
-	if ((var && ((var->storage == StorageClassUniform || var->storage == StorageClassStorageBuffer) &&
-	             is_array(get<SPIRType>(var->basetype)))) ||
-	    (!ptr_chain && should_dereference(base)))
+
+	bool declared_as_pointer = false;
+
+	if (var)
+	{
+		bool is_buffer_variable = var->storage == StorageClassUniform || var->storage == StorageClassStorageBuffer;
+		declared_as_pointer = is_buffer_variable && is_array(get<SPIRType>(var->basetype));
+	}
+
+	if (declared_as_pointer || (!ptr_chain && should_dereference(base)))
 		return join("->", to_member_name(type, index));
 	else
 		return join(".", to_member_name(type, index));
@@ -6259,7 +6399,7 @@ string CompilerMSL::type_to_glsl(const SPIRType &type, uint32_t id)
 	// Pointer?
 	if (type.pointer)
 	{
-		type_name = join(get_type_address_space(type), " ", type_to_glsl(get<SPIRType>(type.parent_type), id));
+		type_name = join(get_type_address_space(type, id), " ", type_to_glsl(get<SPIRType>(type.parent_type), id));
 		switch (type.basetype)
 		{
 		case SPIRType::Image:
@@ -7391,3 +7531,174 @@ std::string CompilerMSL::to_initializer_expression(const SPIRVariable &var)
 	else
 		return CompilerGLSL::to_initializer_expression(var);
 }
+
+bool CompilerMSL::descriptor_set_is_argument_buffer(uint32_t desc_set) const
+{
+	if (!msl_options.argument_buffers)
+		return false;
+	if (desc_set >= kMaxArgumentBuffers)
+		return false;
+
+	return (argument_buffer_discrete_mask & (1u << desc_set)) == 0;
+}
+
+void CompilerMSL::analyze_argument_buffers()
+{
+	// Gather all used resources and sort them out into argument buffers.
+	// Each argument buffer corresponds to a descriptor set in SPIR-V.
+	// The [[id(N)]] values used correspond to the resource mapping we have for MSL.
+	// Otherwise, the binding number is used, but this is generally not safe some types like
+	// combined image samplers and arrays of resources. Metal needs different indices here,
+	// while SPIR-V can have one descriptor set binding. To use argument buffers in practice,
+	// you will need to use the remapping from the API.
+	for (auto &id : argument_buffer_ids)
+		id = 0;
+
+	// Output resources, sorted by resource index & type.
+	struct Resource
+	{
+		SPIRVariable *var;
+		string name;
+		SPIRType::BaseType basetype;
+		uint32_t index;
+	};
+	vector<Resource> resources_in_set[kMaxArgumentBuffers];
+
+	ir.for_each_typed_id<SPIRVariable>([&](uint32_t self, SPIRVariable &var) {
+		if ((var.storage == StorageClassUniform || var.storage == StorageClassUniformConstant ||
+		     var.storage == StorageClassStorageBuffer) &&
+		    !is_hidden_variable(var))
+		{
+			uint32_t desc_set = get_decoration(self, DecorationDescriptorSet);
+			// Ignore if it's part of a push descriptor set.
+			if (!descriptor_set_is_argument_buffer(desc_set))
+				return;
+
+			uint32_t var_id = var.self;
+			auto &type = get_variable_data_type(var);
+
+			if (desc_set >= kMaxArgumentBuffers)
+				SPIRV_CROSS_THROW("Descriptor set index is out of range.");
+
+			if (type.basetype == SPIRType::SampledImage)
+			{
+				add_resource_name(var_id);
+
+				uint32_t image_resource_index = get_metal_resource_index(var, SPIRType::Image);
+				uint32_t sampler_resource_index = get_metal_resource_index(var, SPIRType::Sampler);
+
+				// Avoid trivial conflicts where we didn't remap.
+				// This will let us at least compile test cases without having to instrument remaps.
+				if (sampler_resource_index == image_resource_index)
+					sampler_resource_index += type.array.empty() ? 1 : to_array_size_literal(type);
+
+				resources_in_set[desc_set].push_back({ &var, to_name(var_id), SPIRType::Image, image_resource_index });
+
+				if (type.image.dim != DimBuffer && constexpr_samplers.count(var_id) == 0)
+				{
+					resources_in_set[desc_set].push_back(
+					    { &var, to_sampler_expression(var_id), SPIRType::Sampler, sampler_resource_index });
+				}
+			}
+			else if (constexpr_samplers.count(var_id) == 0)
+			{
+				// constexpr samplers are not declared as resources.
+				add_resource_name(var_id);
+				resources_in_set[desc_set].push_back(
+				    { &var, to_name(var_id), type.basetype, get_metal_resource_index(var, type.basetype) });
+			}
+		}
+	});
+
+	for (uint32_t desc_set = 0; desc_set < kMaxArgumentBuffers; desc_set++)
+	{
+		auto &resources = resources_in_set[desc_set];
+		if (resources.empty())
+			continue;
+
+		assert(descriptor_set_is_argument_buffer(desc_set));
+
+		uint32_t next_id = ir.increase_bound_by(3);
+		uint32_t type_id = next_id + 1;
+		uint32_t ptr_type_id = next_id + 2;
+		argument_buffer_ids[desc_set] = next_id;
+
+		auto &buffer_type = set<SPIRType>(type_id);
+		buffer_type.storage = StorageClassUniform;
+		buffer_type.basetype = SPIRType::Struct;
+		set_name(type_id, join("spvDescriptorSetBuffer", desc_set));
+
+		auto &ptr_type = set<SPIRType>(ptr_type_id);
+		ptr_type = buffer_type;
+		ptr_type.pointer = true;
+		ptr_type.pointer_depth = 1;
+		ptr_type.parent_type = type_id;
+
+		uint32_t buffer_variable_id = next_id;
+		set<SPIRVariable>(buffer_variable_id, ptr_type_id, StorageClassUniform);
+		set_name(buffer_variable_id, join("spvDescriptorSet", desc_set));
+
+		// Ids must be emitted in ID order.
+		sort(begin(resources), end(resources), [&](const Resource &lhs, const Resource &rhs) -> bool {
+			return tie(lhs.index, lhs.basetype) < tie(rhs.index, rhs.basetype);
+		});
+
+		uint32_t member_index = 0;
+		for (auto &resource : resources)
+		{
+			auto &var = *resource.var;
+			auto &type = get_variable_data_type(var);
+			string mbr_name = ensure_valid_name(resource.name, "m");
+			set_member_name(buffer_type.self, member_index, mbr_name);
+
+			if (resource.basetype == SPIRType::Sampler && type.basetype != SPIRType::Sampler)
+			{
+				// Have to synthesize a sampler type here.
+
+				bool type_is_array = !type.array.empty();
+				uint32_t sampler_type_id = ir.increase_bound_by(type_is_array ? 2 : 1);
+				auto &new_sampler_type = set<SPIRType>(sampler_type_id);
+				new_sampler_type.basetype = SPIRType::Sampler;
+				new_sampler_type.storage = StorageClassUniformConstant;
+
+				if (type_is_array)
+				{
+					uint32_t sampler_type_array_id = sampler_type_id + 1;
+					auto &sampler_type_array = set<SPIRType>(sampler_type_array_id);
+					sampler_type_array = new_sampler_type;
+					sampler_type_array.array = type.array;
+					sampler_type_array.array_size_literal = type.array_size_literal;
+					sampler_type_array.parent_type = sampler_type_id;
+					buffer_type.member_types.push_back(sampler_type_array_id);
+				}
+				else
+					buffer_type.member_types.push_back(sampler_type_id);
+			}
+			else
+			{
+				if (resource.basetype == SPIRType::Image || resource.basetype == SPIRType::Sampler ||
+				    resource.basetype == SPIRType::SampledImage)
+				{
+					// Drop pointer information when we emit the resources into a struct.
+					buffer_type.member_types.push_back(get_variable_data_type_id(var));
+					set_qualified_name(var.self, join(to_name(buffer_variable_id), ".", mbr_name));
+				}
+				else
+				{
+					// Resources will be declared as pointers not references, so automatically dereference as appropriate.
+					buffer_type.member_types.push_back(var.basetype);
+					if (type.array.empty())
+						set_qualified_name(var.self, join("(*", to_name(buffer_variable_id), ".", mbr_name, ")"));
+					else
+						set_qualified_name(var.self, join(to_name(buffer_variable_id), ".", mbr_name));
+				}
+			}
+
+			set_extended_member_decoration(buffer_type.self, member_index, SPIRVCrossDecorationArgumentBufferID,
+			                               resource.index);
+			set_extended_member_decoration(buffer_type.self, member_index, SPIRVCrossDecorationInterfaceOrigID,
+			                               var.self);
+			member_index++;
+		}
+	}
+}

3rdparty/spirv-cross/spirv_msl.hpp

@@ -54,6 +54,11 @@ struct MSLVertexAttr
 // Matches the binding index of a MSL resource for a binding within a descriptor set.
 // Taken together, the stage, desc_set and binding combine to form a reference to a resource
 // descriptor used in a particular shading stage.
+// If using MSL 2.0 argument buffers, and the descriptor set is not marked as a discrete descriptor set,
+// the binding reference we remap to will become an [[id(N)]] attribute within
+// the "descriptor set" argument buffer structure.
+// For resources which are bound in the "classic" MSL 1.0 way or discrete descriptors, the remap will become a
+// [[buffer(N)]], [[texture(N)]] or [[sampler(N)]] depending on the resource types used.
 struct MSLResourceBinding
 {
 	spv::ExecutionModel stage = spv::ExecutionModelMax;
@@ -148,6 +153,8 @@ static const uint32_t kPushConstDescSet = ~(0u);
 // element to indicate the bindings for the push constants.
 static const uint32_t kPushConstBinding = 0;
 
+static const uint32_t kMaxArgumentBuffers = 8;
+
 // The current version of the aux buffer structure. It must be incremented any time a
 // new field is added to the aux buffer.
 #define SPIRV_CROSS_MSL_AUX_BUFFER_STRUCT_VERSION 1
@@ -180,6 +187,10 @@ public:
 		bool swizzle_texture_samples = false;
 		bool tess_domain_origin_lower_left = false;
 
+		// Enable use of MSL 2.0 indirect argument buffers.
+		// MSL 2.0 must also be enabled.
+		bool argument_buffers = false;
+
 		// Fragment output in MSL must have at least as many components as the render pass.
 		// Add support to explicit pad out components.
 		bool pad_fragment_output_components = false;
@@ -275,6 +286,10 @@ public:
 	// the set/binding combination was used by the MSL code.
 	void add_msl_resource_binding(const MSLResourceBinding &resource);
 
+	// When using MSL argument buffers, we can force "classic" MSL 1.0 binding schemes for certain descriptor sets.
+	// This corresponds to VK_KHR_push_descriptor in Vulkan.
+	void add_discrete_descriptor_set(uint32_t desc_set);
+
 	// Query after compilation is done. This allows you to check if a location or set/binding combination was used by the shader.
 	bool is_msl_vertex_attribute_used(uint32_t location);
 	bool is_msl_resource_binding_used(spv::ExecutionModel model, uint32_t set, uint32_t binding);
@@ -413,7 +428,11 @@ protected:
 	void fix_up_shader_inputs_outputs();
 
 	std::string func_type_decl(SPIRType &type);
-	std::string entry_point_args(bool append_comma);
+	std::string entry_point_args_classic(bool append_comma);
+	std::string entry_point_args_argument_buffer(bool append_comma);
+	std::string entry_point_arg_stage_in();
+	void entry_point_args_builtin(std::string &args);
+	void entry_point_args_discrete_descriptors(std::string &args);
 	std::string to_qualified_member_name(const SPIRType &type, uint32_t index);
 	std::string ensure_valid_name(std::string name, std::string pfx);
 	std::string to_sampler_expression(uint32_t id);
@@ -432,7 +451,7 @@ protected:
 	bool is_member_packable(SPIRType &ib_type, uint32_t index);
 	MSLStructMemberKey get_struct_member_key(uint32_t type_id, uint32_t index);
 	std::string get_argument_address_space(const SPIRVariable &argument);
-	std::string get_type_address_space(const SPIRType &type);
+	std::string get_type_address_space(const SPIRType &type, uint32_t id);
 	SPIRType &get_stage_in_struct_type();
 	SPIRType &get_stage_out_struct_type();
 	SPIRType &get_patch_stage_in_struct_type();
@@ -513,6 +532,11 @@ protected:
 	std::unordered_map<uint32_t, MSLConstexprSampler> constexpr_samplers;
 	std::vector<uint32_t> buffer_arrays;
 
+	uint32_t argument_buffer_ids[kMaxArgumentBuffers];
+	uint32_t argument_buffer_discrete_mask = 0;
+	void analyze_argument_buffers();
+	bool descriptor_set_is_argument_buffer(uint32_t desc_set) const;
+
 	uint32_t get_target_components_for_fragment_location(uint32_t location) const;
 	uint32_t build_extended_vector_type(uint32_t type_id, uint32_t components);
 

3rdparty/spirv-cross/test_shaders.py

@@ -169,6 +169,14 @@ def cross_compile_msl(shader, spirv, opt, paths):
         msl_args.append('--msl-capture-output')
     if '.domain.' in shader:
         msl_args.append('--msl-domain-lower-left')
+    if '.argument.' in shader:
+        msl_args.append('--msl-argument-buffers')
+    if '.discrete.' in shader:
+        # Arbitrary for testing purposes.
+        msl_args.append('--msl-discrete-descriptor-set')
+        msl_args.append('2')
+        msl_args.append('--msl-discrete-descriptor-set')
+        msl_args.append('3')
 
     subprocess.check_call(msl_args)