test/tint/ptr_ref/store/param/workgroup/vec2_f32_in_mat2x2.wgsl.expected.msl

#include <metal_stdlib>

using namespace metal;
struct tint_symbol_5 {
  float2x2 S;
};

void tint_zero_workgroup_memory(uint local_idx, threadgroup float2x2* const tint_symbol_1) {
  if ((local_idx < 1u)) {
    *(tint_symbol_1) = float2x2(float2(0.0f), float2(0.0f));
  }
  threadgroup_barrier(mem_flags::mem_threadgroup);
}

void func(threadgroup float2* const pointer) {
  *(pointer) = float2(0.0f);
}

void tint_symbol_inner(uint local_invocation_index, threadgroup float2x2* const tint_symbol_2) {
  tint_zero_workgroup_memory(local_invocation_index, tint_symbol_2);
  func(&((*(tint_symbol_2))[1]));
}

kernel void tint_symbol(threadgroup tint_symbol_5* tint_symbol_4 [[threadgroup(0)]], uint local_invocation_index [[thread_index_in_threadgroup]]) {
  threadgroup float2x2* const tint_symbol_3 = &((*(tint_symbol_4)).S);
  tint_symbol_inner(local_invocation_index, tint_symbol_3);
  return;
}