Merge pull request #413 from zeux/master
MSL: Order resources by type and binding index in the output
This commit is contained in:
Hans-Kristian Arntzen
GitHub
Коммит
4c1e57ee03
|
|
@ -14,7 +14,7 @@ struct _6
|
|||
float _m0[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_WorkGroupID [[threadgroup_position_in_grid]], device _6& _8 [[buffer(0)]], device _6& _9 [[buffer(1)]])
|
||||
kernel void main0(device _6& _8 [[buffer(0)]], device _6& _9 [[buffer(1)]], uint3 gl_WorkGroupID [[threadgroup_position_in_grid]])
|
||||
{
|
||||
_8._m0[gl_WorkGroupID.x] = _9._m0[gl_WorkGroupID.x] + _8._m0[gl_WorkGroupID.x];
|
||||
}
|
||||
|
|
|
|||
|
|
@ -101,7 +101,7 @@ struct main0_out
|
|||
float4 _entryPointOutput [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(main0_in in [[stage_in]], float4 gl_FragCoord [[position]], constant CB0& _19 [[buffer(0)]], texture3d<float> LightMapTexture [[texture(0)]], sampler LightMapSampler [[sampler(0)]], sampler ShadowMapSampler [[sampler(1)]], texture2d<float> ShadowMapTexture [[texture(1)]], texturecube<float> EnvironmentMapTexture [[texture(2)]], sampler EnvironmentMapSampler [[sampler(2)]], sampler DiffuseMapSampler [[sampler(3)]], texture2d<float> DiffuseMapTexture [[texture(3)]], sampler NormalMapSampler [[sampler(4)]], texture2d<float> NormalMapTexture [[texture(4)]], texture2d<float> NormalDetailMapTexture [[texture(5)]], sampler NormalDetailMapSampler [[sampler(5)]], texture2d<float> StudsMapTexture [[texture(6)]], sampler StudsMapSampler [[sampler(6)]], sampler SpecularMapSampler [[sampler(7)]], texture2d<float> SpecularMapTexture [[texture(7)]])
|
||||
fragment main0_out main0(main0_in in [[stage_in]], constant CB0& _19 [[buffer(0)]], texture3d<float> LightMapTexture [[texture(0)]], texture2d<float> ShadowMapTexture [[texture(1)]], texturecube<float> EnvironmentMapTexture [[texture(2)]], texture2d<float> DiffuseMapTexture [[texture(3)]], texture2d<float> NormalMapTexture [[texture(4)]], texture2d<float> NormalDetailMapTexture [[texture(5)]], texture2d<float> StudsMapTexture [[texture(6)]], texture2d<float> SpecularMapTexture [[texture(7)]], sampler LightMapSampler [[sampler(0)]], sampler ShadowMapSampler [[sampler(1)]], sampler EnvironmentMapSampler [[sampler(2)]], sampler DiffuseMapSampler [[sampler(3)]], sampler NormalMapSampler [[sampler(4)]], sampler NormalDetailMapSampler [[sampler(5)]], sampler StudsMapSampler [[sampler(6)]], sampler SpecularMapSampler [[sampler(7)]], float4 gl_FragCoord [[position]])
|
||||
{
|
||||
main0_out out = {};
|
||||
VertexOutput _128 = _121;
|
||||
|
|
|
|||
|
|
@ -95,7 +95,7 @@ struct main0_out
|
|||
float4 m_5 [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(float4 gl_FragCoord [[position]], constant _6& _7 [[buffer(0)]], texture2d<float> _8 [[texture(0)]], sampler _9 [[sampler(0)]], constant _10& _11 [[buffer(1)]], texture2d<float> _12 [[texture(1)]], sampler _13 [[sampler(1)]], texture2d<float> _14 [[texture(2)]], sampler _15 [[sampler(2)]], constant _18& _19 [[buffer(2)]])
|
||||
fragment main0_out main0(constant _6& _7 [[buffer(0)]], constant _10& _11 [[buffer(1)]], constant _18& _19 [[buffer(2)]], texture2d<float> _8 [[texture(0)]], texture2d<float> _12 [[texture(1)]], texture2d<float> _14 [[texture(2)]], sampler _9 [[sampler(0)]], sampler _13 [[sampler(1)]], sampler _15 [[sampler(2)]], float4 gl_FragCoord [[position]])
|
||||
{
|
||||
main0_out out = {};
|
||||
_28 _77 = _74;
|
||||
|
|
|
|||
|
|
@ -21,7 +21,7 @@ struct SSBO3
|
|||
uint counter;
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(0)]], device SSBO2& _45 [[buffer(1)]], device SSBO3& _48 [[buffer(2)]])
|
||||
kernel void main0(device SSBO& _23 [[buffer(0)]], device SSBO2& _45 [[buffer(1)]], device SSBO3& _48 [[buffer(2)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
float4 _29 = _23.in_data[gl_GlobalInvocationID.x];
|
||||
if (dot(_29, float4(1.0, 5.0, 6.0, 2.0)) > 8.19999980926513671875)
|
||||
|
|
|
|||
|
|
@ -23,7 +23,7 @@ struct SSBO3
|
|||
uint count;
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _22 [[buffer(0)]], device SSBO2& _38 [[buffer(1)]], device SSBO3& _41 [[buffer(2)]])
|
||||
kernel void main0(device SSBO& _22 [[buffer(0)]], device SSBO2& _38 [[buffer(1)]], device SSBO3& _41 [[buffer(2)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
float _28 = _22.in_data[gl_GlobalInvocationID.x];
|
||||
if (_28 > 12.0)
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ struct SSBO2
|
|||
float4 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _28 [[buffer(0)]], device SSBO2& _52 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _28 [[buffer(0)]], device SSBO2& _52 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
int i = 0;
|
||||
float4 _56;
|
||||
|
|
|
|||
|
|
@ -3,7 +3,7 @@
|
|||
|
||||
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)]])
|
||||
kernel void main0(texture2d<float> uImageIn [[texture(0)]], texture2d<float, access::write> uImageOut [[texture(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uImageOut.write(uImageIn.read(uint2((int2(gl_GlobalInvocationID.xy) + int2(uImageIn.get_width(), uImageIn.get_height())))), uint2(int2(gl_GlobalInvocationID.xy)));
|
||||
}
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ struct SSBO2
|
|||
float4 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _24 [[buffer(0)]], device SSBO2& _177 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _24 [[buffer(0)]], device SSBO2& _177 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float4 idat = _24.in_data[ident];
|
||||
|
|
|
|||
|
|
@ -8,7 +8,7 @@ struct SSBO2
|
|||
float3x3 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO2& _22 [[buffer(0)]])
|
||||
kernel void main0(device SSBO2& _22 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
_22.out_data[gl_GlobalInvocationID.x] = float3x3(float3(10.0), float3(20.0), float3(40.0));
|
||||
}
|
||||
|
|
|
|||
|
|
@ -22,7 +22,7 @@ Tx mod(Tx x, Ty y)
|
|||
return x - y * floor(x / y);
|
||||
}
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(0)]], device SSBO2& _33 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _23 [[buffer(0)]], device SSBO2& _33 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
_33.out_data[gl_GlobalInvocationID.x] = mod(_23.in_data[gl_GlobalInvocationID.x], _33.out_data[gl_GlobalInvocationID.x]);
|
||||
_33.out_data[gl_GlobalInvocationID.x] = as_type<float4>(as_type<uint4>(_23.in_data[gl_GlobalInvocationID.x]) % as_type<uint4>(_33.out_data[gl_GlobalInvocationID.x]));
|
||||
|
|
|
|||
|
|
@ -13,7 +13,7 @@ struct SSBO2
|
|||
float4 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(0)]], device SSBO2& _35 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _23 [[buffer(0)]], device SSBO2& _35 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
float4 i;
|
||||
float4 _31 = modf(_23.in_data[gl_GlobalInvocationID.x], i);
|
||||
|
|
|
|||
|
|
@ -10,7 +10,7 @@ struct SSBO2
|
|||
|
||||
constant int _69 = {};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO2& _27 [[buffer(0)]])
|
||||
kernel void main0(device SSBO2& _27 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
if (gl_GlobalInvocationID.x == 2u)
|
||||
{
|
||||
|
|
|
|||
|
|
@ -15,7 +15,7 @@ struct SSBO2
|
|||
float out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _22 [[buffer(0)]], uint gl_LocalInvocationIndex [[thread_index_in_threadgroup]], device SSBO2& _44 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _22 [[buffer(0)]], device SSBO2& _44 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], uint gl_LocalInvocationIndex [[thread_index_in_threadgroup]])
|
||||
{
|
||||
threadgroup float sShared[4];
|
||||
sShared[gl_LocalInvocationIndex] = _22.in_data[gl_GlobalInvocationID.x];
|
||||
|
|
|
|||
|
|
@ -18,7 +18,7 @@ struct SSBO
|
|||
Foo in_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO2& _23 [[buffer(0)]], device SSBO& _30 [[buffer(1)]])
|
||||
kernel void main0(device SSBO2& _23 [[buffer(0)]], device SSBO& _30 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
_23.out_data[gl_GlobalInvocationID.x].m = _30.in_data[gl_GlobalInvocationID.x].m * _30.in_data[gl_GlobalInvocationID.x].m;
|
||||
}
|
||||
|
|
|
|||
|
|
@ -16,7 +16,7 @@ struct SSBO2
|
|||
|
||||
constant uint _98 = {};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _24 [[buffer(0)]], device SSBO2& _89 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _24 [[buffer(0)]], device SSBO2& _89 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
float4 _93;
|
||||
int _94;
|
||||
|
|
|
|||
|
|
@ -28,7 +28,7 @@ struct SSBO2
|
|||
float4 outputs[1];
|
||||
};
|
||||
|
||||
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)]])
|
||||
kernel void main0(device SSBO0& _36 [[buffer(0)]], device SSBO1& _55 [[buffer(1)]], device SSBO2& _66 [[buffer(2)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
_66.outputs[gl_GlobalInvocationID.x] = _36.s0s[gl_GlobalInvocationID.x].a + _55.s1s[gl_GlobalInvocationID.x].a;
|
||||
}
|
||||
|
|
|
|||
|
|
@ -13,7 +13,7 @@ struct SSBO
|
|||
uint inputs[1];
|
||||
};
|
||||
|
||||
kernel void main0(device SSBO2& _10 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(1)]])
|
||||
kernel void main0(device SSBO2& _10 [[buffer(0)]], device SSBO& _23 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
_10.outputs[gl_GlobalInvocationID.x] = _23.inputs[gl_GlobalInvocationID.x] / 29u;
|
||||
}
|
||||
|
|
|
|||
|
|
@ -8,7 +8,7 @@ struct main0_out
|
|||
float4 FragColor [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(texture2d_ms<float> uSampler [[texture(0)]], sampler uSamplerSmplr [[sampler(0)]], texture2d_ms<float> uSamplerArray [[texture(1)]], sampler uSamplerArraySmplr [[sampler(1)]], texture2d_ms<float> uImage [[texture(2)]], texture2d_ms<float> uImageArray [[texture(3)]])
|
||||
fragment main0_out main0(texture2d_ms<float> uSampler [[texture(0)]], texture2d_ms<float> uSamplerArray [[texture(1)]], texture2d_ms<float> uImage [[texture(2)]], texture2d_ms<float> uImageArray [[texture(3)]], sampler uSamplerSmplr [[sampler(0)]], sampler uSamplerArraySmplr [[sampler(1)]])
|
||||
{
|
||||
main0_out out = {};
|
||||
out.FragColor = float4(float(((int(uSampler.get_num_samples()) + int(uSamplerArray.get_num_samples())) + int(uImage.get_num_samples())) + int(uImageArray.get_num_samples())));
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ struct main0_out
|
|||
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)]])
|
||||
fragment main0_out main0(main0_in in [[stage_in]], texture2d<float> TextureBase [[texture(0)]], texture2d<float> TextureDetail [[texture(1)]], sampler TextureBaseSmplr [[sampler(0)]], sampler TextureDetailSmplr [[sampler(1)]])
|
||||
{
|
||||
main0_out out = {};
|
||||
float4 _20 = TextureBase.sample(TextureBaseSmplr, in.VertGeom.xy);
|
||||
|
|
|
|||
|
|
@ -8,7 +8,7 @@ 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)]])
|
||||
fragment main0_out main0(texture2d_ms<float> uSampler [[texture(0)]], sampler uSamplerSmplr [[sampler(0)]], float4 gl_FragCoord [[position]])
|
||||
{
|
||||
main0_out out = {};
|
||||
int2 _17 = int2(gl_FragCoord.xy);
|
||||
|
|
|
|||
|
|
@ -15,7 +15,7 @@ struct main0_out
|
|||
float FragColor [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(main0_in in [[stage_in]], depth2d<float> uShadow2D [[texture(0)]], sampler uShadow2DSmplr [[sampler(0)]], texture1d<float> uSampler1D [[texture(1)]], sampler uSampler1DSmplr [[sampler(1)]], texture2d<float> uSampler2D [[texture(2)]], sampler uSampler2DSmplr [[sampler(2)]], texture3d<float> uSampler3D [[texture(3)]], sampler uSampler3DSmplr [[sampler(3)]])
|
||||
fragment main0_out main0(main0_in in [[stage_in]], depth2d<float> uShadow2D [[texture(0)]], texture1d<float> uSampler1D [[texture(1)]], texture2d<float> uSampler2D [[texture(2)]], texture3d<float> uSampler3D [[texture(3)]], sampler uShadow2DSmplr [[sampler(0)]], sampler uSampler1DSmplr [[sampler(1)]], sampler uSampler2DSmplr [[sampler(2)]], sampler uSampler3DSmplr [[sampler(3)]])
|
||||
{
|
||||
main0_out out = {};
|
||||
float4 _20 = in.vClip4;
|
||||
|
|
|
|||
|
|
@ -167,7 +167,7 @@ void vs_adjust(thread float4& dst_reg0, thread float4& dst_reg1, thread float4&
|
|||
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)]])
|
||||
vertex main0_out main0(constant VertexBuffer& v_227 [[buffer(0)]], constant VertexConstantsBuffer& v_309 [[buffer(1)]], texture2d<uint> buff_in_2 [[texture(0)]], texture2d<uint> buff_in_1 [[texture(1)]], uint gl_VertexIndex [[vertex_id]])
|
||||
{
|
||||
main0_out out = {};
|
||||
float4 dst_reg0 = float4(0.0, 0.0, 0.0, 1.0);
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ struct _6
|
|||
float _m0[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_WorkGroupID [[threadgroup_position_in_grid]], device _6& _8 [[buffer(0)]], device _6& _9 [[buffer(1)]])
|
||||
kernel void main0(device _6& _8 [[buffer(0)]], device _6& _9 [[buffer(1)]], uint3 gl_WorkGroupID [[threadgroup_position_in_grid]])
|
||||
{
|
||||
uint3 _23 = gl_WorkGroupSize;
|
||||
_8._m0[gl_WorkGroupID.x] = _9._m0[gl_WorkGroupID.x] + _8._m0[gl_WorkGroupID.x];
|
||||
|
|
|
|||
|
|
@ -101,7 +101,7 @@ struct main0_out
|
|||
float4 _entryPointOutput [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(main0_in in [[stage_in]], float4 gl_FragCoord [[position]], constant CB0& _19 [[buffer(0)]], texture3d<float> LightMapTexture [[texture(0)]], sampler LightMapSampler [[sampler(0)]], sampler ShadowMapSampler [[sampler(1)]], texture2d<float> ShadowMapTexture [[texture(1)]], texturecube<float> EnvironmentMapTexture [[texture(2)]], sampler EnvironmentMapSampler [[sampler(2)]], sampler DiffuseMapSampler [[sampler(3)]], texture2d<float> DiffuseMapTexture [[texture(3)]], sampler NormalMapSampler [[sampler(4)]], texture2d<float> NormalMapTexture [[texture(4)]], texture2d<float> NormalDetailMapTexture [[texture(5)]], sampler NormalDetailMapSampler [[sampler(5)]], texture2d<float> StudsMapTexture [[texture(6)]], sampler StudsMapSampler [[sampler(6)]], sampler SpecularMapSampler [[sampler(7)]], texture2d<float> SpecularMapTexture [[texture(7)]])
|
||||
fragment main0_out main0(main0_in in [[stage_in]], constant CB0& _19 [[buffer(0)]], texture3d<float> LightMapTexture [[texture(0)]], texture2d<float> ShadowMapTexture [[texture(1)]], texturecube<float> EnvironmentMapTexture [[texture(2)]], texture2d<float> DiffuseMapTexture [[texture(3)]], texture2d<float> NormalMapTexture [[texture(4)]], texture2d<float> NormalDetailMapTexture [[texture(5)]], texture2d<float> StudsMapTexture [[texture(6)]], texture2d<float> SpecularMapTexture [[texture(7)]], sampler LightMapSampler [[sampler(0)]], sampler ShadowMapSampler [[sampler(1)]], sampler EnvironmentMapSampler [[sampler(2)]], sampler DiffuseMapSampler [[sampler(3)]], sampler NormalMapSampler [[sampler(4)]], sampler NormalDetailMapSampler [[sampler(5)]], sampler StudsMapSampler [[sampler(6)]], sampler SpecularMapSampler [[sampler(7)]], float4 gl_FragCoord [[position]])
|
||||
{
|
||||
main0_out out = {};
|
||||
VertexOutput _128 = _121;
|
||||
|
|
|
|||
|
|
@ -95,7 +95,7 @@ struct main0_out
|
|||
float4 m_5 [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(float4 gl_FragCoord [[position]], constant _6& _7 [[buffer(0)]], texture2d<float> _8 [[texture(0)]], sampler _9 [[sampler(0)]], constant _10& _11 [[buffer(1)]], texture2d<float> _12 [[texture(1)]], sampler _13 [[sampler(1)]], texture2d<float> _14 [[texture(2)]], sampler _15 [[sampler(2)]], constant _18& _19 [[buffer(2)]])
|
||||
fragment main0_out main0(constant _6& _7 [[buffer(0)]], constant _10& _11 [[buffer(1)]], constant _18& _19 [[buffer(2)]], texture2d<float> _8 [[texture(0)]], texture2d<float> _12 [[texture(1)]], texture2d<float> _14 [[texture(2)]], sampler _9 [[sampler(0)]], sampler _13 [[sampler(1)]], sampler _15 [[sampler(2)]], float4 gl_FragCoord [[position]])
|
||||
{
|
||||
main0_out out = {};
|
||||
_28 _77 = _74;
|
||||
|
|
|
|||
|
|
@ -21,7 +21,7 @@ struct SSBO3
|
|||
uint counter;
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(0)]], device SSBO2& _45 [[buffer(1)]], device SSBO3& _48 [[buffer(2)]])
|
||||
kernel void main0(device SSBO& _23 [[buffer(0)]], device SSBO2& _45 [[buffer(1)]], device SSBO3& _48 [[buffer(2)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float4 idata = _23.in_data[ident];
|
||||
|
|
|
|||
|
|
@ -23,7 +23,7 @@ struct SSBO3
|
|||
uint count;
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _22 [[buffer(0)]], device SSBO2& _38 [[buffer(1)]], device SSBO3& _41 [[buffer(2)]])
|
||||
kernel void main0(device SSBO& _22 [[buffer(0)]], device SSBO2& _38 [[buffer(1)]], device SSBO3& _41 [[buffer(2)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float idata = _22.in_data[ident];
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ struct SSBO2
|
|||
float4 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _28 [[buffer(0)]], device SSBO2& _52 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _28 [[buffer(0)]], device SSBO2& _52 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
int i = 0;
|
||||
|
|
|
|||
|
|
@ -3,7 +3,7 @@
|
|||
|
||||
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)]])
|
||||
kernel void main0(texture2d<float> uImageIn [[texture(0)]], texture2d<float, access::write> uImageOut [[texture(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
float4 v = uImageIn.read(uint2((int2(gl_GlobalInvocationID.xy) + int2(uImageIn.get_width(), uImageIn.get_height()))));
|
||||
uImageOut.write(v, uint2(int2(gl_GlobalInvocationID.xy)));
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ struct SSBO2
|
|||
float4 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _24 [[buffer(0)]], device SSBO2& _177 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _24 [[buffer(0)]], device SSBO2& _177 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float4 idat = _24.in_data[ident];
|
||||
|
|
|
|||
|
|
@ -8,7 +8,7 @@ struct SSBO2
|
|||
float3x3 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO2& _22 [[buffer(0)]])
|
||||
kernel void main0(device SSBO2& _22 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
_22.out_data[ident] = float3x3(float3(10.0), float3(20.0), float3(40.0));
|
||||
|
|
|
|||
|
|
@ -22,7 +22,7 @@ Tx mod(Tx x, Ty y)
|
|||
return x - y * floor(x / y);
|
||||
}
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(0)]], device SSBO2& _33 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _23 [[buffer(0)]], device SSBO2& _33 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float4 v = mod(_23.in_data[ident], _33.out_data[ident]);
|
||||
|
|
|
|||
|
|
@ -13,7 +13,7 @@ struct SSBO2
|
|||
float4 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(0)]], device SSBO2& _35 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _23 [[buffer(0)]], device SSBO2& _35 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float4 i;
|
||||
|
|
|
|||
|
|
@ -8,7 +8,7 @@ struct SSBO2
|
|||
float4 out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO2& _27 [[buffer(0)]])
|
||||
kernel void main0(device SSBO2& _27 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
if (ident == 2u)
|
||||
|
|
|
|||
|
|
@ -15,7 +15,7 @@ struct SSBO2
|
|||
float out_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _22 [[buffer(0)]], uint gl_LocalInvocationIndex [[thread_index_in_threadgroup]], device SSBO2& _44 [[buffer(1)]])
|
||||
kernel void main0(device SSBO& _22 [[buffer(0)]], device SSBO2& _44 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], uint gl_LocalInvocationIndex [[thread_index_in_threadgroup]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float idata = _22.in_data[ident];
|
||||
|
|
|
|||
|
|
@ -18,7 +18,7 @@ struct SSBO
|
|||
Foo in_data[1];
|
||||
};
|
||||
|
||||
kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO2& _23 [[buffer(0)]], device SSBO& _30 [[buffer(1)]])
|
||||
kernel void main0(device SSBO2& _23 [[buffer(0)]], device SSBO& _30 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
_23.out_data[ident].m = _30.in_data[ident].m * _30.in_data[ident].m;
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ 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)]])
|
||||
kernel void main0(device SSBO& _24 [[buffer(0)]], device SSBO2& _89 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
uint ident = gl_GlobalInvocationID.x;
|
||||
float4 idat = _24.in_data[ident];
|
||||
|
|
|
|||
|
|
@ -40,7 +40,7 @@ float4 overload(thread const S1& s1)
|
|||
return s1.a;
|
||||
}
|
||||
|
||||
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)]])
|
||||
kernel void main0(device SSBO0& _36 [[buffer(0)]], device SSBO1& _55 [[buffer(1)]], device SSBO2& _66 [[buffer(2)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
S0 s0;
|
||||
s0.a = _36.s0s[gl_GlobalInvocationID.x].a;
|
||||
|
|
|
|||
|
|
@ -13,7 +13,7 @@ struct SSBO
|
|||
uint inputs[1];
|
||||
};
|
||||
|
||||
kernel void main0(device SSBO2& _10 [[buffer(0)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device SSBO& _23 [[buffer(1)]])
|
||||
kernel void main0(device SSBO2& _10 [[buffer(0)]], device SSBO& _23 [[buffer(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
||||
{
|
||||
_10.outputs[gl_GlobalInvocationID.x] = _23.inputs[gl_GlobalInvocationID.x] / 29u;
|
||||
}
|
||||
|
|
|
|||
|
|
@ -8,7 +8,7 @@ struct main0_out
|
|||
float4 FragColor [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(texture2d_ms<float> uSampler [[texture(0)]], sampler uSamplerSmplr [[sampler(0)]], texture2d_ms<float> uSamplerArray [[texture(1)]], sampler uSamplerArraySmplr [[sampler(1)]], texture2d_ms<float> uImage [[texture(2)]], texture2d_ms<float> uImageArray [[texture(3)]])
|
||||
fragment main0_out main0(texture2d_ms<float> uSampler [[texture(0)]], texture2d_ms<float> uSamplerArray [[texture(1)]], texture2d_ms<float> uImage [[texture(2)]], texture2d_ms<float> uImageArray [[texture(3)]], sampler uSamplerSmplr [[sampler(0)]], sampler uSamplerArraySmplr [[sampler(1)]])
|
||||
{
|
||||
main0_out out = {};
|
||||
out.FragColor = float4(float(((int(uSampler.get_num_samples()) + int(uSamplerArray.get_num_samples())) + int(uImage.get_num_samples())) + int(uImageArray.get_num_samples())));
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ struct main0_out
|
|||
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)]])
|
||||
fragment main0_out main0(main0_in in [[stage_in]], texture2d<float> TextureBase [[texture(0)]], texture2d<float> TextureDetail [[texture(1)]], sampler TextureBaseSmplr [[sampler(0)]], sampler TextureDetailSmplr [[sampler(1)]])
|
||||
{
|
||||
main0_out out = {};
|
||||
float4 texSample0 = TextureBase.sample(TextureBaseSmplr, in.VertGeom.xy);
|
||||
|
|
|
|||
|
|
@ -8,7 +8,7 @@ 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)]])
|
||||
fragment main0_out main0(texture2d_ms<float> uSampler [[texture(0)]], sampler uSamplerSmplr [[sampler(0)]], float4 gl_FragCoord [[position]])
|
||||
{
|
||||
main0_out out = {};
|
||||
int2 coord = int2(gl_FragCoord.xy);
|
||||
|
|
|
|||
|
|
@ -15,7 +15,7 @@ struct main0_out
|
|||
float FragColor [[color(0)]];
|
||||
};
|
||||
|
||||
fragment main0_out main0(main0_in in [[stage_in]], depth2d<float> uShadow2D [[texture(0)]], sampler uShadow2DSmplr [[sampler(0)]], texture1d<float> uSampler1D [[texture(1)]], sampler uSampler1DSmplr [[sampler(1)]], texture2d<float> uSampler2D [[texture(2)]], sampler uSampler2DSmplr [[sampler(2)]], texture3d<float> uSampler3D [[texture(3)]], sampler uSampler3DSmplr [[sampler(3)]])
|
||||
fragment main0_out main0(main0_in in [[stage_in]], depth2d<float> uShadow2D [[texture(0)]], texture1d<float> uSampler1D [[texture(1)]], texture2d<float> uSampler2D [[texture(2)]], texture3d<float> uSampler3D [[texture(3)]], sampler uShadow2DSmplr [[sampler(0)]], sampler uSampler1DSmplr [[sampler(1)]], sampler uSampler2DSmplr [[sampler(2)]], sampler uSampler3DSmplr [[sampler(3)]])
|
||||
{
|
||||
main0_out out = {};
|
||||
float4 _20 = in.vClip4;
|
||||
|
|
|
|||
|
|
@ -2692,7 +2692,19 @@ string CompilerMSL::entry_point_args(bool append_comma)
|
|||
convert_to_string(nsi_var.first) + ")]]";
|
||||
}
|
||||
|
||||
// Uniforms
|
||||
// Output resources, sorted by resource index & type
|
||||
// We need to sort to work around a bug on macOS 10.13 with NVidia drivers where switching between shaders
|
||||
// with different order of buffers can result in issues with buffer assignments inside the driver.
|
||||
struct Resource
|
||||
{
|
||||
Variant *id;
|
||||
string name;
|
||||
SPIRType::BaseType basetype;
|
||||
uint32_t index;
|
||||
};
|
||||
|
||||
vector<Resource> resources;
|
||||
|
||||
for (auto &id : ids)
|
||||
{
|
||||
if (id.get_type() == TypeVariable)
|
||||
|
|
@ -2706,7 +2718,36 @@ string CompilerMSL::entry_point_args(bool append_comma)
|
|||
var.storage == StorageClassPushConstant || var.storage == StorageClassStorageBuffer) &&
|
||||
!is_hidden_variable(var))
|
||||
{
|
||||
switch (type.basetype)
|
||||
if (type.basetype == SPIRType::SampledImage)
|
||||
{
|
||||
resources.push_back(
|
||||
{ &id, to_name(var_id), SPIRType::Image, get_metal_resource_index(var, SPIRType::Image) });
|
||||
|
||||
if (type.image.dim != DimBuffer)
|
||||
resources.push_back({ &id, to_sampler_expression(var_id), SPIRType::Sampler,
|
||||
get_metal_resource_index(var, SPIRType::Sampler) });
|
||||
}
|
||||
else
|
||||
{
|
||||
resources.push_back(
|
||||
{ &id, to_name(var_id), type.basetype, get_metal_resource_index(var, type.basetype) });
|
||||
}
|
||||
}
|
||||
}
|
||||
}
|
||||
|
||||
std::sort(resources.begin(), resources.end(), [](const Resource &lhs, const Resource &rhs) {
|
||||
return tie(lhs.basetype, lhs.index) < tie(rhs.basetype, rhs.index);
|
||||
});
|
||||
|
||||
for (auto &r : resources)
|
||||
{
|
||||
auto &var = r.id->get<SPIRVariable>();
|
||||
auto &type = get<SPIRType>(var.basetype);
|
||||
|
||||
uint32_t var_id = var.self;
|
||||
|
||||
switch (r.basetype)
|
||||
{
|
||||
case SPIRType::Struct:
|
||||
{
|
||||
|
|
@ -2715,39 +2756,37 @@ string CompilerMSL::entry_point_args(bool append_comma)
|
|||
break;
|
||||
if (!ep_args.empty())
|
||||
ep_args += ", ";
|
||||
ep_args += get_argument_address_space(var) + " " + type_to_glsl(type) + "& " + to_name(var_id);
|
||||
ep_args += " [[buffer(" + convert_to_string(get_metal_resource_index(var, type.basetype)) + ")]]";
|
||||
ep_args += get_argument_address_space(var) + " " + type_to_glsl(type) + "& " + r.name;
|
||||
ep_args += " [[buffer(" + convert_to_string(r.index) + ")]]";
|
||||
break;
|
||||
}
|
||||
case SPIRType::Sampler:
|
||||
if (!ep_args.empty())
|
||||
ep_args += ", ";
|
||||
ep_args += type_to_glsl(type) + " " + to_name(var_id);
|
||||
ep_args += " [[sampler(" + convert_to_string(get_metal_resource_index(var, type.basetype)) + ")]]";
|
||||
ep_args += "sampler " + r.name;
|
||||
ep_args += " [[sampler(" + convert_to_string(r.index) + ")]]";
|
||||
break;
|
||||
case SPIRType::Image:
|
||||
if (!ep_args.empty())
|
||||
ep_args += ", ";
|
||||
ep_args += type_to_glsl(type, var_id) + " " + to_name(var_id);
|
||||
ep_args += " [[texture(" + convert_to_string(get_metal_resource_index(var, type.basetype)) + ")]]";
|
||||
break;
|
||||
case SPIRType::SampledImage:
|
||||
if (!ep_args.empty())
|
||||
ep_args += ", ";
|
||||
ep_args += type_to_glsl(type, var_id) + " " + to_name(var_id);
|
||||
ep_args +=
|
||||
" [[texture(" + convert_to_string(get_metal_resource_index(var, SPIRType::Image)) + ")]]";
|
||||
if (type.image.dim != DimBuffer)
|
||||
{
|
||||
ep_args += ", sampler " + to_sampler_expression(var_id);
|
||||
ep_args +=
|
||||
" [[sampler(" + convert_to_string(get_metal_resource_index(var, SPIRType::Sampler)) + ")]]";
|
||||
}
|
||||
ep_args += type_to_glsl(type, var_id) + " " + r.name;
|
||||
ep_args += " [[texture(" + convert_to_string(r.index) + ")]]";
|
||||
break;
|
||||
default:
|
||||
SPIRV_CROSS_THROW("Unexpected resource type");
|
||||
break;
|
||||
}
|
||||
}
|
||||
|
||||
// Builtin variables
|
||||
for (auto &id : ids)
|
||||
{
|
||||
if (id.get_type() == TypeVariable)
|
||||
{
|
||||
auto &var = id.get<SPIRVariable>();
|
||||
|
||||
uint32_t var_id = var.self;
|
||||
|
||||
if (var.storage == StorageClassInput && is_builtin_variable(var))
|
||||
{
|
||||
if (!ep_args.empty())
|
||||
|
|
|
|||
Загрузка…
Ссылка в новой задаче