| #pragma clang diagnostic ignored "-Wmissing-prototypes" |
| |
| #include <metal_stdlib> |
| #include <simd/simd.h> |
| |
| using namespace metal; |
| |
| struct SSBOCol |
| { |
| float3x2 col_major0; |
| float3x2 col_major1; |
| }; |
| |
| struct SSBORow |
| { |
| float2x3 row_major0; |
| float2x3 row_major1; |
| }; |
| |
| constant uint3 gl_WorkGroupSize [[maybe_unused]] = uint3(1u); |
| |
| static inline __attribute__((always_inline)) |
| void load_store_to_variable_col_major(device SSBOCol& v_29) |
| { |
| float3x2 loaded = v_29.col_major0; |
| v_29.col_major1 = loaded; |
| } |
| |
| static inline __attribute__((always_inline)) |
| void load_store_to_variable_row_major(device SSBORow& v_41) |
| { |
| float3x2 loaded = transpose(v_41.row_major0); |
| v_41.row_major0 = transpose(loaded); |
| } |
| |
| static inline __attribute__((always_inline)) |
| void copy_col_major_to_col_major(device SSBOCol& v_29) |
| { |
| v_29.col_major0 = v_29.col_major1; |
| } |
| |
| static inline __attribute__((always_inline)) |
| void copy_col_major_to_row_major(device SSBOCol& v_29, device SSBORow& v_41) |
| { |
| v_41.row_major0 = transpose(v_29.col_major0); |
| } |
| |
| static inline __attribute__((always_inline)) |
| void copy_row_major_to_col_major(device SSBOCol& v_29, device SSBORow& v_41) |
| { |
| v_29.col_major0 = transpose(v_41.row_major0); |
| } |
| |
| static inline __attribute__((always_inline)) |
| void copy_row_major_to_row_major(device SSBORow& v_41) |
| { |
| v_41.row_major0 = v_41.row_major1; |
| } |
| |
| static inline __attribute__((always_inline)) |
| void copy_columns(device SSBOCol& v_29, device SSBORow& v_41) |
| { |
| v_29.col_major0[1] = float2(v_41.row_major0[0][1], v_41.row_major0[1][1]); |
| v_41.row_major0[0][1] = v_29.col_major0[1].x; |
| v_41.row_major0[1][1] = v_29.col_major0[1].y; |
| } |
| |
| static inline __attribute__((always_inline)) |
| void copy_elements(device SSBOCol& v_29, device SSBORow& v_41) |
| { |
| ((device float*)&v_29.col_major0[0])[1u] = ((device float*)&v_41.row_major0[1u])[0]; |
| ((device float*)&v_41.row_major0[1u])[0] = ((device float*)&v_29.col_major0[0])[1u]; |
| } |
| |
| kernel void main0(device SSBOCol& v_29 [[buffer(0)]], device SSBORow& v_41 [[buffer(1)]]) |
| { |
| load_store_to_variable_col_major(v_29); |
| load_store_to_variable_row_major(v_41); |
| copy_col_major_to_col_major(v_29); |
| copy_col_major_to_row_major(v_29, v_41); |
| copy_row_major_to_col_major(v_29, v_41); |
| copy_row_major_to_row_major(v_41); |
| copy_columns(v_29, v_41); |
| copy_elements(v_29, v_41); |
| } |
| |