Merge branch 'master' of git://github.com/brenwill/SPIRV-Cross into pr-109

reference/shaders-msl/comp/defer-parens.comp

@@ -0,0 +1,23 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct SSBO
+{
+    float4 data;
+    int index;
+};
+
+kernel void main0(device SSBO& _13 [[buffer(0)]])
+{
+    float4 d = _13.data;
+    _13.data = float4(d.x, d.yz + float2(10.0), d.w);
+    _13.data = (d + d) + d;
+    _13.data = (d.yz + float2(10.0)).xxyy;
+    float t = (d.yz + float2(10.0)).y;
+    _13.data = float4(t);
+    t = (d.zw + float2(10.0))[_13.index];
+    _13.data = float4(t);
+}
+

reference/shaders-msl/comp/read-write-only.comp

@@ -0,0 +1,29 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct SSBO2
+{
+    float4 data4;
+    float4 data5;
+};
+
+struct SSBO0
+{
+    float4 data0;
+    float4 data1;
+};
+
+struct SSBO1
+{
+    float4 data2;
+    float4 data3;
+};
+
+kernel void main0(device SSBO2& _10 [[buffer(0)]], device SSBO0& _15 [[buffer(1)]], device SSBO1& _21 [[buffer(2)]])
+{
+    _10.data4 = _15.data0 + _21.data2;
+    _10.data5 = _15.data1 + _21.data3;
+}
+

reference/shaders-msl/flatten/basic.flatten.vert

@@ -0,0 +1,31 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct UBO
+{
+    float4x4 uMVP;
+};
+
+struct main0_in
+{
+    float4 aVertex [[attribute(0)]];
+    float3 aNormal [[attribute(1)]];
+};
+
+struct main0_out
+{
+    float3 vNormal [[user(locn0)]];
+    float4 gl_Position [[position]];
+    float gl_PointSize;
+};
+
+vertex main0_out main0(main0_in in [[stage_in]], constant UBO& _16 [[buffer(0)]])
+{
+    main0_out out = {};
+    out.gl_Position = _16.uMVP * in.aVertex;
+    out.vNormal = in.aNormal;
+    return out;
+}
+

reference/shaders-msl/flatten/multiindex.flatten.vert

@@ -0,0 +1,28 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct UBO
+{
+    float4 Data[3][5];
+};
+
+struct main0_in
+{
+    int2 aIndex [[attribute(0)]];
+};
+
+struct main0_out
+{
+    float4 gl_Position [[position]];
+    float gl_PointSize;
+};
+
+vertex main0_out main0(main0_in in [[stage_in]], constant UBO& _20 [[buffer(0)]])
+{
+    main0_out out = {};
+    out.gl_Position = _20.Data[in.aIndex.x][in.aIndex.y];
+    return out;
+}
+

reference/shaders-msl/flatten/push-constant.flatten.vert

@@ -0,0 +1,33 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct PushMe
+{
+    float4x4 MVP;
+    float2x2 Rot;
+    float Arr[4];
+};
+
+struct main0_in
+{
+    float4 Pos [[attribute(1)]];
+    float2 Rot [[attribute(0)]];
+};
+
+struct main0_out
+{
+    float2 vRot [[user(locn0)]];
+    float4 gl_Position [[position]];
+    float gl_PointSize;
+};
+
+vertex main0_out main0(main0_in in [[stage_in]], constant PushMe& registers [[buffer(0)]])
+{
+    main0_out out = {};
+    out.gl_Position = registers.MVP * in.Pos;
+    out.vRot = (registers.Rot * in.Rot) + float2(registers.Arr[2]);
+    return out;
+}
+

reference/shaders-msl/flatten/struct.flatten.vert

@@ -0,0 +1,41 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct Light
+{
+    float3 Position;
+    float Radius;
+    float4 Color;
+};
+
+struct UBO
+{
+    float4x4 uMVP;
+    Light light;
+};
+
+struct main0_in
+{
+    float4 aVertex [[attribute(0)]];
+    float3 aNormal [[attribute(1)]];
+};
+
+struct main0_out
+{
+    float4 vColor [[user(locn0)]];
+    float4 gl_Position [[position]];
+    float gl_PointSize;
+};
+
+vertex main0_out main0(main0_in in [[stage_in]], constant UBO& _18 [[buffer(0)]])
+{
+    main0_out out = {};
+    out.gl_Position = _18.uMVP * in.aVertex;
+    out.vColor = float4(0.0);
+    float3 L = in.aVertex.xyz - _18.light.Position;
+    out.vColor += ((_18.light.Color * clamp(1.0 - (length(L) / _18.light.Radius), 0.0, 1.0)) * dot(in.aNormal, normalize(L)));
+    return out;
+}
+

reference/shaders-msl/flatten/swizzle.flatten.vert

@@ -0,0 +1,48 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct UBO
+{
+    float4 A;
+    float2 B0;
+    float2 B1;
+    float C0;
+    float3 C1;
+    float3 D0;
+    float D1;
+    float E0;
+    float E1;
+    float E2;
+    float E3;
+    float F0;
+    float2 F1;
+    float F2;
+};
+
+struct main0_out
+{
+    float4 oF [[user(locn0)]];
+    float4 oE [[user(locn1)]];
+    float4 oD [[user(locn2)]];
+    float4 oC [[user(locn3)]];
+    float4 oB [[user(locn4)]];
+    float4 oA [[user(locn5)]];
+    float4 gl_Position [[position]];
+    float gl_PointSize;
+};
+
+vertex main0_out main0(constant UBO& _22 [[buffer(0)]])
+{
+    main0_out out = {};
+    out.gl_Position = float4(0.0);
+    out.oA = _22.A;
+    out.oB = float4(_22.B0, _22.B1);
+    out.oC = float4(_22.C0, _22.C1);
+    out.oD = float4(_22.D0, _22.D1);
+    out.oE = float4(_22.E0, _22.E1, _22.E2, _22.E3);
+    out.oF = float4(_22.F0, _22.F1, _22.F2);
+    return out;
+}
+

reference/shaders-msl/flatten/types.flatten.frag

@@ -0,0 +1,35 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct UBO1
+{
+    int4 c;
+    int4 d;
+};
+
+struct UBO2
+{
+    uint4 e;
+    uint4 f;
+};
+
+struct UBO0
+{
+    float4 a;
+    float4 b;
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0(constant UBO1& _14 [[buffer(0)]], constant UBO2& _29 [[buffer(1)]], constant UBO0& _41 [[buffer(2)]])
+{
+    main0_out out = {};
+    out.FragColor = ((((float4(_14.c) + float4(_14.d)) + float4(_29.e)) + float4(_29.f)) + _41.a) + _41.b;
+    return out;
+}
+

reference/shaders-msl/frag/basic.frag

@@ -0,0 +1,23 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_in
+{
+    float2 vTex [[user(locn0)]];
+    float4 vColor [[user(locn1)]];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]], texture2d<float> uTex [[texture(0)]], sampler uTexSmplr [[sampler(0)]])
+{
+    main0_out out = {};
+    out.FragColor = in.vColor * uTex.sample(uTexSmplr, in.vTex.xy);
+    return out;
+}
+

reference/shaders-msl/frag/composite-extract-forced-temporary.frag

@@ -0,0 +1,23 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_in
+{
+    float2 vTexCoord [[user(locn0)]];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]], texture2d<float> Texture [[texture(0)]], sampler TextureSmplr [[sampler(0)]])
+{
+    main0_out out = {};
+    float f = Texture.sample(TextureSmplr, in.vTexCoord.xy).x;
+    out.FragColor = float4(f * f);
+    return out;
+}
+

reference/shaders-msl/frag/constant-array.frag

