• Home
  • Line#
  • Scopes#
  • Navigate#
  • Raw
  • Download
1#include <metal_stdlib>
2#include <simd/simd.h>
3
4using namespace metal;
5
6struct Baz
7{
8    int e;
9    int f;
10};
11
12struct Foo
13{
14    int a;
15    int b;
16};
17
18struct Bar
19{
20    int c;
21    int d;
22};
23
24constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(3u, 3u, 2u);
25
26struct spvDescriptorSetBuffer0
27{
28    constant Foo* m_34 [[id(0)]];
29    constant Bar* m_40 [[id(1)]];
30};
31
32struct spvDescriptorSetBuffer1
33{
34    device Baz* baz [[id(0)]][3][3][2];
35};
36
37kernel void main0(constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]], constant spvDescriptorSetBuffer1& spvDescriptorSet1 [[buffer(1)]], constant uint* spvDynamicOffsets [[buffer(23)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
38{
39    constant auto& _34 = *(constant Foo* )((constant char* )spvDescriptorSet0.m_34 + spvDynamicOffsets[0]);
40    device Baz* baz[3][3][2] =
41    {
42        {
43            {
44                (device Baz* )((device char* )spvDescriptorSet1.baz[0][0][0] + spvDynamicOffsets[1]),
45                (device Baz* )((device char* )spvDescriptorSet1.baz[0][0][1] + spvDynamicOffsets[2]),
46            },
47            {
48                (device Baz* )((device char* )spvDescriptorSet1.baz[0][1][0] + spvDynamicOffsets[3]),
49                (device Baz* )((device char* )spvDescriptorSet1.baz[0][1][1] + spvDynamicOffsets[4]),
50            },
51            {
52                (device Baz* )((device char* )spvDescriptorSet1.baz[0][2][0] + spvDynamicOffsets[5]),
53                (device Baz* )((device char* )spvDescriptorSet1.baz[0][2][1] + spvDynamicOffsets[6]),
54            },
55        },
56        {
57            {
58                (device Baz* )((device char* )spvDescriptorSet1.baz[1][0][0] + spvDynamicOffsets[7]),
59                (device Baz* )((device char* )spvDescriptorSet1.baz[1][0][1] + spvDynamicOffsets[8]),
60            },
61            {
62                (device Baz* )((device char* )spvDescriptorSet1.baz[1][1][0] + spvDynamicOffsets[9]),
63                (device Baz* )((device char* )spvDescriptorSet1.baz[1][1][1] + spvDynamicOffsets[10]),
64            },
65            {
66                (device Baz* )((device char* )spvDescriptorSet1.baz[1][2][0] + spvDynamicOffsets[11]),
67                (device Baz* )((device char* )spvDescriptorSet1.baz[1][2][1] + spvDynamicOffsets[12]),
68            },
69        },
70        {
71            {
72                (device Baz* )((device char* )spvDescriptorSet1.baz[2][0][0] + spvDynamicOffsets[13]),
73                (device Baz* )((device char* )spvDescriptorSet1.baz[2][0][1] + spvDynamicOffsets[14]),
74            },
75            {
76                (device Baz* )((device char* )spvDescriptorSet1.baz[2][1][0] + spvDynamicOffsets[15]),
77                (device Baz* )((device char* )spvDescriptorSet1.baz[2][1][1] + spvDynamicOffsets[16]),
78            },
79            {
80                (device Baz* )((device char* )spvDescriptorSet1.baz[2][2][0] + spvDynamicOffsets[17]),
81                (device Baz* )((device char* )spvDescriptorSet1.baz[2][2][1] + spvDynamicOffsets[18]),
82            },
83        },
84    };
85
86    uint3 coords = gl_GlobalInvocationID;
87    baz[coords.x][coords.y][coords.z]->e = _34.a + (*spvDescriptorSet0.m_40).c;
88    baz[coords.x][coords.y][coords.z]->f = _34.b * (*spvDescriptorSet0.m_40).d;
89}
90
91