SKIP: FAILED

../../src/tint/lang/msl/writer/printer/printer.cc:500 internal compiler error: Inner = struct @align(4) {
  scalar_i32:i32 @offset(0)
  scalar_f32:f32 @offset(4)
}

S = struct @align(16) {
  scalar_f32:f32 @offset(0)
  scalar_i32:i32 @offset(4)
  scalar_u32:u32 @offset(8)
  vec2_f32:vec2<f32> @offset(16)
  vec2_i32:vec2<i32> @offset(24)
  vec2_u32:vec2<u32> @offset(32)
  vec3_f32:vec3<f32> @offset(48)
  vec3_i32:vec3<i32> @offset(64)
  vec3_u32:vec3<u32> @offset(80)
  vec4_f32:vec4<f32> @offset(96)
  vec4_i32:vec4<i32> @offset(112)
  vec4_u32:vec4<u32> @offset(128)
  mat2x2_f32:mat2x2<f32> @offset(144)
  mat2x3_f32:mat2x3<f32> @offset(160)
  mat2x4_f32:mat2x4<f32> @offset(192)
  mat3x2_f32:mat3x2<f32> @offset(224)
  mat3x3_f32:mat3x3<f32> @offset(256)
  mat3x4_f32:mat3x4<f32> @offset(304)
  mat4x2_f32:mat4x2<f32> @offset(352)
  mat4x3_f32:mat4x3<f32> @offset(384)
  mat4x4_f32:mat4x4<f32> @offset(448)
  arr2_vec3_f32:array<vec3<f32>, 2> @offset(512)
  struct_inner:Inner @offset(544)
  array_struct_inner:array<Inner, 4> @offset(552)
}

$B1: {  # root
  %sb:ptr<storage, S, read_write> = var @binding_point(0, 0)
}