@@ -0,0 +1,40 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct Foobar
+{
+    float a;
+    float b;
+};
+
+struct main0_in
+{
+    int index [[user(locn0)]];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+float4 resolve(thread const Foobar& f)
+{
+    return float4(f.a + f.b);
+}
+
+fragment main0_out main0(main0_in in [[stage_in]])
+{
+    main0_out out = {};
+    float4 indexable[3] = {float4(1.0), float4(2.0), float4(3.0)};
+    float4 indexable_1[2][2] = {{float4(1.0), float4(2.0)}, {float4(8.0), float4(10.0)}};
+    Foobar param = {10.0, 20.0};
+    Foobar indexable_2[2] = {{10.0, 40.0}, {90.0, 70.0}};
+    Foobar param_1 = indexable_2[in.index];
+    out.FragColor = ((indexable[in.index] + (indexable_1[in.index][in.index + 1])) + resolve(param)) + resolve(param_1);
+    return out;
+}
+

reference/shaders-msl/frag/flush_params.frag

@@ -0,0 +1,38 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct Structy
+{
+    float4 c;
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+void foo2(thread Structy& f)
+{
+    f.c = float4(10.0);
+}
+
+Structy foo()
+{
+    Structy param;
+    foo2(param);
+    Structy f = param;
+    return f;
+}
+
+fragment main0_out main0()
+{
+    main0_out out = {};
+    Structy s = foo();
+    out.FragColor = s.c;
+    return out;
+}
+

reference/shaders-msl/frag/mix.frag

@@ -0,0 +1,48 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_in
+{
+    float vIn3 [[user(locn3)]];
+    float vIn2 [[user(locn2)]];
+    float4 vIn1 [[user(locn1)]];
+    float4 vIn0 [[user(locn0)]];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]])
+{
+    main0_out out = {};
+    bool4 l = bool4(false, true, false, false);
+    out.FragColor = float4(l.x ? in.vIn1.x : in.vIn0.x, l.y ? in.vIn1.y : in.vIn0.y, l.z ? in.vIn1.z : in.vIn0.z, l.w ? in.vIn1.w : in.vIn0.w);
+    bool f = true;
+    out.FragColor = float4(f ? in.vIn3 : in.vIn2);
+    float4 _35;
+    if (f)
+    {
+        _35 = in.vIn0;
+    }
+    else
+    {
+        _35 = in.vIn1;
+    }
+    out.FragColor = _35;
+    float _44;
+    if (f)
+    {
+        _44 = in.vIn2;
+    }
+    else
+    {
+        _44 = in.vIn3;
+    }
+    out.FragColor = float4(_44);
+    return out;
+}
+

reference/shaders-msl/frag/pls.frag

@@ -0,0 +1,31 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_in
+{
+    float4 PLSIn3 [[user(locn0)]];
+    float4 PLSIn2 [[user(locn1)]];
+    float4 PLSIn1 [[user(locn1)]];
+    float4 PLSIn0 [[user(locn0)]];
+};
+
+struct main0_out
+{
+    float4 PLSOut0 [[color(0)]];
+    float4 PLSOut1 [[color(1)]];
+    float4 PLSOut2 [[color(2)]];
+    float4 PLSOut3 [[color(3)]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]])
+{
+    main0_out out = {};
+    out.PLSOut0 = in.PLSIn0 * 2.0;
+    out.PLSOut1 = in.PLSIn1 * 6.0;
+    out.PLSOut2 = in.PLSIn2 * 7.0;
+    out.PLSOut3 = in.PLSIn3 * 4.0;
+    return out;
+}
+

reference/shaders-msl/frag/sampler-ms.frag

@@ -0,0 +1,18 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0(float4 gl_FragCoord [[position]], texture2d_ms<float> uSampler [[texture(0)]], sampler uSamplerSmplr [[sampler(0)]])
+{
+    main0_out out = {};
+    int2 coord = int2(gl_FragCoord.xy);
+    out.FragColor = ((uSampler.read(uint2(coord.xy), 0) + uSampler.read(uint2(coord.xy), 1)) + uSampler.read(uint2(coord.xy), 2)) + uSampler.read(uint2(coord.xy), 3);
+    return out;
+}
+

reference/shaders-msl/frag/sampler.frag

@@ -0,0 +1,31 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_in
+{
+    float2 vTex [[user(locn0)]];
+    float4 vColor [[user(locn1)]];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+float4 sample_texture(thread const texture2d<float> tex, thread const sampler& texSmplr, thread const float2& uv)
+{
+    return tex.sample(texSmplr, uv.xy);
+}
+
+fragment main0_out main0(main0_in in [[stage_in]], texture2d<float> uTex [[texture(0)]], sampler uTexSmplr [[sampler(0)]])
+{
+    main0_out out = {};
+    float2 param = in.vTex;
+    out.FragColor = in.vColor * sample_texture(uTex, uTexSmplr, param);
+    return out;
+}
+

reference/shaders-msl/frag/swizzle.frag

@@ -0,0 +1,28 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_in
+{
+    float2 vUV [[user(locn2)]];
+    float3 vNormal [[user(locn1)]];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]], texture2d<float> samp [[texture(0)]], sampler sampSmplr [[sampler(0)]])
+{
+    main0_out out = {};
+    out.FragColor = float4(samp.sample(sampSmplr, in.vUV.xy).xyz, 1.0);
+    out.FragColor = float4(samp.sample(sampSmplr, in.vUV.xy).xz, 1.0, 4.0);
+    out.FragColor = float4(samp.sample(sampSmplr, in.vUV.xy).xx, samp.sample(sampSmplr, (in.vUV + float2(0.100000001490116119384765625)).xy).yy);
+    out.FragColor = float4(in.vNormal, 1.0);
+    out.FragColor = float4(in.vNormal + float3(1.7999999523162841796875), 1.0);
+    out.FragColor = float4(in.vUV, in.vUV + float2(1.7999999523162841796875));
+    return out;
+}
+

reference/shaders-msl/frag/ubo_layout.frag

@@ -0,0 +1,32 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct Str
+{
+    float4x4 foo;
+};
+
+struct UBO1
+{
+    Str foo;
+};
+
+struct UBO2
+{
+    Str foo;
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0(constant UBO1& ubo1 [[buffer(0)]], constant UBO2& ubo0 [[buffer(1)]])
+{
+    main0_out out = {};
+    out.FragColor = transpose(ubo1.foo.foo)[0] + ubo0.foo.foo[0];
+    return out;
+}
+

reference/shaders-msl/legacy/vert/transpose.legacy.vert

@@ -0,0 +1,34 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct Buffer
+{
+    float4x4 MVPRowMajor;
+    float4x4 MVPColMajor;
+    float4x4 M;
+};
+
+struct main0_in
+{
+    float4 Position [[attribute(0)]];
+};
+
+struct main0_out
+{
+    float4 gl_Position [[position]];
+    float gl_PointSize;
+};
+
+vertex main0_out main0(main0_in in [[stage_in]], constant Buffer& _13 [[buffer(0)]])
+{
+    main0_out out = {};
+    float4 c0 = _13.M * (in.Position * _13.MVPRowMajor);
+    float4 c1 = _13.M * (_13.MVPColMajor * in.Position);
+    float4 c2 = _13.M * (_13.MVPRowMajor * in.Position);
+    float4 c3 = _13.M * (in.Position * _13.MVPColMajor);
+    out.gl_Position = ((c0 + c1) + c2) + c3;
+    return out;
+}
+

reference/shaders-msl/vert/ubo.vert

@@ -0,0 +1,31 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct UBO
+{
+    float4x4 mvp;
+};
+
+struct main0_in
+{
+    float4 aVertex [[attribute(0)]];
+    float3 aNormal [[attribute(1)]];
+};
+
+struct main0_out
+{
+    float3 vNormal [[user(locn0)]];
+    float4 gl_Position [[position]];
+    float gl_PointSize;
+};
+
+vertex main0_out main0(main0_in in [[stage_in]], constant UBO& _16 [[buffer(0)]])
+{
+    main0_out out = {};
+    out.gl_Position = _16.mvp * in.aVertex;
+    out.vNormal = in.aNormal;
+    return out;
+}
+

reference/shaders-msl/vulkan/frag/push-constant.vk.frag

@@ -0,0 +1,28 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct PushConstants
+{
+    float4 value0;
+    float4 value1;
+};
+
+struct main0_in
+{
+    float4 vColor [[user(locn0)]];
+};
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]], constant PushConstants& push [[buffer(0)]])
+{
+    main0_out out = {};
+    out.FragColor = (in.vColor + push.value0) + push.value1;
+    return out;
+}
+

reference/shaders-msl/vulkan/vert/vulkan-vertex.vk.vert

@@ -0,0 +1,18 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_out
+{
+    float4 gl_Position [[position]];
+    float gl_PointSize;
+};
+
+vertex main0_out main0(uint gl_VertexIndex [[vertex_id]], uint gl_InstanceIndex [[instance_id]])
+{
+    main0_out out = {};
+    out.gl_Position = float4(1.0, 2.0, 3.0, 4.0) * float(gl_VertexIndex + gl_InstanceIndex);
+    return out;
+}
+

