#include <metal_stdlib>
using namespace metal;

struct tint_module_vars_struct {
  thread float3x3* m;
  device float3x3* out;
};

void tint_store_and_preserve_padding(device float3x3* const target, float3x3 value_param) {
  (*target)[0u] = value_param[0u];
  (*target)[1u] = value_param[1u];
  (*target)[2u] = value_param[2u];
}

kernel void f(device float3x3* out [[buffer(0)]]) {
  thread float3x3 m = float3x3(float3(0.0f, 1.0f, 2.0f), float3(3.0f, 4.0f, 5.0f), float3(6.0f, 7.0f, 8.0f));
  tint_module_vars_struct const tint_module_vars = tint_module_vars_struct{.m=(&m), .out=out};
  tint_store_and_preserve_padding(tint_module_vars.out, (*tint_module_vars.m));
}