| SKIP: FAILED |
| |
| ../../src/tint/lang/msl/writer/printer/printer.cc:500 internal compiler error: S = struct @align(16) { |
| m:mat3x3<f32> @offset(0) |
| } |
| |
| S2 = struct @align(16) { |
| m:array<mat3x3<f32>, 1> @offset(0) |
| } |
| |
| S3 = struct @align(16) { |
| s:S @offset(0) |
| } |
| |
| S4 = struct @align(16) { |
| s:array<S, 1> @offset(0) |
| } |
| |
| $B1: { # root |
| %buffer0:ptr<storage, mat3x3<f32>, read_write> = var @binding_point(0, 0) |
| %buffer1:ptr<storage, S, read_write> = var @binding_point(0, 1) |
| %buffer2:ptr<storage, S2, read_write> = var @binding_point(0, 2) |
| %buffer3:ptr<storage, S3, read_write> = var @binding_point(0, 3) |
| %buffer4:ptr<storage, S4, read_write> = var @binding_point(0, 4) |
| %buffer5:ptr<storage, array<mat3x3<f32>, 1>, read_write> = var @binding_point(0, 5) |
| %buffer6:ptr<storage, array<S, 1>, read_write> = var @binding_point(0, 6) |
| %buffer7:ptr<storage, array<S2, 1>, read_write> = var @binding_point(0, 7) |
| } |
| |
| %tint_symbol = @compute @workgroup_size(1, 1, 1) func():void { |
| $B2: { |
| %m:ptr<function, mat3x3<f32>, read_write> = var |
| loop [i: $B3, b: $B4, c: $B5] { # loop_1 |
| $B3: { # initializer |
| %c:ptr<function, u32, read_write> = var, 0u |
| next_iteration # -> $B4 |
| } |
| $B4: { # body |
| %12:u32 = load %c |
| %13:bool = lt %12, 3u |
| if %13 [t: $B6, f: $B7] { # if_1 |
| $B6: { # true |
| exit_if # if_1 |
| } |
| $B7: { # false |
| exit_loop # loop_1 |
| } |
| } |
| %14:u32 = load %c |
| %15:ptr<function, vec3<f32>, read_write> = access %m, %14 |
| %16:ptr<function, vec3<f32>, read_write> = let %15 |
| %17:u32 = load %c |
| %18:u32 = mul %17, 3u |
| %19:u32 = add %18, 1u |
| %20:f32 = convert %19 |
| %21:f32 = let %20 |
| %22:u32 = load %c |
| %23:u32 = mul %22, 3u |
| %24:u32 = add %23, 2u |
| %25:f32 = convert %24 |
| %26:f32 = let %25 |
| %27:u32 = load %c |
| %28:u32 = mul %27, 3u |
| %29:u32 = add %28, 3u |
| %30:f32 = convert %29 |
| %31:vec3<f32> = construct %21, %26, %30 |
| store %16, %31 |
| continue # -> $B5 |
| } |
| $B5: { # continuing |
| %32:u32 = load %c |
| %33:u32 = add %32, 1u |
| store %c, %33 |
| next_iteration # -> $B4 |
| } |
| } |
| %34:mat3x3<f32> = load %m |
| %a:mat3x3<f32> = let %34 |
| %36:void = call %tint_store_and_preserve_padding, %buffer0, %a |
| %38:mat3x3<f32> = load %m |
| %39:S = construct %38 |
| %a_1:S = let %39 # %a_1: 'a' |
| %41:void = call %tint_store_and_preserve_padding_1, %buffer1, %a_1 |
| %43:mat3x3<f32> = load %m |
| %44:array<mat3x3<f32>, 1> = construct %43 |
| %45:S2 = construct %44 |
| %a_2:S2 = let %45 # %a_2: 'a' |
| %47:void = call %tint_store_and_preserve_padding_2, %buffer2, %a_2 |
| %49:mat3x3<f32> = load %m |
| %50:S = construct %49 |
| %51:S3 = construct %50 |
| %a_3:S3 = let %51 # %a_3: 'a' |
| %53:void = call %tint_store_and_preserve_padding_3, %buffer3, %a_3 |
| %55:mat3x3<f32> = load %m |
| %56:S = construct %55 |
| %57:array<S, 1> = construct %56 |
| %58:S4 = construct %57 |
| %a_4:S4 = let %58 # %a_4: 'a' |
| %60:void = call %tint_store_and_preserve_padding_4, %buffer4, %a_4 |
| %62:mat3x3<f32> = load %m |
| %63:array<mat3x3<f32>, 1> = construct %62 |
| %a_5:array<mat3x3<f32>, 1> = let %63 # %a_5: 'a' |
| %65:void = call %tint_store_and_preserve_padding_5, %buffer5, %a_5 |
| %67:mat3x3<f32> = load %m |
| %68:S = construct %67 |
| %69:array<S, 1> = construct %68 |
| %a_6:array<S, 1> = let %69 # %a_6: 'a' |
| %71:void = call %tint_store_and_preserve_padding_6, %buffer6, %a_6 |
| %73:mat3x3<f32> = load %m |
| %74:array<mat3x3<f32>, 1> = construct %73 |
| %75:S2 = construct %74 |
| %76:array<S2, 1> = construct %75 |
| %a_7:array<S2, 1> = let %76 # %a_7: 'a' |
| %78:void = call %tint_store_and_preserve_padding_7, %buffer7, %a_7 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding = func(%target:ptr<storage, mat3x3<f32>, read_write>, %value_param:mat3x3<f32>):void { |
| $B8: { |
| %82:ptr<storage, vec3<f32>, read_write> = access %target, 0u |
| %83:vec3<f32> = access %value_param, 0u |
| store %82, %83 |
| %84:ptr<storage, vec3<f32>, read_write> = access %target, 1u |
| %85:vec3<f32> = access %value_param, 1u |
| store %84, %85 |
| %86:ptr<storage, vec3<f32>, read_write> = access %target, 2u |
| %87:vec3<f32> = access %value_param, 2u |
| store %86, %87 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_1 = func(%target_1:ptr<storage, S, read_write>, %value_param_1:S):void { # %tint_store_and_preserve_padding_1: 'tint_store_and_preserve_padding', %target_1: 'target', %value_param_1: 'value_param' |
| $B9: { |
| %90:ptr<storage, mat3x3<f32>, read_write> = access %target_1, 0u |
| %91:mat3x3<f32> = access %value_param_1, 0u |
| %92:void = call %tint_store_and_preserve_padding, %90, %91 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_2 = func(%target_2:ptr<storage, S2, read_write>, %value_param_2:S2):void { # %tint_store_and_preserve_padding_2: 'tint_store_and_preserve_padding', %target_2: 'target', %value_param_2: 'value_param' |
| $B10: { |
| %95:ptr<storage, array<mat3x3<f32>, 1>, read_write> = access %target_2, 0u |
| %96:array<mat3x3<f32>, 1> = access %value_param_2, 0u |
| %97:void = call %tint_store_and_preserve_padding_5, %95, %96 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_5 = func(%target_3:ptr<storage, array<mat3x3<f32>, 1>, read_write>, %value_param_3:array<mat3x3<f32>, 1>):void { # %tint_store_and_preserve_padding_5: 'tint_store_and_preserve_padding', %target_3: 'target', %value_param_3: 'value_param' |
| $B11: { |
| loop [i: $B12, b: $B13, c: $B14] { # loop_2 |
| $B12: { # initializer |
| next_iteration 0u # -> $B13 |
| } |
| $B13 (%idx:u32): { # body |
| %101:bool = gte %idx, 1u |
| if %101 [t: $B15] { # if_2 |
| $B15: { # true |
| exit_loop # loop_2 |
| } |
| } |
| %102:ptr<storage, mat3x3<f32>, read_write> = access %target_3, %idx |
| %103:mat3x3<f32> = access %value_param_3, %idx |
| %104:void = call %tint_store_and_preserve_padding, %102, %103 |
| continue # -> $B14 |
| } |
| $B14: { # continuing |
| %105:u32 = add %idx, 1u |
| next_iteration %105 # -> $B13 |
| } |
| } |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_3 = func(%target_4:ptr<storage, S3, read_write>, %value_param_4:S3):void { # %tint_store_and_preserve_padding_3: 'tint_store_and_preserve_padding', %target_4: 'target', %value_param_4: 'value_param' |
| $B16: { |
| %108:ptr<storage, S, read_write> = access %target_4, 0u |
| %109:S = access %value_param_4, 0u |
| %110:void = call %tint_store_and_preserve_padding_1, %108, %109 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_4 = func(%target_5:ptr<storage, S4, read_write>, %value_param_5:S4):void { # %tint_store_and_preserve_padding_4: 'tint_store_and_preserve_padding', %target_5: 'target', %value_param_5: 'value_param' |
| $B17: { |
| %113:ptr<storage, array<S, 1>, read_write> = access %target_5, 0u |
| %114:array<S, 1> = access %value_param_5, 0u |
| %115:void = call %tint_store_and_preserve_padding_6, %113, %114 |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_6 = func(%target_6:ptr<storage, array<S, 1>, read_write>, %value_param_6:array<S, 1>):void { # %tint_store_and_preserve_padding_6: 'tint_store_and_preserve_padding', %target_6: 'target', %value_param_6: 'value_param' |
| $B18: { |
| loop [i: $B19, b: $B20, c: $B21] { # loop_3 |
| $B19: { # initializer |
| next_iteration 0u # -> $B20 |
| } |
| $B20 (%idx_1:u32): { # body |
| %119:bool = gte %idx_1, 1u |
| if %119 [t: $B22] { # if_3 |
| $B22: { # true |
| exit_loop # loop_3 |
| } |
| } |
| %120:ptr<storage, S, read_write> = access %target_6, %idx_1 |
| %121:S = access %value_param_6, %idx_1 |
| %122:void = call %tint_store_and_preserve_padding_1, %120, %121 |
| continue # -> $B21 |
| } |
| $B21: { # continuing |
| %123:u32 = add %idx_1, 1u |
| next_iteration %123 # -> $B20 |
| } |
| } |
| ret |
| } |
| } |
| %tint_store_and_preserve_padding_7 = func(%target_7:ptr<storage, array<S2, 1>, read_write>, %value_param_7:array<S2, 1>):void { # %tint_store_and_preserve_padding_7: 'tint_store_and_preserve_padding', %target_7: 'target', %value_param_7: 'value_param' |
| $B23: { |
| loop [i: $B24, b: $B25, c: $B26] { # loop_4 |
| $B24: { # initializer |
| next_iteration 0u # -> $B25 |
| } |
| $B25 (%idx_2:u32): { # body |
| %127:bool = gte %idx_2, 1u |
| if %127 [t: $B27] { # if_4 |
| $B27: { # true |
| exit_loop # loop_4 |
| } |
| } |
| %128:ptr<storage, S2, read_write> = access %target_7, %idx_2 |
| %129:S2 = access %value_param_7, %idx_2 |
| %130:void = call %tint_store_and_preserve_padding_2, %128, %129 |
| continue # -> $B26 |
| } |
| $B26: { # continuing |
| %131:u32 = add %idx_2, 1u |
| next_iteration %131 # -> $B25 |
| } |
| } |
| 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. * |
| ******************************************************************** |