%tint_symbol = @compute @workgroup_size(1, 1, 1) func():void {
  $B2: {
    %3:ptr<storage, f32, read_write> = access %sb, 0u
    store %3, 0.0f
    %4:ptr<storage, i32, read_write> = access %sb, 1u
    store %4, 0i
    %5:ptr<storage, u32, read_write> = access %sb, 2u
    store %5, 0u
    %6:ptr<storage, vec2<f32>, read_write> = access %sb, 3u
    store %6, vec2<f32>(0.0f)
    %7:ptr<storage, vec2<i32>, read_write> = access %sb, 4u
    store %7, vec2<i32>(0i)
    %8:ptr<storage, vec2<u32>, read_write> = access %sb, 5u
    store %8, vec2<u32>(0u)
    %9:ptr<storage, vec3<f32>, read_write> = access %sb, 6u
    store %9, vec3<f32>(0.0f)
    %10:ptr<storage, vec3<i32>, read_write> = access %sb, 7u
    store %10, vec3<i32>(0i)
    %11:ptr<storage, vec3<u32>, read_write> = access %sb, 8u
    store %11, vec3<u32>(0u)
    %12:ptr<storage, vec4<f32>, read_write> = access %sb, 9u
    store %12, vec4<f32>(0.0f)
    %13:ptr<storage, vec4<i32>, read_write> = access %sb, 10u
    store %13, vec4<i32>(0i)
    %14:ptr<storage, vec4<u32>, read_write> = access %sb, 11u
    store %14, vec4<u32>(0u)
    %15:ptr<storage, mat2x2<f32>, read_write> = access %sb, 12u
    store %15, mat2x2<f32>(vec2<f32>(0.0f))
    %16:ptr<storage, mat2x3<f32>, read_write> = access %sb, 13u
    %17:void = call %tint_store_and_preserve_padding, %16, mat2x3<f32>(vec3<f32>(0.0f))
    %19:ptr<storage, mat2x4<f32>, read_write> = access %sb, 14u
    store %19, mat2x4<f32>(vec4<f32>(0.0f))
    %20:ptr<storage, mat3x2<f32>, read_write> = access %sb, 15u
    store %20, mat3x2<f32>(vec2<f32>(0.0f))
    %21:ptr<storage, mat3x3<f32>, read_write> = access %sb, 16u
    %22:void = call %tint_store_and_preserve_padding_1, %21, mat3x3<f32>(vec3<f32>(0.0f))
    %24:ptr<storage, mat3x4<f32>, read_write> = access %sb, 17u
    store %24, mat3x4<f32>(vec4<f32>(0.0f))
    %25:ptr<storage, mat4x2<f32>, read_write> = access %sb, 18u
    store %25, mat4x2<f32>(vec2<f32>(0.0f))
    %26:ptr<storage, mat4x3<f32>, read_write> = access %sb, 19u
    %27:void = call %tint_store_and_preserve_padding_2, %26, mat4x3<f32>(vec3<f32>(0.0f))
    %29:ptr<storage, mat4x4<f32>, read_write> = access %sb, 20u
    store %29, mat4x4<f32>(vec4<f32>(0.0f))
    %30:ptr<storage, array<vec3<f32>, 2>, read_write> = access %sb, 21u
    %31:void = call %tint_store_and_preserve_padding_3, %30, array<vec3<f32>, 2>(vec3<f32>(0.0f))
    %33:ptr<storage, Inner, read_write> = access %sb, 22u
    store %33, Inner(0i, 0.0f)
    %34:ptr<storage, array<Inner, 4>, read_write> = access %sb, 23u
    store %34, array<Inner, 4>(Inner(0i, 0.0f))
    ret
  }
}
%tint_store_and_preserve_padding = func(%target:ptr<storage, mat2x3<f32>, read_write>, %value_param:mat2x3<f32>):void {
  $B3: {
    %37:ptr<storage, vec3<f32>, read_write> = access %target, 0u
    %38:vec3<f32> = access %value_param, 0u
    store %37, %38
    %39:ptr<storage, vec3<f32>, read_write> = access %target, 1u
    %40:vec3<f32> = access %value_param, 1u
    store %39, %40
    ret
  }
}
%tint_store_and_preserve_padding_1 = func(%target_1:ptr<storage, mat3x3<f32>, read_write>, %value_param_1:mat3x3<f32>):void {  # %tint_store_and_preserve_padding_1: 'tint_store_and_preserve_padding', %target_1: 'target', %value_param_1: 'value_param'
  $B4: {
    %43:ptr<storage, vec3<f32>, read_write> = access %target_1, 0u
    %44:vec3<f32> = access %value_param_1, 0u
    store %43, %44
    %45:ptr<storage, vec3<f32>, read_write> = access %target_1, 1u
    %46:vec3<f32> = access %value_param_1, 1u
    store %45, %46
    %47:ptr<storage, vec3<f32>, read_write> = access %target_1, 2u
    %48:vec3<f32> = access %value_param_1, 2u
    store %47, %48
    ret
  }
}
%tint_store_and_preserve_padding_2 = func(%target_2:ptr<storage, mat4x3<f32>, read_write>, %value_param_2:mat4x3<f32>):void {  # %tint_store_and_preserve_padding_2: 'tint_store_and_preserve_padding', %target_2: 'target', %value_param_2: 'value_param'
  $B5: {
    %51:ptr<storage, vec3<f32>, read_write> = access %target_2, 0u
    %52:vec3<f32> = access %value_param_2, 0u
    store %51, %52
    %53:ptr<storage, vec3<f32>, read_write> = access %target_2, 1u
    %54:vec3<f32> = access %value_param_2, 1u
    store %53, %54
    %55:ptr<storage, vec3<f32>, read_write> = access %target_2, 2u
    %56:vec3<f32> = access %value_param_2, 2u
    store %55, %56
    %57:ptr<storage, vec3<f32>, read_write> = access %target_2, 3u
    %58:vec3<f32> = access %value_param_2, 3u
    store %57, %58
    ret
  }
}
%tint_store_and_preserve_padding_3 = func(%target_3:ptr<storage, array<vec3<f32>, 2>, read_write>, %value_param_3:array<vec3<f32>, 2>):void {  # %tint_store_and_preserve_padding_3: 'tint_store_and_preserve_padding', %target_3: 'target', %value_param_3: 'value_param'
  $B6: {
    loop [i: $B7, b: $B8, c: $B9] {  # loop_1
      $B7: {  # initializer
        next_iteration 0u  # -> $B8
      }
      $B8 (%idx:u32): {  # body
        %62:bool = gte %idx, 2u
        if %62 [t: $B10] {  # if_1
          $B10: {  # true
            exit_loop  # loop_1
          }
        }
        %63:ptr<storage, vec3<f32>, read_write> = access %target_3, %idx
        %64:vec3<f32> = access %value_param_3, %idx
        store %63, %64
        continue  # -> $B9
      }
      $B9: {  # continuing
        %65:u32 = add %idx, 1u
        next_iteration %65  # -> $B8
      }
    }
    ret
  }
}

unhandled variable address space
********************************************************************
*  The tint shader compiler has encountered an unexpected error.   *
*                                                                  *
*  Please help us fix this issue by submitting a bug report at     *
*  crbug.com/tint with the source program that triggered the bug.  *
********************************************************************