Fixes from review of PR #373.

Code fixes from review.
Refactor MSL tests back to using the SPIRV-Tools
and glslang loaded by checkout_glslang_spirv_tools.sh.

reference/opt/shaders-msl/asm/comp/storage-buffer-basic.asm.comp

@@ -16,7 +16,6 @@ struct _6
 
 kernel void main0(uint3 gl_WorkGroupID [[threadgroup_position_in_grid]], device _6& _8 [[buffer(0)]], device _6& _9 [[buffer(1)]])
 {
-    uint3 _23 = gl_WorkGroupSize;
     _8._m0[gl_WorkGroupID.x] = _9._m0[gl_WorkGroupID.x] + _8._m0[gl_WorkGroupID.x];
 }
 

reference/opt/shaders-msl/asm/frag/default-member-names.asm.frag

@@ -25,7 +25,7 @@ struct _10
     _9 _m12;
 };
 
-constant float _57 = {};
+constant _10 _51 = {};
 
 struct main0_out
 {
@@ -35,7 +35,7 @@ struct main0_out
 fragment main0_out main0()
 {
     main0_out out = {};
-    out.m_3 = float4(_57);
+    out.m_3 = float4(_51._m0, _51._m1, _51._m2, _51._m3);
     return out;
 }
 

reference/opt/shaders-msl/asm/frag/phi-loop-variable.asm.frag

@@ -5,5 +5,10 @@ using namespace metal;
 
 fragment void main0()
 {
+    int _16;
+    int _23;
+    for (int _22 = 35; _22 >= 0; _23 = _22 - 1, _22 = _23)
+    {
+    }
 }
 

reference/opt/shaders-msl/asm/frag/undef-variable-store.asm.frag

@@ -16,8 +16,10 @@ fragment main0_out main0()
     main0_out out = {};
     float4 _20;
     float4 _21;
+    float4 _51;
     float4 _52;
-    do
+    _51 = _50;
+    for (;;)
     {
         if (0.0 != 0.0)
         {
@@ -31,7 +33,7 @@ fragment main0_out main0()
         }
         _52 = _38;
         break;
-    } while (false);
+    }
     out._entryPointOutput = _52;
     return out;
 }

reference/opt/shaders-msl/comp/builtins.comp

@@ -3,15 +3,7 @@
 
 using namespace metal;
 
-constant uint3 gl_WorkGroupSize = uint3(8u, 4u, 2u);
-
 kernel void main0(uint3 gl_LocalInvocationID [[thread_position_in_threadgroup]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], uint gl_LocalInvocationIndex [[thread_index_in_threadgroup]], uint3 gl_NumWorkGroups [[threadgroups_per_grid]], uint3 gl_WorkGroupID [[threadgroup_position_in_grid]])
 {
-    uint3 local_id = gl_LocalInvocationID;
-    uint3 global_id = gl_GlobalInvocationID;
-    uint local_index = gl_LocalInvocationIndex;
-    uint3 work_group_size = gl_WorkGroupSize;
-    uint3 num_work_groups = gl_NumWorkGroups;
-    uint3 work_group_id = gl_WorkGroupID;
 }
 

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

@@ -12,12 +12,10 @@ struct SSBO
 kernel void main0(device SSBO& _13 [[buffer(0)]])
 {
     float4 _17 = _13.data;
-    float2 _27 = float2(10.0);
+    _13.data = float4(_17.x, _17.yz + float2(10.0), _17.w);
-    float2 _28 = _17.yz + _27;
-    _13.data = float4(_17.x, _28, _17.w);
     _13.data = (_17 + _17) + _17;
-    _13.data = _28.xxyy;
+    _13.data = (_17.yz + float2(10.0)).xxyy;
-    _13.data = float4(_28.y);
+    _13.data = float4((_17.yz + float2(10.0)).y);
-    _13.data = float4((_17.zw + _27)[_13.index]);
+    _13.data = float4((_17.zw + float2(10.0))[_13.index]);
 }
 

reference/opt/shaders-msl/comp/image.comp

@@ -5,7 +5,6 @@ using namespace metal;
 
 kernel void main0(texture2d<float> uImageIn [[texture(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], texture2d<float, access::write> uImageOut [[texture(1)]])
 {
-    float4 v = uImageIn.read(uint2((int2(gl_GlobalInvocationID.xy) + int2(uImageIn.get_width(), uImageIn.get_height()))));
+    uImageOut.write(uImageIn.read(uint2((int2(gl_GlobalInvocationID.xy) + int2(uImageIn.get_width(), uImageIn.get_height())))), uint2(int2(gl_GlobalInvocationID.xy)));
-    uImageOut.write(v, uint2(int2(gl_GlobalInvocationID.xy)));
 }
 

reference/opt/shaders-msl/comp/mat3.comp

@@ -10,7 +10,6 @@ struct SSBO2
 
 kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO2& _22 [[buffer(0)]])
 {
-    uint ident = gl_GlobalInvocationID.x;
+    _22.out_data[gl_GlobalInvocationID.x] = float3x3(float3(10.0), float3(20.0), float3(40.0));
-    _22.out_data[ident] = float3x3(float3(10.0), float3(20.0), float3(40.0));
 }
 

reference/opt/shaders-msl/comp/return.comp

@@ -12,6 +12,8 @@ constant int _69 = {};
 
 kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO2& _27 [[buffer(0)]])
 {
+    int _29;
+    int _65;
     if (gl_GlobalInvocationID.x == 2u)
     {
         _27.out_data[gl_GlobalInvocationID.x] = float4(20.0);
@@ -24,7 +26,7 @@ kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], devic
             return;
         }
     }