shaders-msl/comp/defer-parens.comp

@@ -0,0 +1,30 @@
+#version 310 es
+layout(local_size_x = 1) in;
+
+layout(binding = 0, std430) buffer SSBO
+{
+	vec4 data;
+	int index;
+};
+
+void main()
+{
+	// Tests defer-parens behavior where a binary expression is OpCompositeExtracted chained together
+	// with an OpCompositeConstruct optimization.
+	vec4 d = data;
+	data = vec4(d.x, d.yz + 10.0, d.w);
+
+	// Verify binary ops.
+	data = d + d + d;
+
+	// Verify swizzles.
+	data = (d.yz + 10.0).xxyy;
+
+	// OpCompositeExtract
+	float t = (d.yz + 10.0).y;
+	data = vec4(t);
+
+	// OpVectorExtractDynamic
+	t = (d.zw + 10.0)[index];
+	data = vec4(t);
+}

shaders-msl/comp/read-write-only.comp

@@ -0,0 +1,26 @@
+#version 310 es
+layout(local_size_x = 1) in;
+
+layout(binding = 0, std430) readonly buffer SSBO0
+{
+   vec4 data0;
+   vec4 data1;
+};
+
+layout(binding = 1, std430) restrict buffer SSBO1
+{
+   vec4 data2;
+   vec4 data3;
+};
+
+layout(binding = 2, std430) restrict writeonly buffer SSBO2
+{
+   vec4 data4;
+   vec4 data5;
+};
+
+void main()
+{
+   data4 = data0 + data2;
+   data5 = data1 + data3;
+}

shaders-msl/flatten/basic.flatten.vert

@@ -0,0 +1,15 @@
+#version 310 es
+
+layout(std140) uniform UBO
+{
+    mat4 uMVP;
+};
+in vec4 aVertex;
+in vec3 aNormal;
+out vec3 vNormal;
+
+void main()
+{
+    gl_Position = uMVP * aVertex;
+    vNormal = aNormal;
+}

shaders-msl/flatten/multiindex.flatten.vert

@@ -0,0 +1,13 @@
+#version 310 es
+
+layout(std140) uniform UBO
+{
+    vec4 Data[3][5];
+};
+
+in ivec2 aIndex;
+
+void main()
+{
+    gl_Position = Data[aIndex.x][aIndex.y];
+}

shaders-msl/flatten/push-constant.flatten.vert

@@ -0,0 +1,17 @@
+#version 310 es
+
+layout(push_constant, std430) uniform PushMe
+{
+   mat4 MVP;
+   mat2 Rot; // The MatrixStride will be 8 here.
+   float Arr[4];
+} registers;
+
+layout(location = 0) in vec2 Rot;
+layout(location = 1) in vec4 Pos;
+layout(location = 0) out vec2 vRot;
+void main()
+{
+   gl_Position = registers.MVP * Pos;
+   vRot = registers.Rot * Rot + registers.Arr[2]; // Constant access should work even if array stride is just 4 here.
+}

shaders-msl/flatten/struct.flatten.vert

@@ -0,0 +1,30 @@
+#version 310 es
+
+struct Light
+{
+    vec3 Position;
+    float Radius;
+
+    vec4 Color;
+};
+
+layout(std140) uniform UBO
+{
+    mat4 uMVP;
+
+    Light light;
+};
+
+in vec4 aVertex;
+in vec3 aNormal;
+out vec4 vColor;
+
+void main()
+{
+    gl_Position = uMVP * aVertex;
+
+    vColor = vec4(0.0);
+
+    vec3 L = aVertex.xyz - light.Position;
+    vColor += dot(aNormal, normalize(L)) * (clamp(1.0 - length(L) / light.Radius, 0.0, 1.0) * light.Color);
+}

shaders-msl/flatten/swizzle.flatten.vert

@@ -0,0 +1,42 @@
+#version 310 es
+
+// comments note the 16b alignment boundaries (see GL spec 7.6.2.2 Standard Uniform Block Layout)
+layout(std140) uniform UBO
+{
+    // 16b boundary
+    vec4 A;
+    // 16b boundary
+    vec2 B0;
+    vec2 B1;
+    // 16b boundary
+    float C0;
+    // 16b boundary (vec3 is aligned to 16b)
+    vec3 C1;
+    // 16b boundary
+    vec3 D0;
+    float D1;
+    // 16b boundary
+    float E0;
+    float E1;
+    float E2;
+    float E3;
+    // 16b boundary
+    float F0;
+    vec2 F1;
+    // 16b boundary (vec2 before us is aligned to 8b)
+    float F2;
+};
+
+out vec4 oA, oB, oC, oD, oE, oF;
+
+void main()
+{
+    gl_Position = vec4(0.0);
+
+    oA = A;
+    oB = vec4(B0, B1);
+    oC = vec4(C0, C1);
+    oD = vec4(D0, D1);
+    oE = vec4(E0, E1, E2, E3);
+    oF = vec4(F0, F1, F2);
+}

shaders-msl/flatten/types.flatten.frag

@@ -0,0 +1,27 @@
+#version 310 es
+precision mediump float;
+
+layout(std140, binding = 0) uniform UBO0
+{
+   vec4 a;
+   vec4 b;
+};
+
+layout(std140, binding = 0) uniform UBO1
+{
+   ivec4 c;
+   ivec4 d;
+};
+
+layout(std140, binding = 0) uniform UBO2
+{
+   uvec4 e;
+   uvec4 f;
+};
+
+layout(location = 0) out vec4 FragColor;
+
+void main()
+{
+   FragColor = vec4(c) + vec4(d) + vec4(e) + vec4(f) + a + b; 
+}

shaders-msl/frag/basic.frag

@@ -0,0 +1,13 @@
+#version 310 es
+precision mediump float;
+
+in vec4 vColor;
+in vec2 vTex;
+layout(binding = 0) uniform sampler2D uTex;
+layout(location = 0) out vec4 FragColor;
+
+void main()
+{
+    FragColor = vColor * texture(uTex, vTex);
+}
+

shaders-msl/frag/composite-extract-forced-temporary.frag

@@ -0,0 +1,11 @@
+#version 310 es
+precision mediump float;
+layout(binding = 0) uniform sampler2D Texture;
+layout(location = 0) out vec4 FragColor;
+layout(location = 0) in vec2 vTexCoord;
+
+void main()
+{
+	float f = texture(Texture, vTexCoord).x;
+	FragColor = vec4(f * f);
+}

shaders-msl/frag/constant-array.frag

@@ -0,0 +1,21 @@
+#version 310 es
+precision mediump float;
+layout(location = 0) out vec4 FragColor;
+
+layout(location = 0) flat in int index;
+
+struct Foobar { float a; float b; };
+
+vec4 resolve(Foobar f)
+{
+   return vec4(f.a + f.b);
+}
+
+void main()
+{
+   const vec4 foo[3] = vec4[](vec4(1.0), vec4(2.0), vec4(3.0));
+   const vec4 foobars[2][2] = vec4[][](vec4[](vec4(1.0), vec4(2.0)), vec4[](vec4(8.0), vec4(10.0)));
+   const Foobar foos[2] = Foobar[](Foobar(10.0, 40.0), Foobar(90.0, 70.0));
+
+   FragColor = foo[index] + foobars[index][index + 1] + resolve(Foobar(10.0, 20.0)) + resolve(foos[index]);
+}

shaders-msl/frag/flush_params.frag

@@ -0,0 +1,27 @@
+#version 310 es
+precision mediump float;
+
+layout(location = 0) out vec4 FragColor;
+
+struct Structy
+{
+   vec4 c;
+};
+
+void foo2(out Structy f)
+{
+   f.c = vec4(10.0);
+}
+
+Structy foo()
+{
+   Structy f;
+   foo2(f);
+   return f;
+}
+
+void main()
+{
+   Structy s = foo();
+   FragColor = s.c;
+}

shaders-msl/frag/mix.frag

@@ -0,0 +1,20 @@
+#version 310 es
+precision mediump float;
+
+layout(location = 0) in vec4 vIn0;
+layout(location = 1) in vec4 vIn1;
+layout(location = 2) in float vIn2;
+layout(location = 3) in float vIn3;
+layout(location = 0) out vec4 FragColor;
+
+void main()
+{
+    bvec4 l = bvec4(false, true, false, false);
+    FragColor = mix(vIn0, vIn1, l);
+
+    bool f = true;
+    FragColor = vec4(mix(vIn2, vIn3, f));
+
+    FragColor = f ? vIn0 : vIn1;
+    FragColor = vec4(f ? vIn2 : vIn3);
+}

