91 lines
3.2 KiB
Plaintext
91 lines
3.2 KiB
Plaintext
#include <metal_stdlib>
|
|
#include <simd/simd.h>
|
|
|
|
using namespace metal;
|
|
|
|
struct Baz
|
|
{
|
|
int e;
|
|
int f;
|
|
};
|
|
|
|
struct Foo
|
|
{
|
|
int a;
|
|
int b;
|
|
};
|
|
|
|
struct Bar
|
|
{
|
|
int c;
|
|
int d;
|
|
};
|
|
|
|
constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(3u, 3u, 2u);
|
|
|
|
struct spvDescriptorSetBuffer0
|
|
{
|
|
constant Foo* m_34 [[id(0)]];
|
|
constant Bar* m_40 [[id(1)]];
|
|
};
|
|
|
|
struct spvDescriptorSetBuffer1
|
|
{
|
|
device Baz* baz [[id(0)]][3][3][2];
|
|
};
|
|
|
|
kernel void main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], constant spvDescriptorSetBuffer1& spvDescriptorSet1 [[buffer(1)]], constant uint* spvDynamicOffsets [[buffer(23)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
|
|
{
|
|
constant auto& _34 = *(constant Foo* )((constant char* )spvDescriptorSet0.m_34 + spvDynamicOffsets[0]);
|
|
device Baz* baz[3][3][2] =
|
|
{
|
|
{
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[0][0][0] + spvDynamicOffsets[1]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[0][0][1] + spvDynamicOffsets[2]),
|
|
},
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[0][1][0] + spvDynamicOffsets[3]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[0][1][1] + spvDynamicOffsets[4]),
|
|
},
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[0][2][0] + spvDynamicOffsets[5]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[0][2][1] + spvDynamicOffsets[6]),
|
|
},
|
|
},
|
|
{
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[1][0][0] + spvDynamicOffsets[7]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[1][0][1] + spvDynamicOffsets[8]),
|
|
},
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[1][1][0] + spvDynamicOffsets[9]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[1][1][1] + spvDynamicOffsets[10]),
|
|
},
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[1][2][0] + spvDynamicOffsets[11]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[1][2][1] + spvDynamicOffsets[12]),
|
|
},
|
|
},
|
|
{
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[2][0][0] + spvDynamicOffsets[13]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[2][0][1] + spvDynamicOffsets[14]),
|
|
},
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[2][1][0] + spvDynamicOffsets[15]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[2][1][1] + spvDynamicOffsets[16]),
|
|
},
|
|
{
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[2][2][0] + spvDynamicOffsets[17]),
|
|
(device Baz* )((device char* )spvDescriptorSet1.baz[2][2][1] + spvDynamicOffsets[18]),
|
|
},
|
|
},
|
|
};
|
|
|
|
uint3 coords = gl_GlobalInvocationID;
|
|
baz[coords.x][coords.y][coords.z]->e = _34.a + (*spvDescriptorSet0.m_40).c;
|
|
baz[coords.x][coords.y][coords.z]->f = _34.b * (*spvDescriptorSet0.m_40).d;
|
|
}
|
|
|