-    while (true)
+    for (int _68 = 0; _68 < 20; _65 = _69 + 1, _68 = _65)
     {
         return;
     }

reference/opt/shaders-msl/comp/rmw-opt.comp

@@ -20,6 +20,7 @@ kernel void main0(device SSBO& _9 [[buffer(0)]])
     _9.a ^= 10;
     _9.a %= 40;
     _9.a |= 1;
-    _9.a = 0;
+    bool _65 = false && true;
+    _9.a = int(_65 && (true || _65));
 }
 

reference/opt/shaders-msl/comp/torture-loop.comp

@@ -0,0 +1,85 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct SSBO
+{
+    float4x4 mvp;
+    float4 in_data[1];
+};
+
+struct SSBO2
+{
+    float4 out_data[1];
+};
+
+constant uint _98 = {};
+
+kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _24 [[buffer(0)]], device SSBO2& _89 [[buffer(1)]])
+{
+    float4 _30;
+    float4 _46;
+    int _33;
+    int _48;
+    int _40;
+    int _77;
+    uint _12;
+    uint _75;
+    float4 _71;
+    uint _73;
+    int _83;
+    _30 = _24.in_data[gl_GlobalInvocationID.x];
+    float4 _93;
+    int _94;
+    _93 = _30;
+    _94 = 0;
+    for (;;)
+    {
+        _40 = _94 + 1;
+        if (_40 < 10)
+        {
+            _46 = _93 * 2.0;
+            _48 = _40 + 1;
+            _93 = _46;
+            _94 = _48;
+            continue;
+        }
+        else
+        {
+            break;
+        }
+    }
+    float4 _95;
+    int _96;
+    _95 = _93;
+    _96 = _40;
+    float4 _100;
+    uint _101;
+    for (uint _97 = 0u, _99 = _98; _97 < 16u; _75 = _97 + uint(1), _77 = _96 + 1, _95 = _100, _96 = _77, _97 = _75, _99 = _101)
+    {
+        _100 = _95;
+        _101 = 0u;
+        for (; _101 < 30u; _73 = _101 + uint(1), _100 = _71, _101 = _73)
+        {
+            _71 = _24.mvp * _100;
+        }
+    }
+    int _102;
+    _102 = _96;
+    for (;;)
+    {
+        _83 = _102 + 1;
+        if (_83 > 10)
+        {
+            _102 = _83;
+            continue;
+        }
+        else
+        {
+            break;
+        }
+    }
+    _89.out_data[gl_GlobalInvocationID.x] = _95;
+}
+

reference/opt/shaders-msl/comp/type-alias.comp

@@ -3,11 +3,21 @@
 
 using namespace metal;
 
+struct S0
+{
+    float4 a;
+};
+
 struct S1
 {
     float4 a;
 };
 
+struct SSBO0
+{
+    S0 s0s[1];
+};
+
 struct SSBO1
 {
     S1 s1s[1];
@@ -18,8 +28,8 @@ struct SSBO2
     float4 outputs[1];
 };
 
-kernel void main0(device SSBO1& _36 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO1& _55 [[buffer(1)]], device SSBO2& _66 [[buffer(2)]])
+kernel void main0(device SSBO0& _36 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO1& _55 [[buffer(1)]], device SSBO2& _66 [[buffer(2)]])
 {
-    _66.outputs[gl_GlobalInvocationID.x] = _36.s1s[gl_GlobalInvocationID.x].a + _55.s1s[gl_GlobalInvocationID.x].a;
+    _66.outputs[gl_GlobalInvocationID.x] = _36.s0s[gl_GlobalInvocationID.x].a + _55.s1s[gl_GlobalInvocationID.x].a;
 }
 

reference/opt/shaders-msl/comp/udiv.comp

@@ -3,13 +3,18 @@
 
 using namespace metal;
 
+struct SSBO2
+{
+    uint outputs[1];
+};
+
 struct SSBO
 {
     uint inputs[1];
 };
 
-kernel void main0(device SSBO& _10 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(1)]])
+kernel void main0(device SSBO2& _10 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(1)]])
 {
-    _10.inputs[gl_GlobalInvocationID.x] = _23.inputs[gl_GlobalInvocationID.x] / 29u;
+    _10.outputs[gl_GlobalInvocationID.x] = _23.inputs[gl_GlobalInvocationID.x] / 29u;
 }
 

reference/opt/shaders-msl/desktop-only/frag/image-ms.desktop.frag

@@ -5,9 +5,7 @@ using namespace metal;
 
 fragment void main0(texture2d_ms<float> uImageMS [[texture(0)]], texture2d_array<float, access::read_write> uImageArray [[texture(1)]], texture2d<float, access::write> uImage [[texture(2)]])
 {
-    float4 a = uImageMS.read(uint2(int2(1, 2)), 2);
+    uImage.write(uImageMS.read(uint2(int2(1, 2)), 2), uint2(int2(2, 3)));
-    float4 b = uImageArray.read(uint2(int3(1, 2, 4).xy), uint(int3(1, 2, 4).z));
+    uImageArray.write(uImageArray.read(uint2(int3(1, 2, 4).xy), uint(int3(1, 2, 4).z)), uint2(int3(2, 3, 7).xy), uint(int3(2, 3, 7).z));
-    uImage.write(a, uint2(int2(2, 3)));
-    uImageArray.write(b, uint2(int3(2, 3, 7).xy), uint(int3(2, 3, 7).z));
 }
 

