[tint][ir] Print all operands to various instructions - Improve a variety of instructions to print all of the operands, instead of just the expected one, when disassembling. - Also removes spurious ','s - Updating affected tests Issue: 399124608 Change-Id: I56265732224c9781100eb394e4537e4f8c25d9aa Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/228075 Auto-Submit: Ryan Harrison <rharrison@chromium.org> Reviewed-by: dan sinclair <dsinclair@chromium.org> Commit-Queue: Ryan Harrison <rharrison@chromium.org> Reviewed-by: James Price <jrprice@google.com>
diff --git a/src/tint/lang/core/ir/disassembler.cc b/src/tint/lang/core/ir/disassembler.cc index c092f88..3d915bf 100644 --- a/src/tint/lang/core/ir/disassembler.cc +++ b/src/tint/lang/core/ir/disassembler.cc
@@ -497,28 +497,17 @@ [&](const Discard* d) { EmitInstructionName(d); }, [&](const Store* s) { EmitInstructionName(s); - out_ << " "; - EmitOperand(s, Store::kToOperandOffset); - out_ << ", "; - EmitOperand(s, Store::kFromOperandOffset); + EmitOperandList(s); }, [&](const StoreVectorElement* s) { EmitInstructionName(s); - if (s->Operands().Length() > 0) { - out_ << " "; - EmitOperandList(s); - } + EmitOperandList(s); }, [&](const UserCall* uc) { EmitValueWithType(uc); out_ << " = "; EmitInstructionName(uc); - out_ << " "; - EmitOperand(uc, UserCall::kFunctionOperandOffset); - if (!uc->Args().IsEmpty()) { - out_ << ", "; - } - EmitOperandList(uc, UserCall::kArgsOperandOffset); + EmitOperandList(uc); }, [&](const BuiltinCall* c) { EmitValueWithType(c); @@ -537,10 +526,7 @@ out_ << ">"; } - if (!c->Args().IsEmpty()) { - out_ << " "; - EmitOperandList(c, BuiltinCall::kArgsOperandOffset); - } + EmitOperandList(c, BuiltinCall::kArgsOperandOffset); }, [&](const MemberBuiltinCall* c) { EmitValueWithType(c); @@ -548,30 +534,20 @@ EmitOperand(c, MemberBuiltinCall::kObjectOperandOffset); out_ << "."; EmitInstructionName(c); - if (!c->Args().IsEmpty()) { - out_ << " "; - EmitOperandList(c, UserCall::kArgsOperandOffset); - } + EmitOperandList(c, UserCall::kArgsOperandOffset); }, [&](const Override* o) { EmitValueWithType(o); out_ << " = "; EmitInstructionName(o); - if (o->Initializer()) { - out_ << ", "; - EmitOperand(o, Var::kInitializerOperandOffset); - } + EmitOperandList(o); out_ << " @id(" << o->OverrideId().value << ")"; }, [&](const Var* v) { EmitValueWithType(v); out_ << " = "; EmitInstructionName(v); - - if (v->Operands().Length() > 0) { - out_ << " "; - EmitOperandList(v); - } + EmitOperandList(v); if (v->BindingPoint().has_value()) { out_ << " "; @@ -640,10 +616,7 @@ EmitValueWithType(inst); out_ << " = "; EmitInstructionName(inst); - if (!inst->Operands().IsEmpty()) { - out_ << " "; - EmitOperandList(inst); - } + EmitOperandList(inst); }); { // Add a comment if the result IDs don't match their names @@ -674,6 +647,8 @@ for (size_t i = start_index, n = inst->Operands().Length(); i < n; i++) { if (i != start_index) { out_ << ", "; + } else { + out_ << " "; } EmitOperand(inst, i); } @@ -684,6 +659,8 @@ for (size_t i = start_index; i < n; i++) { if (i != start_index) { out_ << ", "; + } else { + out_ << " "; } EmitOperand(inst, i); } @@ -894,12 +871,12 @@ auto next_iter_values = bi->NextIterValues(); auto exit_values = bi->ExitValues(); if (!next_iter_values.IsEmpty()) { - out_ << " " << StyleLabel("next_iteration") << ": [ "; + out_ << " " << StyleLabel("next_iteration") << ": ["; EmitOperandList(bi, ir::BreakIf::kArgsOperandOffset, next_iter_values.Length()); out_ << " ]"; } if (!exit_values.IsEmpty()) { - out_ << " " << StyleLabel("exit_loop") << ": [ "; + out_ << " " << StyleLabel("exit_loop") << ": ["; EmitOperandList(bi, ir::BreakIf::kArgsOperandOffset + next_iter_values.Length()); out_ << " ]"; } @@ -914,8 +891,7 @@ return std::nullopt; }); - if (args_offset && !term->Args().IsEmpty()) { - out_ << " "; + if (args_offset) { EmitOperandList(term, *args_offset); } @@ -947,11 +923,7 @@ SourceMarker sm(this); EmitValueWithType(b); out_ << " = " << NameOf(b->Op()); - - if (b->Operands().Length() > 0) { - out_ << " "; - EmitOperandList(b); - } + EmitOperandList(b); sm.Store(b); } @@ -960,11 +932,7 @@ SourceMarker sm(this); EmitValueWithType(u); out_ << " = " << NameOf(u->Op()); - - if (u->Operands().Length() > 0) { - out_ << " "; - EmitOperandList(u); - } + EmitOperandList(u); sm.Store(u); }
diff --git a/src/tint/lang/core/ir/referenced_module_decls_test.cc b/src/tint/lang/core/ir/referenced_module_decls_test.cc index 6b5dbb4..de9dadf 100644 --- a/src/tint/lang/core/ir/referenced_module_decls_test.cc +++ b/src/tint/lang/core/ir/referenced_module_decls_test.cc
@@ -93,9 +93,9 @@ %a:ptr<workgroup, u32, read_write> = var undef %b:ptr<workgroup, u32, read_write> = var undef %3:i32 = add 1i, 2i - %o:i32 = override, %3 @id(1) + %o:i32 = override %3 @id(1) %c:ptr<workgroup, u32, read_write> = var undef - %p:i32 = override @id(0) + %p:i32 = override undef @id(0) %7:i32 = mul 2i, 4i } @@ -137,8 +137,8 @@ %a:ptr<workgroup, u32, read_write> = var undef %b:ptr<workgroup, u32, read_write> = var undef %c:ptr<workgroup, u32, read_write> = var undef - %d:i32 = override @id(0) - %e:i32 = override @id(1) + %d:i32 = override undef @id(0) + %e:i32 = override undef @id(1) } %foo = func():void { @@ -190,7 +190,7 @@ %a:ptr<workgroup, u32, read_write> = var undef %b:ptr<workgroup, u32, read_write> = var undef %c:ptr<workgroup, u32, read_write> = var undef - %d:i32 = override @id(0) + %d:i32 = override undef @id(0) } %foo = func():void { @@ -260,7 +260,7 @@ %a:ptr<workgroup, u32, read_write> = var undef %b:ptr<workgroup, u32, read_write> = var undef %c:ptr<workgroup, u32, read_write> = var undef - %c_1:i32 = override @id(0) # %c_1: 'c' + %c_1:i32 = override undef @id(0) # %c_1: 'c' } %foo = func():void { @@ -328,7 +328,7 @@ auto* src = R"( $B1: { # root %a:ptr<workgroup, u32, read_write> = var undef - %b:i32 = override @id(0) + %b:i32 = override undef @id(0) %c:ptr<workgroup, u32, read_write> = var undef } @@ -423,7 +423,7 @@ auto* src = R"( $B1: { # root - %o:u32 = override @id(1) + %o:u32 = override undef @id(1) } %foo = @compute @workgroup_size(%o, 1u, 1u) func():void { @@ -455,7 +455,7 @@ auto* src = R"( $B1: { # root - %o:u32 = override @id(1) + %o:u32 = override undef @id(1) %a:ptr<workgroup, array<i32, %o>, read_write> = var undef }
diff --git a/src/tint/lang/core/ir/transform/single_entry_point_test.cc b/src/tint/lang/core/ir/transform/single_entry_point_test.cc index 416570f..44bb8c8 100644 --- a/src/tint/lang/core/ir/transform/single_entry_point_test.cc +++ b/src/tint/lang/core/ir/transform/single_entry_point_test.cc
@@ -280,9 +280,9 @@ auto* src = R"( $B1: { # root - %o1:i32 = override @id(1) - %o2:i32 = override @id(2) - %o3:i32 = override @id(3) + %o1:i32 = override undef @id(1) + %o2:i32 = override undef @id(2) + %o3:i32 = override undef @id(3) } %foo = @fragment func():void { @@ -302,8 +302,8 @@ auto* expect = R"( $B1: { # root - %o1:i32 = override @id(1) - %o2:i32 = override @id(2) + %o1:i32 = override undef @id(1) + %o2:i32 = override undef @id(2) } %foo = @fragment func():void { @@ -350,9 +350,9 @@ %3:i32 = add %2, 4i %4:i32 = mul 3i, 5i %5:i32 = add %4, 5i - %o1:i32 = override, %1 @id(1) - %o2:i32 = override, %3 @id(2) - %o3:i32 = override, %5 @id(3) + %o1:i32 = override %1 @id(1) + %o2:i32 = override %3 @id(2) + %o3:i32 = override %5 @id(3) } %foo = @fragment func():void { @@ -375,8 +375,8 @@ %1:i32 = mul 2i, 4i %2:i32 = mul 2i, 4i %3:i32 = add %2, 4i - %o1:i32 = override, %1 @id(1) - %o2:i32 = override, %3 @id(2) + %o1:i32 = override %1 @id(1) + %o2:i32 = override %3 @id(2) } %foo = @fragment func():void { @@ -509,10 +509,10 @@ auto* src = R"( $B1: { # root - %unused_override:i32 = override @id(5) - %o1:i32 = override @id(1) - %o2:i32 = override @id(2) - %o3:i32 = override @id(3) + %unused_override:i32 = override undef @id(5) + %o1:i32 = override undef @id(1) + %o2:i32 = override undef @id(2) + %o3:i32 = override undef @id(3) } %unused_func = func():void { @@ -550,9 +550,9 @@ auto* expect = R"( $B1: { # root - %o1:i32 = override @id(1) - %o2:i32 = override @id(2) - %o3:i32 = override @id(3) + %o1:i32 = override undef @id(1) + %o2:i32 = override undef @id(2) + %o3:i32 = override undef @id(3) } %f1 = func():void { @@ -765,13 +765,13 @@ auto* src = R"( $B1: { # root - %o1:i32 = override @id(1) - %o2:i32 = override @id(2) - %o3:i32 = override @id(3) - %o4:i32 = override @id(4) - %o5:i32 = override @id(5) - %o6:i32 = override @id(6) - %o7:i32 = override @id(7) + %o1:i32 = override undef @id(1) + %o2:i32 = override undef @id(2) + %o3:i32 = override undef @id(3) + %o4:i32 = override undef @id(4) + %o5:i32 = override undef @id(5) + %o6:i32 = override undef @id(6) + %o7:i32 = override undef @id(7) } %foo = @fragment func():void { @@ -797,8 +797,8 @@ auto* expect = R"( $B1: { # root - %o1:i32 = override @id(1) - %o5:i32 = override @id(5) + %o1:i32 = override undef @id(1) + %o5:i32 = override undef @id(5) } %foo = @fragment func():void { @@ -831,9 +831,9 @@ auto* src = R"( $B1: { # root - %o1:i32 = override @id(1) - %o2:i32 = override @id(2) - %o3:i32 = override @id(3) + %o1:i32 = override undef @id(1) + %o2:i32 = override undef @id(2) + %o3:i32 = override undef @id(3) } %foo = @compute @workgroup_size(%o1, %o2, %o1) func():void { @@ -850,8 +850,8 @@ auto* expect = R"( $B1: { # root - %o1:i32 = override @id(1) - %o2:i32 = override @id(2) + %o1:i32 = override undef @id(1) + %o2:i32 = override undef @id(2) } %foo = @compute @workgroup_size(%o1, %o2, %o1) func():void { @@ -891,8 +891,8 @@ auto* src = R"( $B1: { # root - %o1:i32 = override @id(1) - %o2:i32 = override @id(2) + %o1:i32 = override undef @id(1) + %o2:i32 = override undef @id(2) %a:ptr<workgroup, array<i32, %o1>, read_write> = var undef %b:ptr<workgroup, array<i32, %o2>, read_write> = var undef } @@ -913,7 +913,7 @@ auto* expect = R"( $B1: { # root - %o1:i32 = override @id(1) + %o1:i32 = override undef @id(1) %a:ptr<workgroup, array<i32, %o1>, read_write> = var undef }
diff --git a/src/tint/lang/core/ir/transform/substitute_overrides_test.cc b/src/tint/lang/core/ir/transform/substitute_overrides_test.cc index 359b4df..74d3c8b 100644 --- a/src/tint/lang/core/ir/transform/substitute_overrides_test.cc +++ b/src/tint/lang/core/ir/transform/substitute_overrides_test.cc
@@ -69,7 +69,7 @@ auto* src = R"( $B1: { # root - %a:i32 = override @id(1) + %a:i32 = override undef @id(1) } )"; @@ -94,7 +94,7 @@ auto* src = R"( $B1: { # root - %a:u32 = override, 2u @id(1) + %a:u32 = override 2u @id(1) } %foo = func():u32 { @@ -132,7 +132,7 @@ auto* src = R"( $B1: { # root - %a:u32 = override, 2u @id(1) + %a:u32 = override 2u @id(1) } %foo = func():u32 { @@ -171,7 +171,7 @@ auto* src = R"( $B1: { # root - %a:u32 = override @id(1) + %a:u32 = override undef @id(1) } %foo = func():u32 { @@ -214,7 +214,7 @@ auto* src = R"( $B1: { # root %1:u32 = add 2u, 4u - %a:u32 = override, %1 @id(1) + %a:u32 = override %1 @id(1) } %foo = func():u32 { @@ -256,7 +256,7 @@ auto* src = R"( $B1: { # root %1:u32 = add 2u, 4u - %a:u32 = override, %1 @id(1) + %a:u32 = override %1 @id(1) } %foo = func():u32 { @@ -301,9 +301,9 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %2:u32 = add %x, 4u - %a:u32 = override, %2 @id(1) + %a:u32 = override %2 @id(1) } %foo = func():u32 { @@ -349,9 +349,9 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %2:u32 = add %x, 4u - %a:u32 = override, %2 @id(1) + %a:u32 = override %2 @id(1) } %foo = func():u32 { @@ -392,9 +392,9 @@ auto* src = R"( $B1: { # root - %x:f32 = override @id(2) + %x:f32 = override undef @id(2) %2:f32 = add %x, 4.0f - %a:f32 = override, %2 @id(1) + %a:f32 = override %2 @id(1) } %foo = func():f32 { @@ -434,9 +434,9 @@ auto* src = R"( $B1: { # root - %x:f32 = override @id(2) + %x:f32 = override undef @id(2) %2:f32 = add %x, 340282346638528859811704183484516925440.0f - %a:f32 = override, %2 @id(1) + %a:f32 = override %2 @id(1) } %foo = func():f32 { @@ -475,9 +475,9 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %2:u32 = add %x, 4u - %a:u32 = override, %2 @id(1) + %a:u32 = override %2 @id(1) } %foo = @compute @workgroup_size(%a, %x, %a) func():void { @@ -527,9 +527,9 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %2:u32 = add %x, 4u - %a:u32 = override, %2 @id(1) + %a:u32 = override %2 @id(1) } %foo = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -577,7 +577,7 @@ auto* src = R"( $B1: { # root - %x:f32 = override @id(2) + %x:f32 = override undef @id(2) } %foo = @fragment func():void { @@ -633,9 +633,9 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %2:u32 = add %x, 4u - %a:u32 = override, %2 @id(1) + %a:u32 = override %2 @id(1) } %foo = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -694,9 +694,9 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %2:u32 = add %x, 4u - %a:u32 = override, %2 @id(1) + %a:u32 = override %2 @id(1) } %foo = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -755,9 +755,9 @@ auto* src = R"( $B1: { # root - %x:f32 = override @id(2) + %x:f32 = override undef @id(2) %2:f32 = add %x, 4.0f - %a:f32 = override, %2 @id(1) + %a:f32 = override %2 @id(1) } %foo = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -817,9 +817,9 @@ auto* src = R"( $B1: { # root - %x:f32 = override @id(2) + %x:f32 = override undef @id(2) %2:f32 = add %x, 4.0f - %a:f32 = override, %2 @id(1) + %a:f32 = override %2 @id(1) } %foo = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -872,7 +872,7 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %v:ptr<workgroup, array<i32, %x>, read_write> = var undef } @@ -920,7 +920,7 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %2:u32 = mul %x, 2u %v:ptr<workgroup, array<i32, %2>, read_write> = var undef } @@ -973,7 +973,7 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) %v:ptr<workgroup, array<i32, %x>, read_write> = var undef } @@ -1034,8 +1034,8 @@ auto* src = R"( $B1: { # root - %cond:bool = override @id(0) - %one_f32:f32 = override, 1.0f @id(2) + %cond:bool = override undef @id(0) + %one_f32:f32 = override 1.0f @id(2) %3:bool = constexpr_if %cond [t: $B2, f: $B3] { # constexpr_if_1 $B2: { # true %4:f32 = div %one_f32, 0.0f @@ -1046,7 +1046,7 @@ exit_if false # constexpr_if_1 } } - %foo:bool = override, %3 @id(1) + %foo:bool = override %3 @id(1) } %foo2 = func():bool { @@ -1098,8 +1098,8 @@ auto* src = R"( $B1: { # root - %cond:bool = override @id(0) - %one_f32:f32 = override, 1.0f @id(2) + %cond:bool = override undef @id(0) + %one_f32:f32 = override 1.0f @id(2) %3:bool = constexpr_if %cond [t: $B2, f: $B3] { # constexpr_if_1 $B2: { # true %4:f32 = div %one_f32, 0.0f @@ -1110,7 +1110,7 @@ exit_if false # constexpr_if_1 } } - %foo:bool = override, %3 @id(1) + %foo:bool = override %3 @id(1) } %foo2 = func():bool { @@ -1157,8 +1157,8 @@ auto* src = R"( $B1: { # root - %cond:bool = override @id(0) - %one_f32:f32 = override, 1.0f @id(2) + %cond:bool = override undef @id(0) + %one_f32:f32 = override 1.0f @id(2) %3:bool = constexpr_if %cond [t: $B2, f: $B3] { # constexpr_if_1 $B2: { # true %4:f32 = div %one_f32, 1.0f @@ -1169,7 +1169,7 @@ exit_if true # constexpr_if_1 } } - %foo:bool = override, %3 @id(1) + %foo:bool = override %3 @id(1) } %foo2 = func():bool { @@ -1233,8 +1233,8 @@ auto* src = R"( $B1: { # root - %cond:bool = override @id(0) - %zero_f32:f32 = override, 0.0f @id(2) + %cond:bool = override undef @id(0) + %zero_f32:f32 = override 0.0f @id(2) %3:bool = constexpr_if %cond [t: $B2, f: $B3] { # constexpr_if_1 $B2: { # true %4:bool = constexpr_if %cond [t: $B4, f: $B5] { # constexpr_if_2 @@ -1255,7 +1255,7 @@ exit_if false # constexpr_if_1 } } - %foo:bool = override, %3 @id(1) + %foo:bool = override %3 @id(1) } %foo2 = func():bool { @@ -1310,8 +1310,8 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(1) - %y:bool = override @id(2) + %x:u32 = override undef @id(1) + %y:bool = override undef @id(2) } %foo = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -1376,7 +1376,7 @@ auto* src = R"( $B1: { # root - %y:bool = override @id(1) + %y:bool = override undef @id(1) } %foo = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -1439,7 +1439,7 @@ auto* src = R"( $B1: { # root - %x:u32 = override @id(0) + %x:u32 = override undef @id(0) } %foo2 = func():u32 { @@ -1480,7 +1480,7 @@ auto* src = R"( $B1: { # root - %x:i32 = override @id(0) + %x:i32 = override undef @id(0) %arr:ptr<storage, array<u32>, read_write> = var undef @binding_point(0, 0) } @@ -1530,10 +1530,10 @@ auto* src = R"( $B1: { # root - %o0:f16 = override @id(0) - %o1:f16 = override @id(1) - %o2:f16 = override @id(2) - %o3:f16 = override @id(3) + %o0:f16 = override undef @id(0) + %o1:f16 = override undef @id(1) + %o2:f16 = override undef @id(2) + %o3:f16 = override undef @id(3) %5:vec4<f16> = construct %o0, %o1, %o2, %o3 %6:vec4<f16> = ceil %5 %global:ptr<private, vec4<f16>, read_write> = var %6 @@ -1581,7 +1581,7 @@ }); auto* src = R"( $B1: { # root - %x:u32 = override @id(2) + %x:u32 = override undef @id(2) } )"; @@ -1603,7 +1603,7 @@ }); auto* src = R"( $B1: { # root - %x:i32 = override @id(2) + %x:i32 = override undef @id(2) } )"; @@ -1625,7 +1625,7 @@ }); auto* src = R"( $B1: { # root - %x:f32 = override @id(2) + %x:f32 = override undef @id(2) } )"; @@ -1647,7 +1647,7 @@ }); auto* src = R"( $B1: { # root - %x:f16 = override @id(2) + %x:f16 = override undef @id(2) } )";
diff --git a/src/tint/lang/core/ir/validator_call_test.cc b/src/tint/lang/core/ir/validator_call_test.cc index da3f59f..b4e2ca9 100644 --- a/src/tint/lang/core/ir/validator_call_test.cc +++ b/src/tint/lang/core/ir/validator_call_test.cc
@@ -225,7 +225,7 @@ ASSERT_NE(res, Success); EXPECT_THAT(res.Failure().reason.Str(), testing::HasSubstr(R"(:8:15 error: call: expected at least 1 operands, got 0 - %3:void = call undef + %3:void = call ^^^^ )")) << res.Failure().reason.Str(); }
diff --git a/src/tint/lang/core/ir/validator_test.cc b/src/tint/lang/core/ir/validator_test.cc index 93fdd9e..df86834 100644 --- a/src/tint/lang/core/ir/validator_test.cc +++ b/src/tint/lang/core/ir/validator_test.cc
@@ -1208,7 +1208,7 @@ res.Failure().reason.Str(), testing::HasSubstr( R"(:2:12 error: override: root block: invalid instruction: tint::core::ir::Override - %a:u32 = override, 1u @id(0) + %a:u32 = override 1u @id(0) ^^^^^^^^ )")) << res.Failure().reason.Str(); } @@ -1255,7 +1255,7 @@ EXPECT_THAT( res.Failure().reason.Str(), testing::HasSubstr(R"(:2:18 error: override: override type 'vec3<u32>' is not a scalar - %1:vec3<u32> = override @id(0) + %1:vec3<u32> = override undef @id(0) ^^^^^^^^ )")) << res.Failure().reason.Str(); } @@ -1273,7 +1273,7 @@ res.Failure().reason.Str(), testing::HasSubstr( R"(:2:12 error: override: override type 'u32' does not match initializer type 'i32' - %1:u32 = override, 1i @id(0) + %1:u32 = override 1i @id(0) ^^^^^^^^ )")) << res.Failure().reason.Str(); } @@ -1291,7 +1291,7 @@ ASSERT_NE(res, Success); EXPECT_THAT(res.Failure().reason.Str(), testing::HasSubstr(R"(:3:12 error: override: duplicate override id encountered: 2 - %2:i32 = override @id(2) + %2:i32 = override undef @id(2) ^^^^^^^^ )")) << res.Failure().reason.Str(); }
diff --git a/src/tint/lang/wgsl/reader/program_to_ir/program_to_ir_test.cc b/src/tint/lang/wgsl/reader/program_to_ir/program_to_ir_test.cc index fdbde28..58f9a0e 100644 --- a/src/tint/lang/wgsl/reader/program_to_ir/program_to_ir_test.cc +++ b/src/tint/lang/wgsl/reader/program_to_ir/program_to_ir_test.cc
@@ -1195,7 +1195,7 @@ EXPECT_EQ(m.SourceOf(override).range.begin, loc); EXPECT_EQ(core::ir::Disassembler(m).Plain(), R"($B1: { # root - %a:i32 = override @id(0) + %a:i32 = override undef @id(0) } )"); @@ -1218,7 +1218,7 @@ EXPECT_FLOAT_EQ(1.0f, init->Value()->ValueAs<float>()); EXPECT_EQ(core::ir::Disassembler(m).Plain(), R"($B1: { # root - %a:f32 = override, 1.0f @id(0) + %a:f32 = override 1.0f @id(0) } )"); @@ -1241,7 +1241,7 @@ EXPECT_EQ(3u, init->Value()->ValueAs<uint32_t>()); EXPECT_EQ(core::ir::Disassembler(m).Plain(), R"($B1: { # root - %a:u32 = override, 3u @id(0) + %a:u32 = override 3u @id(0) } )"); @@ -1258,8 +1258,8 @@ auto m = res.Move(); EXPECT_EQ(core::ir::Disassembler(m).Plain(), R"($B1: { # root - %a:u32 = override @id(0) - %b:bool = override @id(1) + %a:u32 = override undef @id(0) + %b:bool = override undef @id(1) %3:bool = constexpr_if %b [t: $B2, f: $B3] { # constexpr_if_1 $B2: { # true %4:u32 = div 1u, %a @@ -1270,7 +1270,7 @@ exit_if false # constexpr_if_1 } } - %c:bool = override, %3 @id(2) + %c:bool = override %3 @id(2) } )"); @@ -1290,8 +1290,8 @@ ASSERT_EQ(1u, m.functions.Length()); EXPECT_EQ(core::ir::Disassembler(m).Plain(), R"($B1: { # root - %a:u32 = override @id(0) - %b:bool = override @id(1) + %a:u32 = override undef @id(0) + %b:bool = override undef @id(1) } %test_function = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -1327,7 +1327,7 @@ ASSERT_EQ(1u, m.functions.Length()); EXPECT_EQ(core::ir::Disassembler(m).Plain(), R"($B1: { # root - %a:u32 = override @id(0) + %a:u32 = override undef @id(0) } %test_function = @compute @workgroup_size(1u, 1u, 1u) func():void { @@ -1359,9 +1359,9 @@ auto m = res.Move(); EXPECT_EQ(core::ir::Disassembler(m).Plain(), R"($B1: { # root - %z:u32 = override @id(0) + %z:u32 = override undef @id(0) %2:u32 = add %z, 2u - %a:u32 = override, %2 @id(1) + %a:u32 = override %2 @id(1) } )"); @@ -1381,7 +1381,7 @@ auto m = res.Move(); EXPECT_EQ(core::ir::Disassembler(m).Plain(), R"($B1: { # root - %x:i32 = override, 1i @id(0) + %x:i32 = override 1i @id(0) %arr:ptr<workgroup, array<u32, %x>, read_write> = var undef }