Merge pull request #1675 from KhronosGroup/fix-1670
MSL: Support array-of-component IO variables in vertex/fragment
This commit is contained in:
commit
0214990e7c
99
reference/opt/shaders-msl/frag/array-component-io.frag
Normal file
99
reference/opt/shaders-msl/frag/array-component-io.frag
Normal file
@ -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;
|
||||
}
|
||||
|
@ -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;
|
||||
}
|
||||
|
100
reference/opt/shaders-msl/vert/array-component-io.vert
Normal file
100
reference/opt/shaders-msl/vert/array-component-io.vert
Normal file
@ -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;
|
||||
}
|
||||
|
@ -5,7 +5,7 @@ using namespace metal;
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
float4 FragColor0 [[color(0)]];
|
||||
float4 m_location_0 [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0()
|
||||
@ -17,9 +17,9 @@ fragment main0_out main0()
|
||||
FragColor0 = 1.0;
|
||||
FragColor1 = float2(2.0, 3.0);
|
||||
FragColor3 = 4.0;
|
||||
out.FragColor0.x = FragColor0;
|
||||
out.FragColor0.yz = FragColor1;
|
||||
out.FragColor0.w = FragColor3;
|
||||
out.m_location_0.x = FragColor0;
|
||||
out.m_location_0.yz = FragColor1;
|
||||
out.m_location_0.w = FragColor3;
|
||||
return out;
|
||||
}
|
||||
|
||||
|
@ -5,7 +5,7 @@ using namespace metal;
|
||||
|
||||
struct main0_out
|
||||
{
|
||||
float4 FragColor0 [[color(0)]];
|
||||
float3 m_location_0 [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0()
|
||||
@ -15,8 +15,8 @@ fragment main0_out main0()
|
||||
float2 FragColor1 = {};
|
||||
FragColor0 = 1.0;
|
||||
FragColor1 = float2(2.0, 3.0);
|
||||
out.FragColor0.x = FragColor0;
|
||||
out.FragColor0.yz = FragColor1;
|
||||
out.m_location_0.x = FragColor0;
|
||||
out.m_location_0.yz = FragColor1;
|
||||
return out;
|
||||
}
|
||||
|
||||
|
@ -11,7 +11,7 @@ struct main0_out
|
||||
|
||||
struct main0_in
|
||||
{
|
||||
float4 Foo3 [[attribute(0)]];
|
||||
float4 m_location_0 [[attribute(0)]];
|
||||
};
|
||||
|
||||
vertex main0_out main0(main0_in in [[stage_in]])
|
||||
@ -19,8 +19,8 @@ vertex main0_out main0(main0_in in [[stage_in]])
|
||||
main0_out out = {};
|
||||
float3 Foo3 = {};
|
||||
float Foo1 = {};
|
||||
Foo3 = in.Foo3.xyz;
|
||||
Foo1 = in.Foo3.w;
|
||||
Foo3 = in.m_location_0.xyz;
|
||||
Foo1 = in.m_location_0.w;
|
||||
out.gl_Position = float4(Foo3, Foo1);
|
||||
out.Foo = Foo3 + float3(Foo1);
|
||||
return out;
|
||||
|
99
reference/shaders-msl/frag/array-component-io.frag
Normal file
99
reference/shaders-msl/frag/array-component-io.frag
Normal file
@ -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;
|
||||
}
|
||||
|
98
reference/shaders-msl/vert/array-component-io.for-tess.vert
Normal file
98
reference/shaders-msl/vert/array-component-io.for-tess.vert
Normal file
@ -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;
|
||||
}
|
||||
|
100
reference/shaders-msl/vert/array-component-io.vert
Normal file
100
reference/shaders-msl/vert/array-component-io.vert
Normal file
@ -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;
|
||||
}
|
||||
|
21
shaders-msl/frag/array-component-io.frag
Normal file
21
shaders-msl/frag/array-component-io.frag
Normal file
@ -0,0 +1,21 @@
|
||||
#version 450
|
||||
|
||||
layout(location = 1, component = 0) out float A[2];
|
||||
layout(location = 1, component = 2) out vec2 B[2];
|
||||
layout(location = 0, component = 1) out float C[3];
|
||||
layout(location = 0, component = 3) out float D;
|
||||
|
||||
layout(location = 1, component = 0) flat in float InA[2];
|
||||
layout(location = 1, component = 2) flat in vec2 InB[2];
|
||||
layout(location = 0, component = 1) flat in float InC[3];
|
||||
layout(location = 3, component = 1) sample in float InD;
|
||||
layout(location = 4, component = 2) noperspective in float InE;
|
||||
layout(location = 5, component = 3) centroid in float InF;
|
||||
|
||||
void main()
|
||||
{
|
||||
A = InA;
|
||||
B = InB;
|
||||
C = InC;
|
||||
D = InD + InE + InF;
|
||||
}
|
21
shaders-msl/vert/array-component-io.for-tess.vert
Normal file
21
shaders-msl/vert/array-component-io.for-tess.vert
Normal file
@ -0,0 +1,21 @@
|
||||
#version 450
|
||||
|
||||
layout(location = 1, component = 0) out float A[2];
|
||||
layout(location = 1, component = 2) out vec2 B[2];
|
||||
layout(location = 0, component = 1) out float C[3];
|
||||
layout(location = 0, component = 3) out float D;
|
||||
|
||||
layout(location = 1, component = 0) in float InA[2];
|
||||
layout(location = 1, component = 2) in vec2 InB[2];
|
||||
layout(location = 0, component = 1) in float InC[3];
|
||||
layout(location = 0, component = 3) in float InD;
|
||||
layout(location = 4) in vec4 Pos;
|
||||
|
||||
void main()
|
||||
{
|
||||
gl_Position = Pos;
|
||||
A = InA;
|
||||
B = InB;
|
||||
C = InC;
|
||||
D = InD;
|
||||
}
|
21
shaders-msl/vert/array-component-io.vert
Normal file
21
shaders-msl/vert/array-component-io.vert
Normal file
@ -0,0 +1,21 @@
|
||||
#version 450
|
||||
|
||||
layout(location = 1, component = 0) out float A[2];
|
||||
layout(location = 1, component = 2) out vec2 B[2];
|
||||
layout(location = 0, component = 1) out float C[3];
|
||||
layout(location = 0, component = 3) out float D;
|
||||
|
||||
layout(location = 1, component = 0) in float InA[2];
|
||||
layout(location = 1, component = 2) in vec2 InB[2];
|
||||
layout(location = 0, component = 1) in float InC[3];
|
||||
layout(location = 0, component = 3) in float InD;
|
||||
layout(location = 4) in vec4 Pos;
|
||||
|
||||
void main()
|
||||
{
|
||||
gl_Position = Pos;
|
||||
A = InA;
|
||||
B = InB;
|
||||
C = InC;
|
||||
D = InD;
|
||||
}
|
216
spirv_msl.cpp
216
spirv_msl.cpp
@ -1991,6 +1991,92 @@ uint32_t CompilerMSL::build_msl_interpolant_type(uint32_t type_id, bool is_noper
|
||||
return new_type_id;
|
||||
}
|
||||
|
||||
bool CompilerMSL::add_component_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref,
|
||||
SPIRVariable &var,
|
||||
const SPIRType &type,
|
||||
InterfaceBlockMeta &meta)
|
||||
{
|
||||
// Deal with Component decorations.
|
||||
const InterfaceBlockMeta::LocationMeta *location_meta = nullptr;
|
||||
uint32_t location = ~0u;
|
||||
if (has_decoration(var.self, DecorationLocation))
|
||||
{
|
||||
location = get_decoration(var.self, DecorationLocation);
|
||||
auto location_meta_itr = meta.location_meta.find(location);
|
||||
if (location_meta_itr != end(meta.location_meta))
|
||||
location_meta = &location_meta_itr->second;
|
||||
}
|
||||
|
||||
// Check if we need to pad fragment output to match a certain number of components.
|
||||
if (location_meta)
|
||||
{
|
||||
bool pad_fragment_output = has_decoration(var.self, DecorationLocation) &&
|
||||
msl_options.pad_fragment_output_components &&
|
||||
get_entry_point().model == ExecutionModelFragment && storage == StorageClassOutput;
|
||||
|
||||
auto &entry_func = get<SPIRFunction>(ir.default_entry_point);
|
||||
uint32_t start_component = get_decoration(var.self, DecorationComponent);
|
||||
uint32_t type_components = type.vecsize;
|
||||
uint32_t num_components = location_meta->num_components;
|
||||
|
||||
if (pad_fragment_output)
|
||||
{
|
||||
uint32_t locn = get_decoration(var.self, DecorationLocation);
|
||||
num_components = std::max(num_components, get_target_components_for_fragment_location(locn));
|
||||
}
|
||||
|
||||
// We have already declared an IO block member as m_location_N.
|
||||
// Just emit an early-declared variable and fixup as needed.
|
||||
// Arrays need to be unrolled here since each location might need a different number of components.
|
||||
entry_func.add_local_variable(var.self);
|
||||
vars_needing_early_declaration.push_back(var.self);
|
||||
|
||||
if (var.storage == StorageClassInput)
|
||||
{
|
||||
entry_func.fixup_hooks_in.push_back([=, &type, &var]() {
|
||||
if (!type.array.empty())
|
||||
{
|
||||
uint32_t array_size = to_array_size_literal(type);
|
||||
for (uint32_t loc_off = 0; loc_off < array_size; loc_off++)
|
||||
{
|
||||
statement(to_name(var.self), "[", loc_off, "]", " = ", ib_var_ref,
|
||||
".m_location_", location + loc_off,
|
||||
vector_swizzle(type_components, start_component), ";");
|
||||
}
|
||||
}
|
||||
else
|
||||
{
|
||||
statement(to_name(var.self), " = ", ib_var_ref, ".m_location_", location,
|
||||
vector_swizzle(type_components, start_component), ";");
|
||||
}
|
||||
});
|
||||
}
|
||||
else
|
||||
{
|
||||
entry_func.fixup_hooks_out.push_back([=, &type, &var]() {
|
||||
if (!type.array.empty())
|
||||
{
|
||||
uint32_t array_size = to_array_size_literal(type);
|
||||
for (uint32_t loc_off = 0; loc_off < array_size; loc_off++)
|
||||
{
|
||||
statement(ib_var_ref, ".m_location_", location + loc_off,
|
||||
vector_swizzle(type_components, start_component), " = ",
|
||||
to_name(var.self), "[", loc_off, "];");
|
||||
}
|
||||
}
|
||||
else
|
||||
{
|
||||
statement(ib_var_ref, ".m_location_", location,
|
||||
vector_swizzle(type_components, start_component), " = ", to_name(var.self), ";");
|
||||
}
|
||||
});
|
||||
}
|
||||
return true;
|
||||
}
|
||||
else
|
||||
return false;
|
||||
}
|
||||
|
||||
void CompilerMSL::add_plain_variable_to_interface_block(StorageClass storage, const string &ib_var_ref,
|
||||
SPIRType &ib_type, SPIRVariable &var, InterfaceBlockMeta &meta)
|
||||
{
|
||||
@ -2019,65 +2105,14 @@ void CompilerMSL::add_plain_variable_to_interface_block(StorageClass storage, co
|
||||
|
||||
auto &entry_func = get<SPIRFunction>(ir.default_entry_point);
|
||||
|
||||
// Deal with Component decorations.
|
||||
InterfaceBlockMeta::LocationMeta *location_meta = nullptr;
|
||||
if (has_decoration(var.self, DecorationLocation))
|
||||
{
|
||||
auto location_meta_itr = meta.location_meta.find(get_decoration(var.self, DecorationLocation));
|
||||
if (location_meta_itr != end(meta.location_meta))
|
||||
location_meta = &location_meta_itr->second;
|
||||
}
|
||||
if (add_component_variable_to_interface_block(storage, ib_var_ref, var, type, meta))
|
||||
return;
|
||||
|
||||
bool pad_fragment_output = has_decoration(var.self, DecorationLocation) &&
|
||||
msl_options.pad_fragment_output_components &&
|
||||
get_entry_point().model == ExecutionModelFragment && storage == StorageClassOutput;
|
||||
|
||||
// Check if we need to pad fragment output to match a certain number of components.
|
||||
if (location_meta)
|
||||
{
|
||||
start_component = get_decoration(var.self, DecorationComponent);
|
||||
uint32_t num_components = location_meta->num_components;
|
||||
if (pad_fragment_output)
|
||||
{
|
||||
uint32_t locn = get_decoration(var.self, DecorationLocation);
|
||||
num_components = std::max(num_components, get_target_components_for_fragment_location(locn));
|
||||
}
|
||||
|
||||
if (location_meta->ib_index != ~0u)
|
||||
{
|
||||
// We have already declared the variable. Just emit an early-declared variable and fixup as needed.
|
||||
entry_func.add_local_variable(var.self);
|
||||
vars_needing_early_declaration.push_back(var.self);
|
||||
|
||||
if (var.storage == StorageClassInput)
|
||||
{
|
||||
uint32_t ib_index = location_meta->ib_index;
|
||||
entry_func.fixup_hooks_in.push_back([=, &var]() {
|
||||
statement(to_name(var.self), " = ", ib_var_ref, ".", to_member_name(ib_type, ib_index),
|
||||
vector_swizzle(type_components, start_component), ";");
|
||||
});
|
||||
}
|
||||
else
|
||||
{
|
||||
uint32_t ib_index = location_meta->ib_index;
|
||||
entry_func.fixup_hooks_out.push_back([=, &var]() {
|
||||
statement(ib_var_ref, ".", to_member_name(ib_type, ib_index),
|
||||
vector_swizzle(type_components, start_component), " = ", to_name(var.self), ";");
|
||||
});
|
||||
}
|
||||
return;
|
||||
}
|
||||
else
|
||||
{
|
||||
location_meta->ib_index = uint32_t(ib_type.member_types.size());
|
||||
type_id = build_extended_vector_type(type_id, num_components);
|
||||
if (var.storage == StorageClassInput)
|
||||
padded_input = true;
|
||||
else
|
||||
padded_output = true;
|
||||
}
|
||||
}
|
||||
else if (pad_fragment_output)
|
||||
if (pad_fragment_output)
|
||||
{
|
||||
uint32_t locn = get_decoration(var.self, DecorationLocation);
|
||||
target_components = get_target_components_for_fragment_location(locn);
|
||||
@ -2169,11 +2204,8 @@ void CompilerMSL::add_plain_variable_to_interface_block(StorageClass storage, co
|
||||
uint32_t locn = get_decoration(var.self, DecorationLocation);
|
||||
if (storage == StorageClassInput)
|
||||
{
|
||||
type_id = ensure_correct_input_type(var.basetype, locn,
|
||||
location_meta ? location_meta->num_components : 0,
|
||||
meta.strip_array);
|
||||
if (!location_meta)
|
||||
var.basetype = type_id;
|
||||
type_id = ensure_correct_input_type(var.basetype, locn, 0, meta.strip_array);
|
||||
var.basetype = type_id;
|
||||
|
||||
type_id = get_pointee_type_id(type_id);
|
||||
if (meta.strip_array && is_array(get<SPIRType>(type_id)))
|
||||
@ -2193,13 +2225,10 @@ void CompilerMSL::add_plain_variable_to_interface_block(StorageClass storage, co
|
||||
mark_location_as_used_by_shader(locn, type, storage);
|
||||
}
|
||||
|
||||
if (!location_meta)
|
||||
if (get_decoration_bitset(var.self).get(DecorationComponent))
|
||||
{
|
||||
if (get_decoration_bitset(var.self).get(DecorationComponent))
|
||||
{
|
||||
uint32_t component = get_decoration(var.self, DecorationComponent);
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationComponent, component);
|
||||
}
|
||||
uint32_t component = get_decoration(var.self, DecorationComponent);
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationComponent, component);
|
||||
}
|
||||
|
||||
if (get_decoration_bitset(var.self).get(DecorationIndex))
|
||||
@ -2229,10 +2258,7 @@ void CompilerMSL::add_plain_variable_to_interface_block(StorageClass storage, co
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationSample);
|
||||
}
|
||||
|
||||
// If we have location meta, there is no unique OrigID. We won't need it, since we flatten/unflatten
|
||||
// the variable to stack anyways here.
|
||||
if (!location_meta)
|
||||
set_extended_member_decoration(ib_type.self, ib_mbr_idx, SPIRVCrossDecorationInterfaceOrigID, var.self);
|
||||
set_extended_member_decoration(ib_type.self, ib_mbr_idx, SPIRVCrossDecorationInterfaceOrigID, var.self);
|
||||
}
|
||||
|
||||
void CompilerMSL::add_composite_variable_to_interface_block(StorageClass storage, const string &ib_var_ref,
|
||||
@ -2243,6 +2269,9 @@ void CompilerMSL::add_composite_variable_to_interface_block(StorageClass storage
|
||||
auto &var_type = meta.strip_array ? get_variable_element_type(var) : get_variable_data_type(var);
|
||||
uint32_t elem_cnt = 0;
|
||||
|
||||
if (add_component_variable_to_interface_block(storage, ib_var_ref, var, var_type, meta))
|
||||
return;
|
||||
|
||||
if (is_matrix(var_type))
|
||||
{
|
||||
if (is_array(var_type))
|
||||
@ -2339,6 +2368,7 @@ void CompilerMSL::add_composite_variable_to_interface_block(StorageClass storage
|
||||
if (get_decoration_bitset(var.self).get(DecorationLocation))
|
||||
{
|
||||
uint32_t locn = get_decoration(var.self, DecorationLocation) + i;
|
||||
uint32_t comp = get_decoration(var.self, DecorationComponent);
|
||||
if (storage == StorageClassInput)
|
||||
{
|
||||
var.basetype = ensure_correct_input_type(var.basetype, locn, 0, meta.strip_array);
|
||||
@ -2349,6 +2379,8 @@ void CompilerMSL::add_composite_variable_to_interface_block(StorageClass storage
|
||||
ib_type.member_types[ib_mbr_idx] = mbr_type_id;
|
||||
}
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationLocation, locn);
|
||||
if (comp)
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationComponent, comp);
|
||||
mark_location_as_used_by_shader(locn, *usable_type, storage);
|
||||
}
|
||||
else if (is_builtin && is_tessellation_shader() && inputs_by_builtin.count(builtin))
|
||||
@ -3319,7 +3351,6 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch)
|
||||
// Need to deal specially with DecorationComponent.
|
||||
// Multiple variables can alias the same Location, and try to make sure each location is declared only once.
|
||||
// We will swizzle data in and out to make this work.
|
||||
// We only need to consider plain variables here, not composites.
|
||||
// This is only relevant for vertex inputs and fragment outputs.
|
||||
// Technically tessellation as well, but it is too complicated to support.
|
||||
uint32_t component = get_decoration(var_id, DecorationComponent);
|
||||
@ -3329,8 +3360,22 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch)
|
||||
SPIRV_CROSS_THROW("Component decoration is not supported in tessellation shaders.");
|
||||
else if (pack_components)
|
||||
{
|
||||
auto &location_meta = meta.location_meta[location];
|
||||
location_meta.num_components = std::max(location_meta.num_components, component + type.vecsize);
|
||||
uint32_t array_size = 1;
|
||||
if (!type.array.empty())
|
||||
array_size = to_array_size_literal(type);
|
||||
|
||||
for (uint32_t location_offset = 0; location_offset < array_size; location_offset++)
|
||||
{
|
||||
auto &location_meta = meta.location_meta[location + location_offset];
|
||||
location_meta.num_components = std::max(location_meta.num_components, component + type.vecsize);
|
||||
|
||||
// For variables sharing location, decorations and base type must match.
|
||||
location_meta.base_type_id = type.self;
|
||||
location_meta.flat = has_decoration(var.self, DecorationFlat);
|
||||
location_meta.noperspective = has_decoration(var.self, DecorationNoPerspective);
|
||||
location_meta.centroid = has_decoration(var.self, DecorationCentroid);
|
||||
location_meta.sample = has_decoration(var.self, DecorationSample);
|
||||
}
|
||||
}
|
||||
}
|
||||
}
|
||||
@ -3579,6 +3624,31 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch)
|
||||
}
|
||||
}
|
||||
|
||||
// When multiple variables need to access same location,
|
||||
// unroll locations one by one and we will flatten output or input as necessary.
|
||||
for (auto &loc : meta.location_meta)
|
||||
{
|
||||
uint32_t location = loc.first;
|
||||
auto &location_meta = loc.second;
|
||||
|
||||
uint32_t ib_mbr_idx = uint32_t(ib_type.member_types.size());
|
||||
uint32_t type_id = build_extended_vector_type(location_meta.base_type_id, location_meta.num_components);
|
||||
ib_type.member_types.push_back(type_id);
|
||||
|
||||
set_member_name(ib_type.self, ib_mbr_idx, join("m_location_", location));
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationLocation, location);
|
||||
mark_location_as_used_by_shader(location, get<SPIRType>(type_id), storage);
|
||||
|
||||
if (location_meta.flat)
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationFlat);
|
||||
if (location_meta.noperspective)
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationNoPerspective);
|
||||
if (location_meta.centroid)
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationCentroid);
|
||||
if (location_meta.sample)
|
||||
set_member_decoration(ib_type.self, ib_mbr_idx, DecorationSample);
|
||||
}
|
||||
|
||||
// Sort the members of the structure by their locations.
|
||||
MemberSorter member_sorter(ib_type, ir.meta[ib_type_id], MemberSorter::LocationThenBuiltInType);
|
||||
member_sorter.sort();
|
||||
@ -15112,6 +15182,8 @@ bool CompilerMSL::MemberSorter::operator()(uint32_t mbr_idx1, uint32_t mbr_idx2)
|
||||
return mbr_meta2.builtin;
|
||||
else if (mbr_meta1.builtin)
|
||||
return mbr_meta1.builtin_type < mbr_meta2.builtin_type;
|
||||
else if (mbr_meta1.location == mbr_meta2.location)
|
||||
return mbr_meta1.component < mbr_meta2.component;
|
||||
else
|
||||
return mbr_meta1.location < mbr_meta2.location;
|
||||
}
|
||||
|
@ -799,8 +799,12 @@ protected:
|
||||
{
|
||||
struct LocationMeta
|
||||
{
|
||||
uint32_t base_type_id = 0;
|
||||
uint32_t num_components = 0;
|
||||
uint32_t ib_index = ~0u;
|
||||
bool flat = false;
|
||||
bool noperspective = false;
|
||||
bool centroid = false;
|
||||
bool sample = false;
|
||||
};
|
||||
std::unordered_map<uint32_t, LocationMeta> location_meta;
|
||||
bool strip_array = false;
|
||||
@ -815,6 +819,9 @@ protected:
|
||||
SPIRType &ib_type, SPIRVariable &var, InterfaceBlockMeta &meta);
|
||||
void add_plain_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref,
|
||||
SPIRType &ib_type, SPIRVariable &var, InterfaceBlockMeta &meta);
|
||||
bool add_component_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref,
|
||||
SPIRVariable &var, const SPIRType &type,
|
||||
InterfaceBlockMeta &meta);
|
||||
void add_plain_member_variable_to_interface_block(spv::StorageClass storage, const std::string &ib_var_ref,
|
||||
SPIRType &ib_type, SPIRVariable &var, uint32_t index,
|
||||
InterfaceBlockMeta &meta);
|
||||
|
Loading…
Reference in New Issue
Block a user