reference/opt/shaders-msl/flatten/rowmajor.flatten.vert

@@ -0,0 +1,29 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct UBO
+{
+    float4x4 uMVPR;
+    float4x4 uMVPC;
+    float2x4 uMVP;
+};
+
+struct main0_in
+{
+    float4 aVertex [[attribute(0)]];
+};
+
+struct main0_out
+{
+    float4 gl_Position [[position]];
+};
+
+vertex main0_out main0(main0_in in [[stage_in]], constant UBO& _18 [[buffer(0)]])
+{
+    main0_out out = {};
+    out.gl_Position = (_18.uMVPR * in.aVertex) + (in.aVertex * _18.uMVPC);
+    return out;
+}
+

reference/opt/shaders-msl/frag/bitcasting.frag

@@ -0,0 +1,26 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct main0_in
+{
+    float4 VertGeom [[user(locn0)]];
+};
+
+struct main0_out
+{
+    float4 FragColor0 [[color(0)]];
+    float4 FragColor1 [[color(1)]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]], texture2d<float> TextureBase [[texture(0)]], sampler TextureBaseSmplr [[sampler(0)]], texture2d<float> TextureDetail [[texture(1)]], sampler TextureDetailSmplr [[sampler(1)]])
+{
+    main0_out out = {};
+    float4 _20 = TextureBase.sample(TextureBaseSmplr, in.VertGeom.xy);
+    float4 _31 = TextureDetail.sample(TextureDetailSmplr, in.VertGeom.xy, int2(3, 2));
+    out.FragColor0 = as_type<float4>(as_type<int4>(_20)) * as_type<float4>(as_type<int4>(_31));
+    out.FragColor1 = as_type<float4>(as_type<uint4>(_20)) * as_type<float4>(as_type<uint4>(_31));
+    return out;
+}
+

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

@@ -16,8 +16,9 @@ struct main0_out
 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).x;
+    float4 _19 = Texture.sample(TextureSmplr, in.vTexCoord);
-    out.FragColor = float4(f * f);
+    float _22 = _19.x;
+    out.FragColor = float4(_22 * _22);
     return out;
 }
 

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

@@ -1,5 +1,3 @@
-#pragma clang diagnostic ignored "-Wmissing-prototypes"
-
 #include <metal_stdlib>
 #include <simd/simd.h>
 
@@ -21,20 +19,13 @@ 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])) + float4(10.0 + 20.0)) + float4(indexable_2[in.index].a + indexable_2[in.index].b);
-    out.FragColor = ((indexable[in.index] + (indexable_1[in.index][in.index + 1])) + resolve(param)) + resolve(param_1);
     return out;
 }
 

reference/opt/shaders-msl/frag/flush_params.frag

@@ -8,8 +8,6 @@ struct Structy
     float4 c;
 };
 
-constant Structy _47 = {};
-
 struct main0_out
 {
     float4 FragColor [[color(0)]];

reference/opt/shaders-msl/frag/for-loop-init.frag

@@ -11,9 +11,10 @@ struct main0_out
 fragment main0_out main0()
 {
     main0_out out = {};
+    int _12;
+    int _27;
     int _26;
     int _43;
-    int _12;
     int _59;
     int _83;
     int2 _93;
@@ -21,7 +22,7 @@ fragment main0_out main0()
     int _62;
     int _128;
     out.FragColor = 16;
-    while (true)
+    for (int _140 = 0; _140 < 25; _27 = _140 + 1, _140 = _27)
     {
         out.FragColor += 10;
     }
@@ -29,17 +30,17 @@ fragment main0_out main0()
     {
         out.FragColor += 11;
     }
-    int _143;
+    int _142;
-    _143 = 0;
+    _142 = 0;
-    for (; _143 < 20; _59 = _143 + 1, _143 = _59)
+    for (; _142 < 20; _59 = _142 + 1, _142 = _59)
     {
         out.FragColor += 12;
     }
-    _62 = _143 + 3;
+    _62 = _142 + 3;
     out.FragColor += _62;
     if (_62 == 40)
     {
-        for (int _144 = 0; _144 < 40; _83 = _144 + 1, _144 = _83)
+        for (int _143 = 0; _143 < 40; _83 = _143 + 1, _143 = _83)
         {
             out.FragColor += 13;
         }
@@ -49,15 +50,15 @@ fragment main0_out main0()
     {
         out.FragColor += _62;
     }
-    int2 _145;
+    int2 _144;
-    _145 = int2(0);
+    _144 = int2(0);
-    for (; _145.x < 10; _139 = _145, _139.x = _145.x + 4, _145 = _139)
+    for (; _144.x < 10; _139 = _144, _139.x = _144.x + 4, _144 = _139)
     {
-        out.FragColor += _145.y;
+        out.FragColor += _144.y;
     }
-    for (int _146 = _62; _146 < 40; _128 = _146 + 1, _146 = _128)
+    for (int _145 = _62; _145 < 40; _128 = _145 + 1, _145 = _128)
     {
-        out.FragColor += _146;
+        out.FragColor += _145;
     }
     out.FragColor += _62;
     return out;

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

@@ -11,8 +11,8 @@ struct main0_out
 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);
+    int2 _17 = int2(gl_FragCoord.xy);
-    out.FragColor = ((uSampler.read(uint2(coord), 0) + uSampler.read(uint2(coord), 1)) + uSampler.read(uint2(coord), 2)) + uSampler.read(uint2(coord), 3);
+    out.FragColor = ((uSampler.read(uint2(_17), 0) + uSampler.read(uint2(_17), 1)) + uSampler.read(uint2(_17), 2)) + uSampler.read(uint2(_17), 3);
     return out;
 }
 

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

@@ -1,5 +1,3 @@
-#pragma clang diagnostic ignored "-Wmissing-prototypes"
-
 #include <metal_stdlib>
 #include <simd/simd.h>
 
@@ -16,16 +14,10 @@ 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);
-}
-
 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 * uTex.sample(uTexSmplr, in.vTex);
