| #include <metal_stdlib> |
| |
| using namespace metal; |
| struct tint_symbol_6 { |
| float3x2 w; |
| }; |
| |
| void tint_zero_workgroup_memory(uint local_idx, threadgroup float3x2* const tint_symbol) { |
| if ((local_idx < 1u)) { |
| *(tint_symbol) = float3x2(float2(0.0f), float2(0.0f), float2(0.0f)); |
| } |
| threadgroup_barrier(mem_flags::mem_threadgroup); |
| } |
| |
| void f_inner(uint local_invocation_index, threadgroup float3x2* const tint_symbol_1, const constant float3x2* const tint_symbol_2) { |
| tint_zero_workgroup_memory(local_invocation_index, tint_symbol_1); |
| *(tint_symbol_1) = *(tint_symbol_2); |
| (*(tint_symbol_1))[1] = (*(tint_symbol_2))[0]; |
| (*(tint_symbol_1))[1] = (*(tint_symbol_2))[0].yx; |
| (*(tint_symbol_1))[0][1] = (*(tint_symbol_2))[1][0]; |
| } |
| |
| kernel void f(const constant float3x2* tint_symbol_5 [[buffer(0)]], threadgroup tint_symbol_6* tint_symbol_4 [[threadgroup(0)]], uint local_invocation_index [[thread_index_in_threadgroup]]) { |
| threadgroup float3x2* const tint_symbol_3 = &((*(tint_symbol_4)).w); |
| f_inner(local_invocation_index, tint_symbol_3, tint_symbol_5); |
| return; |
| } |
| |