MSL: Handle array with component when we cannot rely on user() attrib.

In these cases, we emit one variable per location, and so we must
flatten stuff.

3 files changed, 297 insertions, 0 deletions

diff --git a/reference/opt/shaders-msl/frag/array-component-io.frag b/reference/opt/shaders-msl/frag/array-component-io.frag
new file mode 100644
index 00000000..9b4c5b52
--- /dev/null
+++ b/reference/opt/shaders-msl/frag/array-component-io.frag
@@ -0,0 +1,99 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+#pragma clang diagnostic ignored "-Wmissing-braces"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+template<typename T, size_t Num>
+struct spvUnsafeArray
+{
+    T elements[Num ? Num : 1];
+    
+    thread T& operator [] (size_t pos) thread
+    {
+        return elements[pos];
+    }
+    constexpr const thread T& operator [] (size_t pos) const thread
+    {
+        return elements[pos];
+    }
+    
+    device T& operator [] (size_t pos) device
+    {
+        return elements[pos];
+    }
+    constexpr const device T& operator [] (size_t pos) const device
+    {
+        return elements[pos];
+    }
+    
+    constexpr const constant T& operator [] (size_t pos) const constant
+    {
+        return elements[pos];
+    }
+    
+    threadgroup T& operator [] (size_t pos) threadgroup
+    {
+        return elements[pos];
+    }
+    constexpr const threadgroup T& operator [] (size_t pos) const threadgroup
+    {
+        return elements[pos];
+    }
+};
+
+struct main0_out
+{
+    float4 m_location_0 [[color(0)]];
+    float4 m_location_1 [[color(1)]];
+    float4 m_location_2 [[color(2)]];
+};
+
+struct main0_in
+{
+    float InC_0 [[user(locn0_1), flat]];
+    float InA_0 [[user(locn1), flat]];
+    float InC_1 [[user(locn1_1), flat]];
+    float2 InB_0 [[user(locn1_2), flat]];
+    float InA_1 [[user(locn2), flat]];
+    float InC_2 [[user(locn2_1), flat]];
+    float2 InB_1 [[user(locn2_2), flat]];
+    float InD [[user(locn3_1), sample_perspective]];
+    float InE [[user(locn4_2), center_no_perspective]];
+    float InF [[user(locn5_3), centroid_perspective]];
+};
+
+fragment main0_out main0(main0_in in [[stage_in]])
+{
+    main0_out out = {};
+    spvUnsafeArray<float, 2> A = {};
+    spvUnsafeArray<float2, 2> B = {};
+    spvUnsafeArray<float, 3> C = {};
+    float D = {};
+    spvUnsafeArray<float, 2> InA = {};
+    spvUnsafeArray<float2, 2> InB = {};
+    spvUnsafeArray<float, 3> InC = {};
+    InA[0] = in.InA_0;
+    InA[1] = in.InA_1;
+    InB[0] = in.InB_0;
+    InB[1] = in.InB_1;
+    InC[0] = in.InC_0;
+    InC[1] = in.InC_1;
+    InC[2] = in.InC_2;
+    A = InA;
+    B = InB;
+    C = InC;
+    D = (in.InD + in.InE) + in.InF;
+    out.m_location_1.x = A[0];
+    out.m_location_2.x = A[1];
+    out.m_location_1.zw = B[0];
+    out.m_location_2.zw = B[1];
+    out.m_location_0.y = C[0];
+    out.m_location_1.y = C[1];
+    out.m_location_2.y = C[2];
+    out.m_location_0.w = D;
+    return out;
+}
+
diff --git a/reference/opt/shaders-msl/vert/array-component-io.for-tess.vert b/reference/opt/shaders-msl/vert/array-component-io.for-tess.vert
new file mode 100644
index 00000000..24958eb5
--- /dev/null
+++ b/reference/opt/shaders-msl/vert/array-component-io.for-tess.vert
@@ -0,0 +1,98 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+#pragma clang diagnostic ignored "-Wmissing-braces"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+template<typename T, size_t Num>
+struct spvUnsafeArray
+{
+    T elements[Num ? Num : 1];
+    
+    thread T& operator [] (size_t pos) thread
+    {
+        return elements[pos];
+    }
+    constexpr const thread T& operator [] (size_t pos) const thread
+    {
+        return elements[pos];
+    }
+    
+    device T& operator [] (size_t pos) device
+    {
+        return elements[pos];
+    }
+    constexpr const device T& operator [] (size_t pos) const device
+    {
+        return elements[pos];
+    }
+    
+    constexpr const constant T& operator [] (size_t pos) const constant
+    {
+        return elements[pos];
+    }
+    
+    threadgroup T& operator [] (size_t pos) threadgroup
+    {
+        return elements[pos];
+    }
+    constexpr const threadgroup T& operator [] (size_t pos) const threadgroup
+    {
+        return elements[pos];
+    }
+};
+
+struct main0_out
+{
+    float4 m_location_0;
+    float4 m_location_1;
+    float4 m_location_2;
+    float4 gl_Position;
+};
+
+struct main0_in
+{
+    float4 m_location_0 [[attribute(0)]];
+    float4 m_location_1 [[attribute(1)]];
+    float4 m_location_2 [[attribute(2)]];
+    float4 Pos [[attribute(4)]];
+};
+
+kernel void main0(main0_in in [[stage_in]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], uint3 spvStageInputSize [[grid_size]], device main0_out* spvOut [[buffer(28)]])
+{
+    spvUnsafeArray<float, 2> A = {};
+    spvUnsafeArray<float2, 2> B = {};
+    spvUnsafeArray<float, 3> C = {};
+    float D = {};
+    spvUnsafeArray<float, 2> InA = {};
+    spvUnsafeArray<float2, 2> InB = {};
+    spvUnsafeArray<float, 3> InC = {};
+    float InD = {};
+    device main0_out& out = spvOut[gl_GlobalInvocationID.y * spvStageInputSize.x + gl_GlobalInvocationID.x];
+    InA[0] = in.m_location_1.x;
+    InA[1] = in.m_location_2.x;
+    InB[0] = in.m_location_1.zw;
+    InB[1] = in.m_location_2.zw;
+    InC[0] = in.m_location_0.y;
+    InC[1] = in.m_location_1.y;
+    InC[2] = in.m_location_2.y;
+    InD = in.m_location_0.w;
+    if (any(gl_GlobalInvocationID >= spvStageInputSize))
+        return;
+    out.gl_Position = in.Pos;
+    A = InA;
+    B = InB;
+    C = InC;
+    D = InD;
+    out.m_location_1.x = A[0];
+    out.m_location_2.x = A[1];
+    out.m_location_1.zw = B[0];
+    out.m_location_2.zw = B[1];
+    out.m_location_0.y = C[0];
+    out.m_location_1.y = C[1];
+    out.m_location_2.y = C[2];
+    out.m_location_0.w = D;
+}
+
diff --git a/reference/opt/shaders-msl/vert/array-component-io.vert b/reference/opt/shaders-msl/vert/array-component-io.vert
new file mode 100644
index 00000000..352c9d2e
--- /dev/null
+++ b/reference/opt/shaders-msl/vert/array-component-io.vert
@@ -0,0 +1,100 @@
+#pragma clang diagnostic ignored "-Wmissing-prototypes"
+#pragma clang diagnostic ignored "-Wmissing-braces"
+
+#include <metal_stdlib>
+#include <simd/simd.h>
+
+using namespace metal;
+
+template<typename T, size_t Num>
+struct spvUnsafeArray
+{
+    T elements[Num ? Num : 1];
+    
+    thread T& operator [] (size_t pos) thread
+    {
+        return elements[pos];
+    }
+    constexpr const thread T& operator [] (size_t pos) const thread
+    {
+        return elements[pos];
+    }
+    
+    device T& operator [] (size_t pos) device
+    {
+        return elements[pos];
+    }
+    constexpr const device T& operator [] (size_t pos) const device
+    {
+        return elements[pos];
+    }
+    
+    constexpr const constant T& operator [] (size_t pos) const constant
+    {
+        return elements[pos];
+    }
+    
+    threadgroup T& operator [] (size_t pos) threadgroup
+    {
+        return elements[pos];
+    }
+    constexpr const threadgroup T& operator [] (size_t pos) const threadgroup
+    {
+        return elements[pos];
+    }
+};
+
+struct main0_out
+{
+    float C_0 [[user(locn0_1)]];
+    float D [[user(locn0_3)]];
+    float A_0 [[user(locn1)]];
+    float C_1 [[user(locn1_1)]];
+    float2 B_0 [[user(locn1_2)]];
+    float A_1 [[user(locn2)]];
+    float C_2 [[user(locn2_1)]];
+    float2 B_1 [[user(locn2_2)]];
+    float4 gl_Position [[position]];
+};
+
+struct main0_in
+{
+    float4 m_location_0 [[attribute(0)]];
+    float4 m_location_1 [[attribute(1)]];
+    float4 m_location_2 [[attribute(2)]];
+    float4 Pos [[attribute(4)]];
+};
+
+vertex main0_out main0(main0_in in [[stage_in]])
+{
+    main0_out out = {};
+    spvUnsafeArray<float, 2> A = {};
+    spvUnsafeArray<float2, 2> B = {};
+    spvUnsafeArray<float, 3> C = {};
+    spvUnsafeArray<float, 2> InA = {};
+    spvUnsafeArray<float2, 2> InB = {};
+    spvUnsafeArray<float, 3> InC = {};
+    float InD = {};
+    InA[0] = in.m_location_1.x;
+    InA[1] = in.m_location_2.x;
+    InB[0] = in.m_location_1.zw;
+    InB[1] = in.m_location_2.zw;
+    InC[0] = in.m_location_0.y;
+    InC[1] = in.m_location_1.y;
+    InC[2] = in.m_location_2.y;
+    InD = in.m_location_0.w;
+    out.gl_Position = in.Pos;
+    A = InA;
+    B = InB;
+    C = InC;
+    out.D = InD;
+    out.A_0 = A[0];
+    out.A_1 = A[1];
+    out.B_0 = B[0];
+    out.B_1 = B[1];
+    out.C_0 = C[0];
+    out.C_1 = C[1];
+    out.C_2 = C[2];
+    return out;
+}
+