-    out.FragColor = in.vColor * sample_texture(uTex, uTexSmplr, param);
     return out;
 }
 

reference/opt/shaders-msl/frag/separate-image-sampler-argument.frag

@@ -1,5 +1,3 @@
-#pragma clang diagnostic ignored "-Wmissing-prototypes"
-
 #include <metal_stdlib>
 #include <simd/simd.h>
 
@@ -10,15 +8,10 @@ struct main0_out
     float4 FragColor [[color(0)]];
 };
 
-float4 samp(thread const texture2d<float> t, thread const sampler s)
-{
-    return t.sample(s, float2(0.5));
-}
-
 fragment main0_out main0(texture2d<float> uDepth [[texture(0)]], sampler uSampler [[sampler(0)]])
 {
     main0_out out = {};
-    out.FragColor = samp(uDepth, uSampler);
+    out.FragColor = uDepth.sample(uSampler, float2(0.5));
     return out;
 }
 

reference/opt/shaders-msl/vulkan/frag/spec-constant.vk.frag

@@ -0,0 +1,22 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+constant float a_tmp [[function_constant(1)]];
+constant float a = is_function_constant_defined(a_tmp) ? a_tmp : 1.0;
+constant float b_tmp [[function_constant(2)]];
+constant float b = is_function_constant_defined(b_tmp) ? b_tmp : 2.0;
+
+struct main0_out
+{
+    float4 FragColor [[color(0)]];
+};
+
+fragment main0_out main0()
+{
+    main0_out out = {};
+    out.FragColor = float4(a + b);
+    return out;
+}
+

reference/shaders-msl-no-opt/comp/torture-loop.comp

@@ -1,51 +0,0 @@
-#include <metal_stdlib>
-#include <simd/simd.h>
-
-using namespace metal;
-
-struct SSBO
-{
-    float4x4 mvp;
-    float4 in_data[1];
-};
-
-struct SSBO2
-{
-    float4 out_data[1];
-};
-
-kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _24 [[buffer(0)]], device SSBO2& _89 [[buffer(1)]])
-{
-    uint ident = gl_GlobalInvocationID.x;
-    float4 idat = _24.in_data[ident];
-    int k = 0;
-    for (;;)
-    {
-        int _39 = k;
-        int _40 = _39 + 1;
-        k = _40;
-        if (_40 < 10)
-        {
-            idat *= 2.0;
-            k++;
-            continue;
-        }
-        else
-        {
-            break;
-        }
-    }
-    for (uint i = 0u; i < 16u; i++, k++)
-    {
-        for (uint j = 0u; j < 30u; j++)
-        {
-            idat = _24.mvp * idat;
-        }
-    }
-    do
-    {
-        k++;
-    } while (k > 10);
-    _89.out_data[ident] = idat;
-}
-

reference/shaders-msl-no-opt/frag/bitcasting.frag

@@ -1,30 +0,0 @@
-#include <metal_stdlib>
-#include <simd/simd.h>
-
-using namespace metal;
-
-struct main0_in
-{
-    float4 VertGeom [[user(locn0)]];
-};
-
-struct main0_out
-{
-    float4 FragColor0 [[color(0)]];
-    float4 FragColor1 [[color(1)]];
-};
-
-fragment main0_out main0(main0_in in [[stage_in]], texture2d<float> TextureBase [[texture(0)]], sampler TextureBaseSmplr [[sampler(0)]], texture2d<float> TextureDetail [[texture(1)]], sampler TextureDetailSmplr [[sampler(1)]])
-{
-    main0_out out = {};
-    float4 texSample0 = TextureBase.sample(TextureBaseSmplr, in.VertGeom.xy);
-    float4 texSample1 = TextureDetail.sample(TextureDetailSmplr, in.VertGeom.xy, int2(3, 2));
-    int4 iResult0 = as_type<int4>(texSample0);
-    int4 iResult1 = as_type<int4>(texSample1);
-    out.FragColor0 = as_type<float4>(iResult0) * as_type<float4>(iResult1);
-    uint4 uResult0 = as_type<uint4>(texSample0);
-    uint4 uResult1 = as_type<uint4>(texSample1);
-    out.FragColor1 = as_type<float4>(uResult0) * as_type<float4>(uResult1);
-    return out;
-}
-

reference/shaders-msl-no-opt/vert/copy.flatten.vert

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

reference/shaders-msl-no-opt/vert/dynamic.flatten.vert

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

reference/shaders-msl-no-opt/vert/functions_nested.vert