shaders-msl/frag/pls.frag

@@ -0,0 +1,20 @@
+#version 310 es
+precision mediump float;
+
+layout(location = 0) in vec4 PLSIn0;
+layout(location = 1) in vec4 PLSIn1;
+in vec4 PLSIn2;
+in vec4 PLSIn3;
+
+layout(location = 0) out vec4 PLSOut0;
+layout(location = 1) out vec4 PLSOut1;
+layout(location = 2) out vec4 PLSOut2;
+layout(location = 3) out vec4 PLSOut3;
+
+void main()
+{
+    PLSOut0 = 2.0 * PLSIn0;
+    PLSOut1 = 6.0 * PLSIn1;
+    PLSOut2 = 7.0 * PLSIn2;
+    PLSOut3 = 4.0 * PLSIn3;
+}

shaders-msl/frag/sampler-ms.frag

@@ -0,0 +1,16 @@
+#version 310 es
+precision mediump float;
+precision highp int;
+
+layout(binding = 0) uniform mediump sampler2DMS uSampler;
+layout(location = 0) out vec4 FragColor;
+
+void main()
+{
+	ivec2 coord = ivec2(gl_FragCoord.xy);
+	FragColor =
+		texelFetch(uSampler, coord, 0) +
+		texelFetch(uSampler, coord, 1) +
+		texelFetch(uSampler, coord, 2) +
+		texelFetch(uSampler, coord, 3);
+}

shaders-msl/frag/sampler.frag

@@ -0,0 +1,18 @@
+#version 310 es
+precision mediump float;
+
+in vec4 vColor;
+in vec2 vTex;
+layout(binding = 0) uniform sampler2D uTex;
+layout(location = 0) out vec4 FragColor;
+
+vec4 sample_texture(sampler2D tex, vec2 uv)
+{
+    return texture(tex, uv);
+}
+
+void main()
+{
+    FragColor = vColor * sample_texture(uTex, vTex);
+}
+

shaders-msl/frag/swizzle.frag

@@ -0,0 +1,17 @@
+#version 310 es
+precision mediump float;
+
+layout(location = 0) uniform sampler2D samp;
+layout(location = 0) out vec4 FragColor;
+layout(location = 1) in vec3 vNormal;
+layout(location = 2) in vec2 vUV;
+
+void main()
+{
+    FragColor = vec4(texture(samp, vUV).xyz, 1.0);
+    FragColor = vec4(texture(samp, vUV).xz, 1.0, 4.0);
+    FragColor = vec4(texture(samp, vUV).xx, texture(samp, vUV + vec2(0.1)).yy);
+    FragColor = vec4(vNormal, 1.0);
+    FragColor = vec4(vNormal + 1.8, 1.0);
+    FragColor = vec4(vUV, vUV + 1.8);
+}

shaders-msl/frag/ubo_layout.frag

@@ -0,0 +1,24 @@
+#version 310 es
+precision mediump float;
+
+layout(location = 0) out vec4 FragColor;
+
+struct Str
+{
+    mat4 foo;
+};
+
+layout(binding = 0, std140) uniform UBO1
+{
+    layout(row_major) Str foo;
+} ubo1;
+
+layout(binding = 1, std140) uniform UBO2
+{
+    layout(column_major) Str foo;
+} ubo0;
+
+void main()
+{
+    FragColor = ubo1.foo.foo[0] + ubo0.foo.foo[0];
+}

shaders-msl/legacy/vert/transpose.legacy.vert

@@ -0,0 +1,20 @@
+#version 310 es
+
+uniform Buffer
+{
+	layout(row_major) mat4 MVPRowMajor;
+	layout(column_major) mat4 MVPColMajor;
+	mat4 M;
+};
+
+layout(location = 0) in vec4 Position;
+
+void main()
+{
+	vec4 c0 = M * (MVPRowMajor * Position);
+	vec4 c1 = M * (MVPColMajor * Position);
+	vec4 c2 = M * (Position * MVPRowMajor);
+	vec4 c3 = M * (Position * MVPColMajor);
+	gl_Position = c0 + c1 + c2 + c3;
+}
+

shaders-msl/vert/ubo.vert

@@ -0,0 +1,16 @@
+#version 310 es
+
+layout(binding = 0, std140) uniform UBO
+{
+   mat4 mvp; 
+};
+
+in vec4 aVertex;
+in vec3 aNormal;
+out vec3 vNormal;
+
+void main()
+{
+    gl_Position = mvp * aVertex;
+    vNormal = aNormal;
+}

shaders-msl/vulkan/frag/push-constant.vk.frag

@@ -0,0 +1,16 @@
+#version 310 es
+precision mediump float;
+
+layout(push_constant, std430) uniform PushConstants
+{
+   vec4 value0;
+   vec4 value1;
+} push;
+
+layout(location = 0) in vec4 vColor;
+layout(location = 0) out vec4 FragColor;
+
+void main()
+{
+   FragColor = vColor + push.value0 + push.value1;
+}

shaders-msl/vulkan/vert/vulkan-vertex.vk.vert

@@ -0,0 +1,6 @@
+#version 310 es
+
+void main()
+{
+   gl_Position = float(gl_VertexIndex + gl_InstanceIndex) * vec4(1.0, 2.0, 3.0, 4.0);
+}

spirv_cross.cpp

