| 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) |
| scalar_f16:f16 @offset(8) |
| } |
| |
| S = struct @align(16) { |
| scalar_f32:f32 @offset(0) |
| scalar_i32:i32 @offset(4) |
| scalar_u32:u32 @offset(8) |
| scalar_f16:f16 @offset(12) |
| vec2_f32:vec2<f32> @offset(16) |
| vec2_i32:vec2<i32> @offset(24) |
| vec2_u32:vec2<u32> @offset(32) |
| vec2_f16:vec2<f16> @offset(40) |
| vec3_f32:vec3<f32> @offset(48) |
| vec3_i32:vec3<i32> @offset(64) |
| vec3_u32:vec3<u32> @offset(80) |
| vec3_f16:vec3<f16> @offset(96) |
| vec4_f32:vec4<f32> @offset(112) |
| vec4_i32:vec4<i32> @offset(128) |
| vec4_u32:vec4<u32> @offset(144) |
| vec4_f16:vec4<f16> @offset(160) |
| mat2x2_f32:mat2x2<f32> @offset(168) |
| mat2x3_f32:mat2x3<f32> @offset(192) |
| mat2x4_f32:mat2x4<f32> @offset(224) |
| mat3x2_f32:mat3x2<f32> @offset(256) |
| mat3x3_f32:mat3x3<f32> @offset(288) |
| mat3x4_f32:mat3x4<f32> @offset(336) |
| mat4x2_f32:mat4x2<f32> @offset(384) |
| mat4x3_f32:mat4x3<f32> @offset(416) |
| mat4x4_f32:mat4x4<f32> @offset(480) |
| mat2x2_f16:mat2x2<f16> @offset(544) |
| mat2x3_f16:mat2x3<f16> @offset(552) |
| mat2x4_f16:mat2x4<f16> @offset(568) |
| mat3x2_f16:mat3x2<f16> @offset(584) |
| mat3x3_f16:mat3x3<f16> @offset(600) |
| mat3x4_f16:mat3x4<f16> @offset(624) |
| mat4x2_f16:mat4x2<f16> @offset(648) |
| mat4x3_f16:mat4x3<f16> @offset(664) |
| mat4x4_f16:mat4x4<f16> @offset(696) |
| arr2_vec3_f32:array<vec3<f32>, 2> @offset(736) |
| arr2_mat4x2_f16:array<mat4x2<f16>, 2> @offset(768) |
| struct_inner:Inner @offset(800) |
| array_struct_inner:array<Inner, 4> @offset(812) |
| } |
| |
| $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, f16, read_write> = access %sb, 3u |
| store %6, 0.0h |
| %7:ptr<storage, vec2<f32>, read_write> = access %sb, 4u |
| store %7, vec2<f32>(0.0f) |
| %8:ptr<storage, vec2<i32>, read_write> = access %sb, 5u |
| store %8, vec2<i32>(0i) |
| %9:ptr<storage, vec2<u32>, read_write> = access %sb, 6u |
| store %9, vec2<u32>(0u) |
| %10:ptr<storage, vec2<f16>, read_write> = access %sb, 7u |
| store %10, vec2<f16>(0.0h) |
| %11:ptr<storage, vec3<f32>, read_write> = access %sb, 8u |
| store %11, vec3<f32>(0.0f) |
| %12:ptr<storage, vec3<i32>, read_write> = access %sb, 9u |
| store %12, vec3<i32>(0i) |
| %13:ptr<storage, vec3<u32>, read_write> = access %sb, 10u |
| store %13, vec3<u32>(0u) |
| %14:ptr<storage, vec3<f16>, read_write> = access %sb, 11u |
| store %14, vec3<f16>(0.0h) |
| %15:ptr<storage, vec4<f32>, read_write> = access %sb, 12u |
| store %15, vec4<f32>(0.0f) |
| %16:ptr<storage, vec4<i32>, read_write> = access %sb, 13u |
| store %16, vec4<i32>(0i) |
| %17:ptr<storage, vec4<u32>, read_write> = access %sb, 14u |
| store %17, vec4<u32>(0u) |
| %18:ptr<storage, vec4<f16>, read_write> = access %sb, 15u |
| store %18, vec4<f16>(0.0h) |
| %19:ptr<storage, mat2x2<f32>, read_write> = access %sb, 16u |
| store %19, mat2x2<f32>(vec2<f32>(0.0f)) |
| %20:ptr<storage, mat2x3<f32>, read_write> = access %sb, 17u |
| %21:void = call %tint_store_and_preserve_padding, %20, mat2x3<f32>(vec3<f32>(0.0f)) |
| %23:ptr<storage, mat2x4<f32>, read_write> = access %sb, 18u |
| store %23, mat2x4<f32>(vec4<f32>(0.0f)) |
| %24:ptr<storage, mat3x2<f32>, read_write> = access %sb, 19u |
| store %24, mat3x2<f32>(vec2<f32>(0.0f)) |
| %25:ptr<storage, mat3x3<f32>, read_write> = access %sb, 20u |
| %26:void = call %tint_store_and_preserve_padding_1, %25, mat3x3<f32>(vec3<f32>(0.0f)) |
| %28:ptr<storage, mat3x4<f32>, read_write> = access %sb, 21u |
| store %28, mat3x4<f32>(vec4<f32>(0.0f)) |
| %29:ptr<storage, mat4x2<f32>, read_write> = access %sb, 22u |
| store %29, mat4x2<f32>(vec2<f32>(0.0f)) |
| %30:ptr<storage, mat4x3<f32>, read_write> = access %sb, 23u |
| %31:void = call %tint_store_and_preserve_padding_2, %30, mat4x3<f32>(vec3<f32>(0.0f)) |
| %33:ptr<storage, mat4x4<f32>, read_write> = access %sb, 24u |
| store %33, mat4x4<f32>(vec4<f32>(0.0f)) |
| %34:ptr<storage, mat2x2<f16>, read_write> = access %sb, 25u |
| store %34, mat2x2<f16>(vec2<f16>(0.0h)) |
| %35:ptr<storage, mat2x3<f16>, read_write> = access %sb, 26u |
| %36:void = call %tint_store_and_preserve_padding_3, %35, mat2x3<f16>(vec3<f16>(0.0h)) |
| %38:ptr<storage, mat2x4<f16>, read_write> = access %sb, 27u |
| store %38, mat2x4<f16>(vec4<f16>(0.0h)) |
| %39:ptr<storage, mat3x2<f16>, read_write> = access %sb, 28u |
| store %39, mat3x2<f16>(vec2<f16>(0.0h)) |
| %40:ptr<storage, mat3x3<f16>, read_write> = access %sb, 29u |
| %41:void = call %tint_store_and_preserve_padding_4, %40, mat3x3<f16>(vec3<f16>(0.0h)) |
| %43:ptr<storage, mat3x4<f16>, read_write> = access %sb, 30u |
| store %43, mat3x4<f16>(vec4<f16>(0.0h)) |
| %44:ptr<storage, mat4x2<f16>, read_write> = access %sb, 31u |
| store %44, mat4x2<f16>(vec2<f16>(0.0h)) |
| %45:ptr<storage, mat4x3<f16>, read_write> = access %sb, 32u |
| %46:void = call %tint_store_and_preserve_padding_5, %45, mat4x3<f16>(vec3<f16>(0.0h)) |
| %48:ptr<storage, mat4x4<f16>, read_write> = access %sb, 33u |
| store %48, mat4x4<f16>(vec4<f16>(0.0h)) |
| %49:ptr<storage, array<vec3<f32>, 2>, read_write> = access %sb, 34u |
| %50:void = call %tint_store_and_preserve_padding_6, %49, array<vec3<f32>, 2>(vec3<f32>(0.0f)) |
| %52:ptr<storage, array<mat4x2<f16>, 2>, read_write> = access %sb, 35u |
| store %52, array<mat4x2<f16>, 2>(mat4x2<f16>(vec2<f16>(0.0h))) |
| %53:ptr<storage, Inner, read_write> = access %sb, 36u |
| %54:void = call %tint_store_and_preserve_padding_7, %53, Inner(0i, 0.0f, 0.0h) |
| %56:ptr<storage, array<Inner, 4>, read_write> = access %sb, 37u |
| %57:void = call %tint_store_and_preserve_padding_8, %56, array<Inner, 4>(Inner(0i, 0.0f, 0.0h)) |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding = func(%target:ptr<storage, mat2x3<f32>, read_write>, %value_param:mat2x3<f32>):void { |
| $B3: { |
| %61:ptr<storage, vec3<f32>, read_write> = access %target, 0u |
| %62:vec3<f32> = access %value_param, 0u |
| store %61, %62 |
| %63:ptr<storage, vec3<f32>, read_write> = access %target, 1u |
| %64:vec3<f32> = access %value_param, 1u |
| store %63, %64 |
| 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: { |
| %67:ptr<storage, vec3<f32>, read_write> = access %target_1, 0u |
| %68:vec3<f32> = access %value_param_1, 0u |
| store %67, %68 |
| %69:ptr<storage, vec3<f32>, read_write> = access %target_1, 1u |
| %70:vec3<f32> = access %value_param_1, 1u |
| store %69, %70 |
| %71:ptr<storage, vec3<f32>, read_write> = access %target_1, 2u |
| %72:vec3<f32> = access %value_param_1, 2u |
| store %71, %72 |
| 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: { |
| %75:ptr<storage, vec3<f32>, read_write> = access %target_2, 0u |
| %76:vec3<f32> = access %value_param_2, 0u |
| store %75, %76 |
| %77:ptr<storage, vec3<f32>, read_write> = access %target_2, 1u |
| %78:vec3<f32> = access %value_param_2, 1u |
| store %77, %78 |
| %79:ptr<storage, vec3<f32>, read_write> = access %target_2, 2u |
| %80:vec3<f32> = access %value_param_2, 2u |
| store %79, %80 |
| %81:ptr<storage, vec3<f32>, read_write> = access %target_2, 3u |
| %82:vec3<f32> = access %value_param_2, 3u |
| store %81, %82 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_3 = func(%target_3:ptr<storage, mat2x3<f16>, read_write>, %value_param_3:mat2x3<f16>):void { # %tint_store_and_preserve_padding_3: 'tint_store_and_preserve_padding', %target_3: 'target', %value_param_3: 'value_param' |
| $B6: { |
| %85:ptr<storage, vec3<f16>, read_write> = access %target_3, 0u |
| %86:vec3<f16> = access %value_param_3, 0u |
| store %85, %86 |
| %87:ptr<storage, vec3<f16>, read_write> = access %target_3, 1u |
| %88:vec3<f16> = access %value_param_3, 1u |
| store %87, %88 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_4 = func(%target_4:ptr<storage, mat3x3<f16>, read_write>, %value_param_4:mat3x3<f16>):void { # %tint_store_and_preserve_padding_4: 'tint_store_and_preserve_padding', %target_4: 'target', %value_param_4: 'value_param' |
| $B7: { |
| %91:ptr<storage, vec3<f16>, read_write> = access %target_4, 0u |
| %92:vec3<f16> = access %value_param_4, 0u |
| store %91, %92 |
| %93:ptr<storage, vec3<f16>, read_write> = access %target_4, 1u |
| %94:vec3<f16> = access %value_param_4, 1u |
| store %93, %94 |
| %95:ptr<storage, vec3<f16>, read_write> = access %target_4, 2u |
| %96:vec3<f16> = access %value_param_4, 2u |
| store %95, %96 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_5 = func(%target_5:ptr<storage, mat4x3<f16>, read_write>, %value_param_5:mat4x3<f16>):void { # %tint_store_and_preserve_padding_5: 'tint_store_and_preserve_padding', %target_5: 'target', %value_param_5: 'value_param' |
| $B8: { |
| %99:ptr<storage, vec3<f16>, read_write> = access %target_5, 0u |
| %100:vec3<f16> = access %value_param_5, 0u |
| store %99, %100 |
| %101:ptr<storage, vec3<f16>, read_write> = access %target_5, 1u |
| %102:vec3<f16> = access %value_param_5, 1u |
| store %101, %102 |
| %103:ptr<storage, vec3<f16>, read_write> = access %target_5, 2u |
| %104:vec3<f16> = access %value_param_5, 2u |
| store %103, %104 |
| %105:ptr<storage, vec3<f16>, read_write> = access %target_5, 3u |
| %106:vec3<f16> = access %value_param_5, 3u |
| store %105, %106 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_6 = func(%target_6:ptr<storage, array<vec3<f32>, 2>, read_write>, %value_param_6:array<vec3<f32>, 2>):void { # %tint_store_and_preserve_padding_6: 'tint_store_and_preserve_padding', %target_6: 'target', %value_param_6: 'value_param' |
| $B9: { |
| loop [i: $B10, b: $B11, c: $B12] { # loop_1 |
| $B10: { # initializer |
| next_iteration 0u # -> $B11 |
| } |
| $B11 (%idx:u32): { # body |
| %110:bool = gte %idx, 2u |
| if %110 [t: $B13] { # if_1 |
| $B13: { # true |
| exit_loop # loop_1 |
| } |
| } |
| %111:ptr<storage, vec3<f32>, read_write> = access %target_6, %idx |
| %112:vec3<f32> = access %value_param_6, %idx |
| store %111, %112 |
| continue # -> $B12 |
| } |
| $B12: { # continuing |
| %113:u32 = add %idx, 1u |
| next_iteration %113 # -> $B11 |
| } |
| } |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_7 = func(%target_7:ptr<storage, Inner, read_write>, %value_param_7:Inner):void { # %tint_store_and_preserve_padding_7: 'tint_store_and_preserve_padding', %target_7: 'target', %value_param_7: 'value_param' |
| $B14: { |
| %116:ptr<storage, i32, read_write> = access %target_7, 0u |
| %117:i32 = access %value_param_7, 0u |
| store %116, %117 |
| %118:ptr<storage, f32, read_write> = access %target_7, 1u |
| %119:f32 = access %value_param_7, 1u |
| store %118, %119 |
| %120:ptr<storage, f16, read_write> = access %target_7, 2u |
| %121:f16 = access %value_param_7, 2u |
| store %120, %121 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_8 = func(%target_8:ptr<storage, array<Inner, 4>, read_write>, %value_param_8:array<Inner, 4>):void { # %tint_store_and_preserve_padding_8: 'tint_store_and_preserve_padding', %target_8: 'target', %value_param_8: 'value_param' |
| $B15: { |
| loop [i: $B16, b: $B17, c: $B18] { # loop_2 |
| $B16: { # initializer |
| next_iteration 0u # -> $B17 |
| } |
| $B17 (%idx_1:u32): { # body |
| %125:bool = gte %idx_1, 4u |
| if %125 [t: $B19] { # if_2 |
| $B19: { # true |
| exit_loop # loop_2 |
| } |
| } |
| %126:ptr<storage, Inner, read_write> = access %target_8, %idx_1 |
| %127:Inner = access %value_param_8, %idx_1 |
| %128:void = call %tint_store_and_preserve_padding_7, %126, %127 |
| continue # -> $B18 |
| } |
| $B18: { # continuing |
| %129:u32 = add %idx_1, 1u |
| next_iteration %129 # -> $B17 |
| } |
| } |
| 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. * |
| ******************************************************************** |