@@ -159,7 +159,8 @@ void vs_adjust(thread float4& dst_reg0, thread float4& dst_reg1, thread float4&
     tmp0.y = float4(dot(float4(in_pos.xyz, 1.0), v_309.vc[5])).y;
     tmp0.z = float4(dot(float4(in_pos.xyz, 1.0), v_309.vc[6])).z;
     float4 tmp1;
-    tmp1 = float4(in_tc0.xy.x, in_tc0.xy.y, tmp1.z, tmp1.w);
+    float4 _359 = float4(in_tc0.xy.x, in_tc0.xy.y, tmp1.z, tmp1.w);
+    tmp1 = _359;
     tmp1.z = v_309.vc[15].x;
     dst_reg7.y = float4(dot(float4(tmp1.xyz, 1.0), v_309.vc[8])).y;
     dst_reg7.x = float4(dot(float4(tmp1.xyz, 1.0), v_309.vc[7])).x;

reference/shaders-msl-no-opt/vert/texture_buffer.vert

@@ -1,17 +0,0 @@
-#include <metal_stdlib>
-#include <simd/simd.h>
-
-using namespace metal;
-
-struct main0_out
-{
-    float4 gl_Position [[position]];
-};
-
-vertex main0_out main0(texture2d<float> uSamp [[texture(0)]], texture2d<float> uSampo [[texture(1)]])
-{
-    main0_out out = {};
-    out.gl_Position = uSamp.read(uint2(10, 0)) + uSampo.read(uint2(100, 0));
-    return out;
-}
-

reference/shaders-msl-no-opt/vulkan/frag/spec-constant.vk.frag

@@ -1,74 +0,0 @@
-#include <metal_stdlib>
-#include <simd/simd.h>
-
-using namespace metal;
-
-constant float a_tmp [[function_constant(1)]];
-constant float a = is_function_constant_defined(a_tmp) ? a_tmp : 1.0;
-constant float b_tmp [[function_constant(2)]];
-constant float b = is_function_constant_defined(b_tmp) ? b_tmp : 2.0;
-constant int c_tmp [[function_constant(3)]];
-constant int c = is_function_constant_defined(c_tmp) ? c_tmp : 3;
-constant int d_tmp [[function_constant(4)]];
-constant int d = is_function_constant_defined(d_tmp) ? d_tmp : 4;
-constant uint e_tmp [[function_constant(5)]];
-constant uint e = is_function_constant_defined(e_tmp) ? e_tmp : 5u;
-constant uint f_tmp [[function_constant(6)]];
-constant uint f = is_function_constant_defined(f_tmp) ? f_tmp : 6u;
-constant bool g_tmp [[function_constant(7)]];
-constant bool g = is_function_constant_defined(g_tmp) ? g_tmp : false;
-constant bool h_tmp [[function_constant(8)]];
-constant bool h = is_function_constant_defined(h_tmp) ? h_tmp : true;
-
-struct main0_out
-{
-    float4 FragColor [[color(0)]];
-};
-
-fragment main0_out main0()
-{
-    main0_out out = {};
-    float t0 = a;
-    float t1 = b;
-    uint c0 = (uint(c) + 0u);
-    int c1 = (-c);
-    int c2 = (~c);
-    int c3 = (c + d);
-    int c4 = (c - d);
-    int c5 = (c * d);
-    int c6 = (c / d);
-    uint c7 = (e / f);
-    int c8 = (c % d);
-    uint c9 = (e % f);
-    int c10 = (c >> d);
-    uint c11 = (e >> f);
-    int c12 = (c << d);
-    int c13 = (c | d);
-    int c14 = (c ^ d);
-    int c15 = (c & d);
-    bool c16 = (g || h);
-    bool c17 = (g && h);
-    bool c18 = (!g);
-    bool c19 = (g == h);
-    bool c20 = (g != h);
-    bool c21 = (c == d);
-    bool c22 = (c != d);
-    bool c23 = (c < d);
-    bool c24 = (e < f);
-    bool c25 = (c > d);
-    bool c26 = (e > f);
-    bool c27 = (c <= d);
-    bool c28 = (e <= f);
-    bool c29 = (c >= d);
-    bool c30 = (e >= f);
-    int c31 = c8 + c3;
-    int c32 = int(e + 0u);
-    bool c33 = (c != int(0u));
-    bool c34 = (e != 0u);
-    int c35 = int(g);
-    uint c36 = uint(g);
-    float c37 = float(g);
-    out.FragColor = float4(t0 + t1);
-    return out;
-}
-

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

@@ -0,0 +1,36 @@
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+struct UBO
+{
+    float4x4 uMVPR;
+    float4x4 uMVPC;
+    float2x4 uMVP;
+};
+
+struct main0_in
+{
+    float4 aVertex [[attribute(0)]];
+};
+
+struct main0_out
+{
+    float4 gl_Position [[position]];
+};
+
+// Implementation of a conversion of matrix content from RowMajor to ColumnMajor organization.
+float2x4 spvConvertFromRowMajor2x4(float2x4 m)
+{
+    return float2x4(float4(m[0][0], m[0][2], m[1][0], m[1][2]), float4(m[0][1], m[0][3], m[1][1], m[1][3]));
+}
+
+vertex main0_out main0(main0_in in [[stage_in]], constant UBO& _18 [[buffer(0)]])
+{
+    main0_out out = {};
+    float2 v = in.aVertex * spvConvertFromRowMajor2x4(_18.uMVP);
+    out.gl_Position = (_18.uMVPR * in.aVertex) + (in.aVertex * _18.uMVPC);
+    return out;
+}
+

reference/shaders-msl/vert/functions_nested.vert

@@ -1,189 +0,0 @@
-#pragma clang diagnostic ignored "-Wmissing-prototypes"
-
-#include <metal_stdlib>
-#include <simd/simd.h>
-
-using namespace metal;
-
-struct attr_desc
-{
-    int type;
-    int attribute_size;
-    int starting_offset;
-    int stride;
-    int swap_bytes;
-    int is_volatile;
-};
-
-struct VertexBuffer
-{
-    float4x4 scale_offset_mat;
-    uint vertex_base_index;
-    int4 input_attributes[16];
-};
-
-struct VertexConstantsBuffer
-{
-    float4 vc[16];
-};
-
-constant float4 _295 = {};
-
-struct main0_out
-{
-    float4 tc0 [[user(locn0)]];
-    float4 back_color [[user(locn10)]];
-    float4 gl_Position [[position]];
-};
-
-attr_desc fetch_desc(thread const int& location, constant VertexBuffer& v_227)
-{
-    int attribute_flags = v_227.input_attributes[location].w;
-    attr_desc result;
-    result.type = v_227.input_attributes[location].x;
-    result.attribute_size = v_227.input_attributes[location].y;
-    result.starting_offset = v_227.input_attributes[location].z;
-    result.stride = attribute_flags & 255;
-    result.swap_bytes = (attribute_flags >> 8) & 1;
-    result.is_volatile = (attribute_flags >> 9) & 1;
-    return result;
-}
-
-uint get_bits(thread const uint4& v, thread const int& swap)
-{
-    if (swap != 0)
-    {
-        return ((v.w | (v.z << uint(8))) | (v.y << uint(16))) | (v.x << uint(24));
-    }
-    return ((v.x | (v.y << uint(8))) | (v.z << uint(16))) | (v.w << uint(24));
-}
-
-float4 fetch_attr(thread const attr_desc& desc, thread const int& vertex_id, thread const texture2d<uint> input_stream)
-{
-    float4 result = float4(0.0, 0.0, 0.0, 1.0);
-    bool reverse_order = false;
-    int first_byte = (vertex_id * desc.stride) + desc.starting_offset;
-    for (int n = 0; n < 4; n++)
-    {
-        if (n == desc.attribute_size)
-        {
-            break;
-        }
-        uint4 tmp;
-        switch (desc.type)
-        {
-            case 0:
-            {
-                int _131 = first_byte;
-                first_byte = _131 + 1;
-                tmp.x = input_stream.read(uint2(_131, 0)).x;
-                int _138 = first_byte;
-                first_byte = _138 + 1;
-                tmp.y = input_stream.read(uint2(_138, 0)).x;
-                uint4 param = tmp;
-                int param_1 = desc.swap_bytes;
-                result[n] = float(get_bits(param, param_1));
-                break;
-            }
-            case 1:
-            {
-                int _156 = first_byte;
-                first_byte = _156 + 1;
-                tmp.x = input_stream.read(uint2(_156, 0)).x;
-                int _163 = first_byte;
-                first_byte = _163 + 1;
-                tmp.y = input_stream.read(uint2(_163, 0)).x;
-                int _170 = first_byte;
-                first_byte = _170 + 1;
-                tmp.z = input_stream.read(uint2(_170, 0)).x;
-                int _177 = first_byte;
-                first_byte = _177 + 1;
-                tmp.w = input_stream.read(uint2(_177, 0)).x;
-                uint4 param_2 = tmp;
-                int param_3 = desc.swap_bytes;
-                result[n] = as_type<float>(get_bits(param_2, param_3));
-                break;
-            }
-            case 2:
-            {
-                int _195 = first_byte;
-                first_byte = _195 + 1;
-                result[n] = float(input_stream.read(uint2(_195, 0)).x);
-                reverse_order = desc.swap_bytes != 0;
-                break;
-            }
-        }
-    }
-    float4 _209;
-    if (reverse_order)
-    {
-        _209 = result.wzyx;
-    }
-    else
-    {
-        _209 = result;
-    }
-    return _209;
-}
-
-float4 read_location(thread const int& location, constant VertexBuffer& v_227, thread uint& gl_VertexIndex, thread texture2d<uint> buff_in_2, thread texture2d<uint> buff_in_1)
-{
-    int param = location;
-    attr_desc desc = fetch_desc(param, v_227);
-    int vertex_id = gl_VertexIndex - int(v_227.vertex_base_index);
-    if (desc.is_volatile != 0)
-    {
-        attr_desc param_1 = desc;
-        int param_2 = vertex_id;
-        return fetch_attr(param_1, param_2, buff_in_2);
-    }
-    else
-    {
-        attr_desc param_3 = desc;
-        int param_4 = vertex_id;
-        return fetch_attr(param_3, param_4, buff_in_1);
-    }
-}
-
-void vs_adjust(thread float4& dst_reg0, thread float4& dst_reg1, thread float4& dst_reg7, constant VertexBuffer& v_227, thread uint& gl_VertexIndex, thread texture2d<uint> buff_in_2, thread texture2d<uint> buff_in_1, constant VertexConstantsBuffer& v_309)
-{
-    int param = 3;
-    float4 in_diff_color = read_location(param, v_227, gl_VertexIndex, buff_in_2, buff_in_1);
-    int param_1 = 0;
-    float4 in_pos = read_location(param_1, v_227, gl_VertexIndex, buff_in_2, buff_in_1);
-    int param_2 = 8;
-    float4 in_tc0 = read_location(param_2, v_227, gl_VertexIndex, buff_in_2, buff_in_1);
-    dst_reg1 = in_diff_color * v_309.vc[13];
-    float4 tmp0;
-    tmp0.x = float4(dot(float4(in_pos.xyz, 1.0), v_309.vc[4])).x;
-    tmp0.y = float4(dot(float4(in_pos.xyz, 1.0), v_309.vc[5])).y;
-    tmp0.z = float4(dot(float4(in_pos.xyz, 1.0), v_309.vc[6])).z;
-    float4 tmp1;
-    tmp1 = float4(in_tc0.xy.x, in_tc0.xy.y, tmp1.z, tmp1.w);
-    tmp1.z = v_309.vc[15].x;
-    dst_reg7.y = float4(dot(float4(tmp1.xyz, 1.0), v_309.vc[8])).y;
-    dst_reg7.x = float4(dot(float4(tmp1.xyz, 1.0), v_309.vc[7])).x;
-    dst_reg0.y = float4(dot(float4(tmp0.xyz, 1.0), v_309.vc[1])).y;
-    dst_reg0.x = float4(dot(float4(tmp0.xyz, 1.0), v_309.vc[0])).x;
-}
-
-vertex main0_out main0(constant VertexBuffer& v_227 [[buffer(0)]], uint gl_VertexIndex [[vertex_id]], texture2d<uint> buff_in_2 [[texture(0)]], texture2d<uint> buff_in_1 [[texture(1)]], constant VertexConstantsBuffer& v_309 [[buffer(1)]])
-{
-    main0_out out = {};
-    float4 dst_reg0 = float4(0.0, 0.0, 0.0, 1.0);
-    float4 dst_reg1 = float4(0.0);
-    float4 dst_reg7 = float4(0.0);
-    float4 param = dst_reg0;
-    float4 param_1 = dst_reg1;
-    float4 param_2 = dst_reg7;
-    vs_adjust(param, param_1, param_2, v_227, gl_VertexIndex, buff_in_2, buff_in_1, v_309);
-    dst_reg0 = param;
-    dst_reg1 = param_1;
-    dst_reg7 = param_2;
-    out.gl_Position = dst_reg0;
-    out.back_color = dst_reg1;
-    out.tc0 = dst_reg7;
-    out.gl_Position *= v_227.scale_offset_mat;
-    return out;
-}
-

shaders-msl/flatten/rowmajor.flatten.vert

@@ -0,0 +1,16 @@
+#version 310 es
+
+layout(std140) uniform UBO
+{
+    layout(column_major) mat4 uMVPR;
+    layout(row_major) mat4 uMVPC;
+    layout(row_major) mat2x4 uMVP;
+};
+
+layout(location = 0) in vec4 aVertex;
+
+void main()
+{
+	vec2 v = aVertex * uMVP;
+	gl_Position = uMVPR * aVertex + uMVPC * aVertex;
+}

spirv_glsl.cpp

@@ -5595,7 +5595,7 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 
 		if (shuffle)
 		{
-			bool allow_fwd = !options.force_temp_use_for_two_vector_shuffles;
+			bool allow_fwd = !backend.force_temp_use_for_two_vector_shuffles;
 			should_fwd = allow_fwd && should_forward(vec0) && should_forward(vec1);
 			trivial_forward = allow_fwd && !expression_is_forwarded(vec0) && !expression_is_forwarded(vec1);
 
@@ -5683,10 +5683,9 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
 	case OpVectorTimesMatrix:
 	case OpMatrixTimesVector:
 	{
-		// If the matrix needs transpose and it is square, just flip the multiply order.
+		// If the matrix needs transpose, just flip the multiply order.
-		SPIRType *t;
 		auto *e = maybe_get<SPIRExpression>(ops[opcode == OpMatrixTimesVector ? 2 : 3]);
-		if (e && e->need_transpose && (t = &get<SPIRType>(e->expression_type)) && t->columns == t->vecsize)
+		if (e && e->need_transpose)
 		{
 			e->need_transpose = false;
 			emit_binary_op(ops[0], ops[1], ops[3], ops[2], "*");

spirv_glsl.hpp

@@ -90,13 +90,6 @@ public:
 		// If disabled on older targets, binding decorations will be stripped.
 		bool enable_420pack_extension = true;
 
-		// If true, all two-vector OpVectorShuffles will be forced to use a temporary variable.
-		// When multiple consecutive two-vector shuffles are performed, using temporary variables
-		// can avoid an exponential expansion of code generated for vector constructors.
-		// Since each constructor requires 4 arguments, shuffles upon shuffles can grow
-		// the number of constructor arguments and code generated by a power of 4.
-		bool force_temp_use_for_two_vector_shuffles = false;
-
 		enum Precision
 		{
 			DontCare,
@@ -328,6 +321,8 @@ protected:
 		bool boolean_mix_support = true;
 		bool allow_precision_qualifiers = false;
 		bool can_swizzle_scalar = false;
+		bool force_temp_use_for_two_vector_shuffles = false;
+
 	} backend;
 
 	void emit_struct(SPIRType &type);

spirv_msl.cpp

@@ -57,6 +57,21 @@ string CompilerMSL::compile()
 	// Force a classic "C" locale, reverts when function returns
 	ClassicLocale classic_locale;
 
+	// Do not deal with GLES-isms like precision, older extensions and such.
+	CompilerGLSL::options.vulkan_semantics = true;
+	CompilerGLSL::options.es = false;
+	CompilerGLSL::options.version = 120;
+	backend.float_literal_suffix = false;
+	backend.uint32_t_literal_suffix = true;
+	backend.basic_int_type = "int";
+	backend.basic_uint_type = "uint";
+	backend.discard_literal = "discard_fragment()";
+	backend.swizzle_is_function = false;
+	backend.shared_is_implied = false;
+	backend.native_row_major_matrix = false;
+	backend.flexible_member_array_supported = false;
+	backend.force_temp_use_for_two_vector_shuffles = true;
+
 	replace_illegal_names();
 
 	non_stage_in_input_var_ids.clear();
@@ -88,20 +103,6 @@ string CompilerMSL::compile()
 	if (options.resolve_specialized_array_lengths)
 		resolve_specialized_array_lengths();
 
-	// Do not deal with GLES-isms like precision, older extensions and such.
-	CompilerGLSL::options.vulkan_semantics = true;
-	CompilerGLSL::options.es = false;
-	CompilerGLSL::options.version = 120;
-	backend.float_literal_suffix = false;
-	backend.uint32_t_literal_suffix = true;
-	backend.basic_int_type = "int";
-	backend.basic_uint_type = "uint";
-	backend.discard_literal = "discard_fragment()";
-	backend.swizzle_is_function = false;
-	backend.shared_is_implied = false;
-	backend.native_row_major_matrix = false;
-	backend.flexible_member_array_supported = false;
-
 	uint32_t pass_count = 0;
 	do
 	{
@@ -1606,6 +1607,24 @@ void CompilerMSL::emit_instruction(const Instruction &instruction)
 			emit_barrier(ops[0], ops[1], ops[2]);
 		break;
 
+	case OpVectorTimesMatrix:
+	case OpMatrixTimesVector:
+	{
+		// If the matrix needs transpose and it is square, just flip the multiply order.
+		uint32_t mtx_id = ops[opcode == OpMatrixTimesVector ? 2 : 3];
+		auto *e = maybe_get<SPIRExpression>(mtx_id);
+		auto &t = expression_type(mtx_id);
+		if (e && e->need_transpose && t.columns == t.vecsize)
+		{
+			e->need_transpose = false;
+			emit_binary_op(ops[0], ops[1], ops[3], ops[2], "*");
+			e->need_transpose = true;
+		}
+		else
+			BOP(*);
+		break;
+	}
+
 		// OpOuterProduct
 
 	default:
@@ -2282,8 +2301,7 @@ string CompilerMSL::to_sampler_expression(uint32_t id)
 bool CompilerMSL::is_non_native_row_major_matrix(uint32_t id)
 {
 	// Natively supported row-major matrices do not need to be converted.
-	// Legacy targets do not support row major.
+	if (backend.native_row_major_matrix)
-	if (backend.native_row_major_matrix && !is_legacy())
 		return false;
 
 	// Non-matrix or column-major matrix types do not need to be converted.
@@ -2300,7 +2318,7 @@ bool CompilerMSL::is_non_native_row_major_matrix(uint32_t id)
 bool CompilerMSL::member_is_non_native_row_major_matrix(const SPIRType &type, uint32_t index)
 {
 	// Natively supported row-major matrices do not need to be converted.
-	if (backend.native_row_major_matrix && !is_legacy())
+	if (backend.native_row_major_matrix)
 		return false;
 
 	// Non-matrix or column-major matrix types do not need to be converted.
@@ -2722,7 +2740,8 @@ string CompilerMSL::entry_point_args(bool append_comma)
 			{
 				if (!ep_args.empty())
 					ep_args += ", ";
-				BuiltIn bi_type = (BuiltIn)get_decoration(var_id, DecorationBuiltIn);
+
+				BuiltIn bi_type = meta[var_id].decoration.builtin_type;
 				ep_args += builtin_type_decl(bi_type) + " " + to_expression(var_id);
 				ep_args += " [[" + builtin_qualifier(bi_type) + "]]";
 			}
@@ -3431,15 +3450,6 @@ bool CompilerMSL::OpCodePreprocessor::handle(Op opcode, const uint32_t *args, ui
 	case OpCopyMemory:
 	case OpCopyMemorySized:
 	case OpImageWrite:
-	case OpLoopMerge:
-	case OpSelectionMerge:
-	case OpLabel:
-	case OpBranch:
-	case OpBranchConditional:
-	case OpSwitch:
-	case OpReturnValue:
-	case OpLifetimeStart:
-	case OpLifetimeStop:
 	case OpAtomicStore:
 	case OpAtomicFlagClear:
 	case OpEmitStreamVertex:

spirv_msl.hpp

@@ -91,6 +91,7 @@ public:
 		{
 			return platform == iOS;
 		}
+
 		bool is_macos()
 		{
 			return platform == macOS;

update_test_shaders.sh

@@ -11,6 +11,7 @@ echo "Using spirv-opt in: $(which spirv-opt)."
 ./test_shaders.py shaders --update --opt || exit 1
 ./test_shaders.py shaders-msl --msl --update || exit 1
 ./test_shaders.py shaders-msl --msl --update --opt || exit 1
+./test_shaders.py shaders-msl-no-opt --msl --update || exit 1
 ./test_shaders.py shaders-hlsl --hlsl --update || exit 1
 ./test_shaders.py shaders-hlsl --hlsl --update --opt || exit 1