@@ -839,6 +839,10 @@ void Compiler::set_member_decoration(uint32_t id, uint32_t index, Decoration dec
 		dec.location = argument;
 		break;
 
+	case DecorationBinding:
+		dec.binding = argument;
+		break;
+
 	case DecorationOffset:
 		dec.offset = argument;
 		break;
@@ -896,6 +900,8 @@ uint32_t Compiler::get_member_decoration(uint32_t id, uint32_t index, Decoration
 		return dec.builtin_type;
 	case DecorationLocation:
 		return dec.location;
+	case DecorationBinding:
+		return dec.binding;
 	case DecorationOffset:
 		return dec.offset;
 	case DecorationSpecId:

spirv_msl.cpp

@@ -1,5 +1,5 @@
 /*
- * Copyright 2015-2016 The Brenwill Workshop Ltd.
+ * Copyright 2016-2017 The Brenwill Workshop Ltd.
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
  * you may not use this file except in compliance with the License.
@@ -25,6 +25,8 @@ using namespace spv;
 using namespace spirv_cross;
 using namespace std;
 
+static const uint32_t k_unknown_location = ~0;
+
 CompilerMSL::CompilerMSL(vector<uint32_t> spirv_)
     : CompilerGLSL(move(spirv_))
 {
@@ -65,6 +67,9 @@ string CompilerMSL::compile(MSLConfiguration &msl_cfg, vector<MSLVertexAttr> *p_
 		for (auto &va : *p_vtx_attrs)
 			vtx_attrs_by_location[va.location] = &va;
 
+	non_stage_in_input_var_ids.clear();
+	pad_type_ids_by_pad_len.clear();
+
 	resource_bindings.clear();
 	if (p_res_bindings)
 		for (auto &rb : *p_res_bindings)
@@ -183,7 +188,7 @@ void CompilerMSL::extract_global_variables_from_functions()
 	for (auto &var : entry_func.local_variables)
 		global_var_ids.insert(var);
 
-	std::unordered_set<uint32_t> added_arg_ids;
+	std::set<uint32_t> added_arg_ids;
 	std::unordered_set<uint32_t> processed_func_ids;
 	extract_global_variables_from_function(entry_point, added_arg_ids, global_var_ids, processed_func_ids);
 }
@@ -191,7 +196,7 @@ void CompilerMSL::extract_global_variables_from_functions()
 // MSL does not support the use of global variables for shader input content.
 // For any global variable accessed directly by the specified function, extract that variable,
 // add it as an argument to that function, and the arg to the added_arg_ids collection.
-void CompilerMSL::extract_global_variables_from_function(uint32_t func_id, std::unordered_set<uint32_t> &added_arg_ids,
+void CompilerMSL::extract_global_variables_from_function(uint32_t func_id, std::set<uint32_t> &added_arg_ids,
                                                          std::unordered_set<uint32_t> &global_var_ids,
                                                          std::unordered_set<uint32_t> &processed_func_ids)
 {
@@ -230,7 +235,7 @@ void CompilerMSL::extract_global_variables_from_function(uint32_t func_id, std::
 			case OpFunctionCall:
 			{
 				uint32_t inner_func_id = ops[2];
-				std::unordered_set<uint32_t> inner_func_args;
+				std::set<uint32_t> inner_func_args;
 				extract_global_variables_from_function(inner_func_id, inner_func_args, global_var_ids,
 				                                       processed_func_ids);
 				added_arg_ids.insert(inner_func_args.begin(), inner_func_args.end());
@@ -361,34 +366,44 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage)
 		{
 			// Flatten the struct members into the interface struct
 			uint32_t mbr_idx = 0;
-			for (auto &member : type.member_types)
+			for (auto &mbr_type_id : type.member_types)
 			{
-				// Add a reference to the member to the interface struct.
+				auto &mbr_type = get<SPIRType>(mbr_type_id);
-				uint32_t ib_mbr_idx = uint32_t(ib_type.member_types.size());
+				if (is_matrix(mbr_type))
-				ib_type.member_types.push_back(member); // membertype.self is different for array types
-
-				// Give the member a name
-				string mbr_name = ensure_valid_name(to_qualified_member_name(type, mbr_idx), "m");
-				set_member_name(ib_type_id, ib_mbr_idx, mbr_name);
-
-				// Update the original variable reference to include the structure reference
-				string qual_var_name = ib_var_ref + "." + mbr_name;
-				set_member_qualified_name(type_id, mbr_idx, qual_var_name);
-
-				// Copy the variable location from the original variable to the member
-				uint32_t locn = get_member_decoration(type_id, mbr_idx, DecorationLocation);
-				set_member_decoration(ib_type_id, ib_mbr_idx, DecorationLocation, locn);
-				mark_location_as_used_by_shader(locn, storage);
-
-				// Mark the member as builtin if needed
-				BuiltIn builtin;
-				if (is_member_builtin(type, mbr_idx, &builtin))
 				{
-					set_member_decoration(ib_type_id, ib_mbr_idx, DecorationBuiltIn, builtin);
+					exclude_member_from_stage_in(type, mbr_idx);
-					if (builtin == BuiltInPosition)
-						qual_pos_var_name = qual_var_name;
 				}
+				else
+				{
+					// Add a reference to the member to the interface struct.
+					uint32_t ib_mbr_idx = uint32_t(ib_type.member_types.size());
+					ib_type.member_types.push_back(mbr_type_id); // membertype.self is different for array types
 
+					// Give the member a name
+					string mbr_name = ensure_valid_name(to_qualified_member_name(type, mbr_idx), "m");
+					set_member_name(ib_type_id, ib_mbr_idx, mbr_name);
+
+					// Update the original variable reference to include the structure reference
+					string qual_var_name = ib_var_ref + "." + mbr_name;
+					set_member_qualified_name(type_id, mbr_idx, qual_var_name);
+
+					// Copy the variable location from the original variable to the member
+					if (get_member_decoration_mask(type_id, mbr_idx) & (1ull << DecorationLocation))
+					{
+						uint32_t locn = get_member_decoration(type_id, mbr_idx, DecorationLocation);
+						set_member_decoration(ib_type_id, ib_mbr_idx, DecorationLocation, locn);
+						mark_location_as_used_by_shader(locn, storage);
+					}
+
+					// Mark the member as builtin if needed
+					BuiltIn builtin;
+					if (is_member_builtin(type, mbr_idx, &builtin))
+					{
+						set_member_decoration(ib_type_id, ib_mbr_idx, DecorationBuiltIn, builtin);
+						if (builtin == BuiltInPosition)
+							qual_pos_var_name = qual_var_name;
+					}
+				}
 				mbr_idx++;
 			}
 		}
@@ -398,33 +413,40 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage)
 		         type.basetype == SPIRType::Float || type.basetype == SPIRType::Double ||
 		         type.basetype == SPIRType::Boolean)
 		{
-			// Add a reference to the variable type to the interface struct.
+			if (is_matrix(type))
-			uint32_t ib_mbr_idx = uint32_t(ib_type.member_types.size());
-			ib_type.member_types.push_back(type_id);
-
-			// Give the member a name
-			string mbr_name = ensure_valid_name(to_expression(p_var->self), "m");
-			set_member_name(ib_type_id, ib_mbr_idx, mbr_name);
-
-			// Update the original variable reference to include the structure reference
-			string qual_var_name = ib_var_ref + "." + mbr_name;
-			meta[p_var->self].decoration.qualified_alias = qual_var_name;
-
-			// Copy the variable location from the original variable to the member
-			auto &dec = meta[p_var->self].decoration;
-			uint32_t locn = dec.location;
-			if (get_decoration_mask(p_var->self) & (1ull << DecorationLocation))
 			{
-				set_member_decoration(ib_type_id, ib_mbr_idx, DecorationLocation, locn);
+				exclude_from_stage_in(*p_var);
 			}
-			mark_location_as_used_by_shader(locn, storage);
+			else
-
-			// Mark the member as builtin if needed
-			if (is_builtin_variable(*p_var))
 			{
-				set_member_decoration(ib_type_id, ib_mbr_idx, DecorationBuiltIn, dec.builtin_type);
+				// Add a reference to the variable type to the interface struct.
-				if (dec.builtin_type == BuiltInPosition)
+				uint32_t ib_mbr_idx = uint32_t(ib_type.member_types.size());
-					qual_pos_var_name = qual_var_name;
+				ib_type.member_types.push_back(type_id);
+
+				// Give the member a name
+				string mbr_name = ensure_valid_name(to_expression(p_var->self), "m");
+				set_member_name(ib_type_id, ib_mbr_idx, mbr_name);
+
+				// Update the original variable reference to include the structure reference
+				string qual_var_name = ib_var_ref + "." + mbr_name;
+				meta[p_var->self].decoration.qualified_alias = qual_var_name;
+
+				// Copy the variable location from the original variable to the member
+				if (get_decoration_mask(p_var->self) & (1ull << DecorationLocation))
+				{
+					uint32_t locn = get_decoration(p_var->self, DecorationLocation);
+					set_member_decoration(ib_type_id, ib_mbr_idx, DecorationLocation, locn);
+					mark_location_as_used_by_shader(locn, storage);
+				}
+
+				// Mark the member as builtin if needed
+				if (is_builtin_variable(*p_var))
+				{
+					uint32_t builtin = get_decoration(p_var->self, DecorationBuiltIn);
+					set_member_decoration(ib_type_id, ib_mbr_idx, DecorationBuiltIn, builtin);
+					if (builtin == BuiltInPosition)
+						qual_pos_var_name = qual_var_name;
+				}
 			}
 		}
 	}
@@ -449,6 +471,169 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage)
 	return ib_var_id;
 }
 
+// Excludes the specified input variable from the stage_in block structure.
+// Instead, the variable is added to a block variable corresponding to a secondary MSL buffer.
+// The main use case for this is when a stage_in variable contains a matrix, which is a rare occurrence.
+void CompilerMSL::exclude_from_stage_in(SPIRVariable &var)
+{
+	uint32_t var_id = var.self;
+
+	if (!(get_decoration_mask(var_id) & (1ull << DecorationLocation)))
+		return;
+
+	uint32_t mbr_type_id = var.basetype;
+	string mbr_name = ensure_valid_name(to_expression(var_id), "m");
+	uint32_t mbr_locn = get_decoration(var_id, DecorationLocation);
+	meta[var_id].decoration.qualified_alias = add_input_buffer_block_member(mbr_type_id, mbr_name, mbr_locn);
+}
+
+// Excludes the specified type member from the stage_in block structure.
+// Instead, the member is added to a block variable corresponding to a secondary MSL buffer.
+// The main use case for this is when a stage_in variable contains a matrix, which is a rare occurrence.
+void CompilerMSL::exclude_member_from_stage_in(const SPIRType &type, uint32_t index)
+{
+	uint32_t type_id = type.self;
+
+	if (!(get_member_decoration_mask(type_id, index) & (1ull << DecorationLocation)))
+		return;
+
+	uint32_t mbr_type_id = type.member_types[index];
+	string mbr_name = ensure_valid_name(to_qualified_member_name(type, index), "m");
+	uint32_t mbr_locn = get_member_decoration(type_id, index, DecorationLocation);
+	string qual_name = add_input_buffer_block_member(mbr_type_id, mbr_name, mbr_locn);
+	set_member_qualified_name(type_id, index, qual_name);
+}
+
+// Adds a member to the input buffer block that corresponds to the MTLBuffer used by an attribute location
+string CompilerMSL::add_input_buffer_block_member(uint32_t mbr_type_id, string mbr_name, uint32_t mbr_locn)
+{
+	mark_location_as_used_by_shader(mbr_locn, StorageClassInput);
+
+	MSLVertexAttr *p_va = vtx_attrs_by_location[mbr_locn];
+	if (!p_va)
+		return "";
+
+	if (p_va->per_instance)
+		needs_instance_idx_arg = true;
+	else
+		needs_vertex_idx_arg = true;
+
+	// The variable that is the block struct.
+	// Record the stride of this struct in its offset decoration.
+	uint32_t ib_var_id = get_input_buffer_block_var_id(p_va->msl_buffer);
+	auto &ib_var = get<SPIRVariable>(ib_var_id);
+	uint32_t ib_type_id = ib_var.basetype;
+	auto &ib_type = get<SPIRType>(ib_type_id);
+	set_decoration(ib_type_id, DecorationOffset, p_va->msl_stride);
+
+	// Add a reference to the variable type to the interface struct.
+	uint32_t ib_mbr_idx = uint32_t(ib_type.member_types.size());
+	ib_type.member_types.push_back(mbr_type_id);
+
+	// Give the member a name
+	set_member_name(ib_type_id, ib_mbr_idx, mbr_name);
+
+	// Set MSL buffer and offset decorations, and indicate no valid attribute location
+	set_member_decoration(ib_type_id, ib_mbr_idx, DecorationBinding, p_va->msl_buffer);
+	set_member_decoration(ib_type_id, ib_mbr_idx, DecorationOffset, p_va->msl_offset);
+	set_member_decoration(ib_type_id, ib_mbr_idx, DecorationLocation, k_unknown_location);
+
+	// Update the original variable reference to include the structure and index reference
+	string idx_var_name = builtin_to_glsl(p_va->per_instance ? BuiltInInstanceIndex : BuiltInVertexIndex);
+	return get_name(ib_var_id) + "[" + idx_var_name + "]." + mbr_name;
+}
+
+// Returns the ID of the input block that will use the specified MSL buffer index,
+// lazily creating an input block variable and type if needed.
+//
+// The use of this block applies only to input variables that have been excluded from the stage_in
+// block, which typically only occurs if an attempt to pass a matrix in the stage_in block.
+uint32_t CompilerMSL::get_input_buffer_block_var_id(uint32_t msl_buffer)
+{
+	uint32_t ib_var_id = non_stage_in_input_var_ids[msl_buffer];
+	if (!ib_var_id)
+	{
+		// No interface block exists yet. Create a new typed variable for this interface block.
+		// The initializer expression is allocated here, but populated when the function
+		// declaraion is emitted, because it is cleared after each compilation pass.
+		uint32_t next_id = increase_bound_by(3);
+		uint32_t ib_type_id = next_id++;
+		auto &ib_type = set<SPIRType>(ib_type_id);
+		ib_type.basetype = SPIRType::Struct;
+		ib_type.storage = StorageClassInput;
+		set_decoration(ib_type_id, DecorationBlock);
+
+		ib_var_id = next_id++;
+		auto &var = set<SPIRVariable>(ib_var_id, ib_type_id, StorageClassInput, 0);
+		var.initializer = next_id++;
+
+		string ib_var_name = stage_in_var_name + convert_to_string(msl_buffer);
+		set_name(ib_var_id, ib_var_name);
+		set_name(ib_type_id, get_entry_point_name() + "_" + ib_var_name);
+
+		// Add the variable to the map of buffer blocks, accessed by the Metal buffer index.
+		non_stage_in_input_var_ids[msl_buffer] = ib_var_id;
+	}
+	return ib_var_id;
+}
+
+// Sort the members of the struct type by offset, and pad where needed.
+void CompilerMSL::pad_input_buffer_block(uint32_t ib_type_id)
+{
+	auto &ib_type = get<SPIRType>(ib_type_id);
+
+	// Sort the members of the interface structure by their offset.
+	MemberSorter member_sorter(ib_type, meta[ib_type_id], MemberSorter::Offset);
+	member_sorter.sort();
+
+	uint32_t curr_offset = 0;
+	for (uint32_t mbr_idx = 0; mbr_idx < ib_type.member_types.size(); mbr_idx++)
+	{
+		uint32_t mbr_offset = get_member_decoration(ib_type_id, mbr_idx, DecorationOffset);
+		uint32_t gap = mbr_offset - curr_offset;
+		if (gap > 0)
+		{
+			ib_type.member_types.insert(ib_type.member_types.begin() + mbr_idx, get_pad_type(gap).self);
+			set_member_name(ib_type_id, mbr_idx, ("pad" + convert_to_string(mbr_idx)));
+			set_member_decoration(ib_type_id, mbr_idx, DecorationOffset, curr_offset);
+			mbr_idx++; // Now move to the actual member
+		}
+		curr_offset = mbr_offset + uint32_t(get_declared_struct_member_size(ib_type, mbr_idx));
+	}
+
+	// Finally, check if we need to pad to the end of the struct to match its stride
+	uint32_t bb_size = get_decoration(ib_type_id, DecorationOffset);
+	uint32_t gap = bb_size - curr_offset;
+	if (gap > 0)
+	{
+		uint32_t mbr_idx = uint32_t(ib_type.member_types.size());
+		ib_type.member_types.push_back(get_pad_type(gap).self);
+		set_member_name(ib_type_id, mbr_idx, ("pad" + convert_to_string(mbr_idx)));
+		set_member_decoration(ib_type_id, mbr_idx, DecorationOffset, curr_offset);
+		mbr_idx++;
+	}
+}
+
+// Returns a char array type suitable for use as a padding member in a packed struct
+SPIRType &CompilerMSL::get_pad_type(uint32_t pad_len)
+{
+	uint32_t pad_type_id = pad_type_ids_by_pad_len[pad_len];
+	if (pad_type_id != 0)
+		return get<SPIRType>(pad_type_id);
+
+	pad_type_id = increase_bound_by(1);
+	auto &pad_type = set<SPIRType>(pad_type_id);
+	pad_type.storage = StorageClassGeneric;
+	pad_type.basetype = SPIRType::Char;
+	pad_type.width = 8;
+	pad_type.array.push_back(pad_len);
+	pad_type.array_size_literal.push_back(true);
+	set_decoration(pad_type.self, DecorationArrayStride, pad_len);
+
+	pad_type_ids_by_pad_len[pad_len] = pad_type_id;
+	return pad_type;
+}
+
 // Emits the file header info
 void CompilerMSL::emit_header()
 {
@@ -530,6 +715,9 @@ void CompilerMSL::emit_resources()
 
 	// Output interface blocks.
 	emit_interface_block(stage_in_var_id);
+	for (auto &nsi_var : non_stage_in_input_var_ids)
+		emit_interface_block(nsi_var.second);
+
 	emit_interface_block(stage_out_var_id);
 	emit_interface_block(stage_uniforms_var_id);
 }
@@ -769,14 +957,17 @@ string CompilerMSL::to_function_name(uint32_t img, const SPIRType &, bool is_fet
 string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool is_fetch, bool, bool is_proj,
                                      uint32_t coord, uint32_t, uint32_t dref, uint32_t grad_x, uint32_t grad_y,
                                      uint32_t lod, uint32_t coffset, uint32_t offset, uint32_t bias, uint32_t comp,
-                                     uint32_t, bool *p_forward)
+                                     uint32_t sample, bool *p_forward)
 {
-	string farg_str = to_sampler_expression(img);
+	string farg_str;
+	if (!is_fetch)
+		farg_str += to_sampler_expression(img);
 
 	// Texture coordinates
 	bool forward = should_forward(coord);
 	auto coord_expr = to_enclosed_expression(coord);
 	string tex_coords = coord_expr;
+	string face_expr;
 	const char *alt_coord = "";
 
 	switch (imgtype.image.dim)
@@ -785,6 +976,9 @@ string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool
 		tex_coords = coord_expr + ".x";
 		remove_duplicate_swizzle(tex_coords);
 
+		if (is_fetch)
+			tex_coords = "uint(" + tex_coords + ")";
+
 		alt_coord = ".y";
 
 		break;
@@ -796,12 +990,18 @@ string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool
 			remove_duplicate_swizzle(coord_x);
 			string coord_y = coord_expr + ".y";
 			remove_duplicate_swizzle(coord_y);
-			tex_coords = "float2(" + coord_x + ", (1.0 - " + coord_y + "))";
+			if (is_fetch)
+				tex_coords = "uint2(" + coord_x + ", (1.0 - " + coord_y + "))";
+			else
+				tex_coords = "float2(" + coord_x + ", (1.0 - " + coord_y + "))";
 		}
 		else
 		{
 			tex_coords = coord_expr + ".xy";
 			remove_duplicate_swizzle(tex_coords);
+
+			if (is_fetch)
+				tex_coords = "uint2(" + tex_coords + ")";
 		}
 
 		alt_coord = ".z";
@@ -809,6 +1009,32 @@ string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool
 		break;
 
 	case Dim3D:
+		if (msl_config.flip_frag_y)
+		{
+			string coord_x = coord_expr + ".x";
+			remove_duplicate_swizzle(coord_x);
+			string coord_y = coord_expr + ".y";
+			remove_duplicate_swizzle(coord_y);
+			string coord_z = coord_expr + ".z";
+			remove_duplicate_swizzle(coord_z);
+			if (is_fetch)
+				tex_coords = "uint3(" + coord_x + ", (1.0 - " + coord_y + "), " + coord_z + ")";
+			else
+				tex_coords = "float3(" + coord_x + ", (1.0 - " + coord_y + "), " + coord_z + ")";
+		}
+		else
+		{
+			tex_coords = coord_expr + ".xyz";
+			remove_duplicate_swizzle(tex_coords);
+
+			if (is_fetch)
+				tex_coords = "uint3(" + tex_coords + ")";
+		}
+
+		alt_coord = ".w";
+
+		break;
+
 	case DimCube:
 		if (msl_config.flip_frag_y)
 		{
@@ -818,12 +1044,31 @@ string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool
 			remove_duplicate_swizzle(coord_y);
 			string coord_z = coord_expr + ".z";
 			remove_duplicate_swizzle(coord_z);
-			tex_coords = "float3(" + coord_x + ", (1.0 - " + coord_y + "), " + coord_z + ")";
+
+			if (is_fetch)
+			{
+				tex_coords = "uint2(" + coord_x + ", (1.0 - " + coord_y + "))";
+				face_expr = coord_z;
+			}
+			else
+				tex_coords = "float3(" + coord_x + ", (1.0 - " + coord_y + "), " + coord_z + ")";
 		}
 		else
 		{
-			tex_coords = coord_expr + ".xyz";
+			if (is_fetch)
-			remove_duplicate_swizzle(tex_coords);
+			{
+				tex_coords = coord_expr + ".xy";
+				remove_duplicate_swizzle(tex_coords);
+				tex_coords = "uint2(" + tex_coords + ")";
+
+				face_expr = coord_expr + ".z";
+				remove_duplicate_swizzle(face_expr);
+			}
+			else
+			{
+				tex_coords = coord_expr + ".xyz";
+				remove_duplicate_swizzle(tex_coords);
+			}
 		}
 
 		alt_coord = ".w";
@@ -834,15 +1079,30 @@ string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool
 		break;
 	}
 
-	// Use alt coord for projection or texture array
+	// If projection, use alt coord as divisor
-	if (imgtype.image.arrayed)
+	if (is_proj)
-		tex_coords += ", " + coord_expr + alt_coord;
+	{
-	else if (is_proj)
+		string divisor = coord_expr + alt_coord;
-		tex_coords += " / " + coord_expr + alt_coord;
+		remove_duplicate_swizzle(divisor);
+		tex_coords += " / " + divisor;
+	}
 
-	farg_str += ", ";
+	if (!farg_str.empty())
+		farg_str += ", ";
 	farg_str += tex_coords;
 
+	// If fetch from cube, add face explicitly
+	if (!face_expr.empty())
+		farg_str += ", uint(" + face_expr + ")";
+
+	// If array, use alt coord
+	if (imgtype.image.arrayed)
+	{
+		string array_index = coord_expr + alt_coord;
+		remove_duplicate_swizzle(array_index);
+		farg_str += ", uint(" + array_index + ")";
+	}
+
 	// Depth compare reference value
 	if (dref)
 	{
@@ -960,6 +1220,12 @@ string CompilerMSL::to_function_args(uint32_t img, const SPIRType &imgtype, bool
 		farg_str += ", " + to_component_argument(comp);
 	}
 
+	if (sample)
+	{
+		farg_str += ", ";
+		farg_str += to_expression(sample);
+	}
+
 	*p_forward = forward;
 
 	return farg_str;
@@ -1080,7 +1346,8 @@ string CompilerMSL::member_attribute_qualifier(const SPIRType &type, uint32_t in
 			}
 		}
 		uint32_t locn = get_ordered_member_location(type.self, index);
-		return string(" [[attribute(") + convert_to_string(locn) + ")]]";
+		if (locn != k_unknown_location)
+			return string(" [[attribute(") + convert_to_string(locn) + ")]]";
 	}
 
 	// Vertex function outputs
@@ -1105,7 +1372,8 @@ string CompilerMSL::member_attribute_qualifier(const SPIRType &type, uint32_t in
 			}
 		}
 		uint32_t locn = get_ordered_member_location(type.self, index);
-		return string(" [[user(locn") + convert_to_string(locn) + ")]]";
+		if (locn != k_unknown_location)
+			return string(" [[user(locn") + convert_to_string(locn) + ")]]";
 	}
 
 	// Fragment function inputs
@@ -1128,7 +1396,8 @@ string CompilerMSL::member_attribute_qualifier(const SPIRType &type, uint32_t in
 			}
 		}
 		uint32_t locn = get_ordered_member_location(type.self, index);
-		return string(" [[user(locn") + convert_to_string(locn) + ")]]";
+		if (locn != k_unknown_location)
+			return string(" [[user(locn") + convert_to_string(locn) + ")]]";
 	}
 
 	// Fragment function outputs
@@ -1147,7 +1416,8 @@ string CompilerMSL::member_attribute_qualifier(const SPIRType &type, uint32_t in
 			}
 		}
 		uint32_t locn = get_ordered_member_location(type.self, index);
-		return string(" [[color(") + convert_to_string(locn) + ")]]";
+		if (locn != k_unknown_location)
+			return string(" [[color(") + convert_to_string(locn) + ")]]";
 	}
 
 	// Compute function inputs
@@ -1310,6 +1580,19 @@ string CompilerMSL::entry_point_args(bool append_comma)
 		ep_args += type_to_glsl(type) + " " + to_name(var.self) + " [[stage_in]]";
 	}
 
+	// Non-stage-in vertex attribute structures
+	for (auto &nsi_var : non_stage_in_input_var_ids)
+	{
+		auto &var = get<SPIRVariable>(nsi_var.second);
+		auto &type = get<SPIRType>(var.basetype);
+
+		if (!ep_args.empty())
+			ep_args += ", ";
+
+		ep_args += "device " + type_to_glsl(type) + "* " + to_name(var.self) + " [[buffer(" +
+		           convert_to_string(nsi_var.first) + ")]]";
+	}
+
 	// Uniforms
 	for (auto &id : ids)
 	{
@@ -1374,6 +1657,13 @@ string CompilerMSL::entry_point_args(bool append_comma)
 		}
 	}
 
+	// 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 (!ep_args.empty() && append_comma)
 		ep_args += ", ";
 
@@ -1763,6 +2053,18 @@ string CompilerMSL::builtin_type_decl(BuiltIn builtin)
 	}
 }
 
+// Returns the declaration of a built-in argument to a function
+string CompilerMSL::built_in_func_arg(BuiltIn builtin, bool prefix_comma)
+{
+	string bi_arg;
+	if (prefix_comma)
+		bi_arg += ", ";
+	bi_arg += builtin_type_decl(builtin);
+	bi_arg += " " + builtin_to_glsl(builtin);
+	bi_arg += " [[" + builtin_qualifier(builtin) + "]]";
+	return bi_arg;
+}
+
 // Returns the effective size of a buffer block struct member.
 size_t CompilerMSL::get_declared_struct_member_size(const SPIRType &struct_type, uint32_t index) const
 {
@@ -1853,11 +2155,11 @@ bool CompilerMSL::OpCodePreprocessor::handle(Op opcode, const uint32_t * /*args*
 // then by the required sorting aspect.
 void CompilerMSL::MemberSorter::sort()
 {
-	// Create a temporary array of consecutive member indices and sort it base on how
+	// Create a temporary array of consecutive member indices and sort it based on how
 	// the members should be reordered, based on builtin and sorting aspect meta info.
 	size_t mbr_cnt = type.member_types.size();
 	vector<uint32_t> mbr_idxs(mbr_cnt);
-	iota(mbr_idxs.begin(), mbr_idxs.end(), 0); // Fill with consecutive indices
+	iota(mbr_idxs.begin(), mbr_idxs.end(), 0);          // Fill with consecutive indices
 	std::sort(mbr_idxs.begin(), mbr_idxs.end(), *this); // Sort member indices based on sorting aspect
 
 	// Move type and meta member info to the order defined by the sorted member indices.

spirv_msl.hpp

@@ -1,5 +1,5 @@
 /*
- * Copyright 2015-2016 The Brenwill Workshop Ltd.
+ * Copyright 2016-2017 The Brenwill Workshop Ltd.
  *
  * Licensed under the Apache License, Version 2.0 (the "License");
  * you may not use this file except in compliance with the License.
@@ -18,6 +18,7 @@
 #define SPIRV_CROSS_MSL_HPP
 
 #include "spirv_glsl.hpp"
+#include <map>
 #include <set>
 #include <unordered_map>
 #include <unordered_set>
@@ -41,6 +42,10 @@ struct MSLConfiguration
 struct MSLVertexAttr
 {
 	uint32_t location = 0;
+	uint32_t msl_buffer = 0;
+	uint32_t msl_offset = 0;
+	uint32_t msl_stride = 0;
+	bool per_instance = false;
 	bool used_by_shader = false;
 };
 
@@ -124,8 +129,8 @@ protected:
 	void localize_global_variables();
 	void extract_global_variables_from_functions();
 
-	std::unordered_map<uint32_t, std::unordered_set<uint32_t>> function_global_vars;
+	std::unordered_map<uint32_t, std::set<uint32_t>> function_global_vars;
-	void extract_global_variables_from_function(uint32_t func_id, std::unordered_set<uint32_t> &added_arg_ids,
+	void extract_global_variables_from_function(uint32_t func_id, std::set<uint32_t> &added_arg_ids,
 	                                            std::unordered_set<uint32_t> &global_var_ids,
 	                                            std::unordered_set<uint32_t> &processed_func_ids);
 	uint32_t add_interface_block(spv::StorageClass storage);
@@ -145,6 +150,7 @@ protected:
 	std::string to_sampler_expression(uint32_t id);
 	std::string builtin_qualifier(spv::BuiltIn builtin);
 	std::string builtin_type_decl(spv::BuiltIn builtin);
+	std::string built_in_func_arg(spv::BuiltIn builtin, bool prefix_comma);
 	std::string member_attribute_qualifier(const SPIRType &type, uint32_t index);
 	std::string argument_decl(const SPIRFunction::Parameter &arg);
 	uint32_t get_metal_resource_index(SPIRVariable &var, SPIRType::BaseType basetype);
@@ -152,17 +158,27 @@ protected:
 	size_t get_declared_type_size(uint32_t type_id) const;
 	size_t get_declared_type_size(uint32_t type_id, uint64_t dec_mask) const;
 	std::string to_component_argument(uint32_t id);
+	void exclude_from_stage_in(SPIRVariable &var);
+	void exclude_member_from_stage_in(const SPIRType &type, uint32_t index);
+	std::string add_input_buffer_block_member(uint32_t mbr_type_id, std::string mbr_name, uint32_t mbr_locn);
+	uint32_t get_input_buffer_block_var_id(uint32_t msl_buffer);
+	void pad_input_buffer_block(uint32_t ib_type_id);
+	SPIRType &get_pad_type(uint32_t pad_len);
 
 	MSLConfiguration msl_config;
 	std::unordered_map<std::string, std::string> func_name_overrides;
 	std::unordered_map<std::string, std::string> var_name_overrides;
 	std::set<uint32_t> custom_function_ops;
 	std::unordered_map<uint32_t, MSLVertexAttr *> vtx_attrs_by_location;
+	std::map<uint32_t, uint32_t> non_stage_in_input_var_ids;
+	std::unordered_map<uint32_t, uint32_t> pad_type_ids_by_pad_len;
 	std::vector<MSLResourceBinding *> resource_bindings;
 	MSLResourceBinding next_metal_resource_index;
 	uint32_t stage_in_var_id = 0;
 	uint32_t stage_out_var_id = 0;
 	uint32_t stage_uniforms_var_id = 0;
+	bool needs_vertex_idx_arg = false;
+	bool needs_instance_idx_arg = false;
 	std::string qual_pos_var_name;
 	std::string stage_in_var_name = "in";
 	std::string stage_out_var_name = "out";

test_shaders.py

@@ -12,7 +12,6 @@ import shutil
 import argparse
 import codecs
 
-METALC = '/Applications/Xcode.app/Contents/Developer/Platforms/iPhoneOS.platform/usr/bin/metal'
 
 def parse_stats(stats):
     m = re.search('([0-9]+) work registers', stats)
@@ -64,8 +63,25 @@ def get_shader_stats(shader):
     returned = stdout.decode('utf-8')
     return parse_stats(returned)
 
+def print_msl_compiler_version():
+    try:
+        subprocess.check_call(['xcrun', '--sdk', 'iphoneos', 'metal', '--version'])
+        print('...are the Metal compiler characteristics.\n')   # display after so xcrun FNF is silent
+    except OSError as e:
+        if (e.errno != os.errno.ENOENT):    # Ignore xcrun not found error
+            raise
+
 def validate_shader_msl(shader):
-    subprocess.check_call([METALC, '-x', 'metal', '-std=ios-metal1.0', '-Werror', shader])
+    msl_path = reference_path(shader[0], shader[1])
+    try:
+        subprocess.check_call(['xcrun', '--sdk', 'iphoneos', 'metal', '-x', 'metal', '-std=ios-metal1.2', '-Werror', msl_path])
+        print('Compiled Metal shader: ' + msl_path)   # display after so xcrun FNF is silent
+    except OSError as oe:
+        if (oe.errno != os.errno.ENOENT):   # Ignore xcrun not found error
+            raise
+    except subprocess.CalledProcessError:
+        print('Error compiling Metal shader: ' + msl_path)
+        sys.exit(1)
 
 def cross_compile_msl(shader):
     spirv_f, spirv_path = tempfile.mkstemp()
@@ -76,10 +92,6 @@ def cross_compile_msl(shader):
     spirv_cross_path = './spirv-cross'
     subprocess.check_call([spirv_cross_path, '--entry', 'main', '--output', msl_path, spirv_path, '--metal'])
     subprocess.check_call(['spirv-val', spirv_path])
-
-    if os.path.exists(METALC):
-        validate_shader_msl(msl_path)
-
     return (spirv_path, msl_path)
 
 def validate_shader_hlsl(shader):
@@ -260,10 +272,11 @@ def test_shader(stats, shader, update, keep):
 
 def test_shader_msl(stats, shader, update, keep):
     joined_path = os.path.join(shader[0], shader[1])
-    print('Testing MSL shader:', joined_path)
+    print('\nTesting MSL shader:', joined_path)
     spirv, msl = cross_compile_msl(joined_path)
     regression_check(shader, msl, update, keep)
     os.remove(spirv)
+    validate_shader_msl(shader)
 
 def test_shader_hlsl(stats, shader, update, keep):
     joined_path = os.path.join(shader[0], shader[1])
@@ -274,6 +287,7 @@ def test_shader_hlsl(stats, shader, update, keep):
 
 def test_shaders_helper(stats, shader_dir, update, malisc, keep, backend):
     for root, dirs, files in os.walk(os.path.join(shader_dir)):
+        files = [ f for f in files if not f.startswith(".") ]   #ignore system files (esp OSX)
         for i in files:
             path = os.path.join(root, i)
             relpath = os.path.relpath(path, shader_dir)
@@ -317,8 +331,8 @@ def main():
         sys.stderr.write('Need shader folder.\n')
         sys.exit(1)
 
-    if os.path.exists(METALC):
+    if args.metal:
-        subprocess.check_call([METALC, '--version'])
+        print_msl_compiler_version()
 
     test_shaders(args.folder, args.update, args.malisc, args.keep, 'metal' if args.metal else ('hlsl' if args.hlsl else 'glsl'))
     if args.malisc: