tint/ir: Add an initializer field to Var

Fixed: tint:1930
Change-Id: I16b4414a2e575a4fb60c70d4eec7a1d69db8818b
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/131742
Commit-Queue: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>

src/tint/ir/builder_impl.cc

@@ -775,9 +775,7 @@ void BuilderImpl::EmitVariable(const ast::Variable* var) {
                 if (!init) {
                     return;
                 }
-
-                auto* store = builder.Store(val, init.Get());
-                current_flow_block->instructions.Push(store);
+                val->initializer = init.Get();
             }
             // Store the declaration so we can get the instruction to store too
             scopes_.Set(v->name->symbol, val);

src/tint/ir/builder_impl_binary_test.cc

@@ -54,7 +54,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundAdd) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -96,7 +96,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundSubtract) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -138,7 +138,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundMultiply) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -180,7 +180,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundDiv) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -222,7 +222,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundModulo) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -264,7 +264,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundAnd) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, bool, read_write> = var private read_write
+%v1:ref<private, bool, read_write> = var private, read_write
 
 
 
@@ -306,7 +306,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundOr) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, bool, read_write> = var private read_write
+%v1:ref<private, bool, read_write> = var private, read_write
 
 
 
@@ -348,7 +348,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundXor) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -379,7 +379,7 @@ func_end
 %fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn4 = block
   %1:bool = call my_func
-  %tint_symbol:bool = var function read_write
+  %tint_symbol:bool = var function, read_write
   store %tint_symbol:bool, %1:bool
   branch %fn5
 
@@ -414,7 +414,7 @@ func_end
 %fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn4 = block
   %1:bool = call my_func
-  %tint_symbol:bool = var function read_write
+  %tint_symbol:bool = var function, read_write
   store %tint_symbol:bool, %1:bool
   branch %fn5
 
@@ -569,7 +569,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundShiftLeft) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -611,7 +611,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundShiftRight) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -645,7 +645,7 @@ func_end
   %fn4 = block
   %1:f32 = call my_func
   %2:bool = lt %1:f32, 2.0f
-  %tint_symbol:bool = var function read_write
+  %tint_symbol:bool = var function, read_write
   store %tint_symbol:bool, %2:bool
   branch %fn5
 

src/tint/ir/builder_impl_call_test.cc

@@ -92,8 +92,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Convert) {
     ASSERT_TRUE(r);
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%i:ref<private, i32, read_write> = var private read_write
-store %i:ref<private, i32, read_write>, 1i
+%i:ref<private, i32, read_write> = var private, read_write, 1i
 
 
 
@@ -116,8 +115,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_ConstructEmpty) {
     ASSERT_TRUE(r);
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%i:ref<private, vec3<f32>, read_write> = var private read_write
-store %i:ref<private, vec3<f32>, read_write>, vec3<f32> 0.0f
+%i:ref<private, vec3<f32>, read_write> = var private, read_write, vec3<f32> 0.0f
 
 
 
@@ -135,8 +133,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Construct) {
     ASSERT_TRUE(r);
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%i:ref<private, f32, read_write> = var private read_write
-store %i:ref<private, f32, read_write>, 1.0f
+%i:ref<private, f32, read_write> = var private, read_write, 1.0f
 
 
 

src/tint/ir/builder_impl_store_test.cc

@@ -37,7 +37,7 @@ TEST_F(IR_BuilderImplTest, EmitStatement_Assign) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%a:ref<private, u32, read_write> = var private read_write
+%a:ref<private, u32, read_write> = var private, read_write
 
 
 

src/tint/ir/builder_impl_unary_test.cc

@@ -91,7 +91,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_AddressOf) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, i32, read_write> = var private read_write
+%v1:ref<private, i32, read_write> = var private, read_write
 
 
 
@@ -117,7 +117,7 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Indirection) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%v1:ref<private, i32, read_write> = var private read_write
+%v1:ref<private, i32, read_write> = var private, read_write
 
 
 

src/tint/ir/builder_impl_var_test.cc

@@ -34,7 +34,7 @@ TEST_F(IR_BuilderImplTest, Emit_GlobalVar_NoInit) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%a:ref<private, u32, read_write> = var private read_write
+%a:ref<private, u32, read_write> = var private, read_write
 
 
 
@@ -50,8 +50,7 @@ TEST_F(IR_BuilderImplTest, Emit_GlobalVar_Init) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%a:ref<private, u32, read_write> = var private read_write
-store %a:ref<private, u32, read_write>, 2u
+%a:ref<private, u32, read_write> = var private, read_write, 2u
 
 
 
@@ -69,7 +68,7 @@ TEST_F(IR_BuilderImplTest, Emit_Var_NoInit) {
     EXPECT_EQ(Disassemble(m),
               R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn2 = block
-  %a:ref<function, u32, read_write> = var function read_write
+  %a:ref<function, u32, read_write> = var function, read_write
   ret
 func_end
 
@@ -88,8 +87,7 @@ TEST_F(IR_BuilderImplTest, Emit_Var_Init) {
     EXPECT_EQ(Disassemble(m),
               R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn2 = block
-  %a:ref<function, u32, read_write> = var function read_write
-  store %a:ref<function, u32, read_write>, 2u
+  %a:ref<function, u32, read_write> = var function, read_write, 2u
   ret
 func_end
 

src/tint/ir/disassembler.cc

@@ -406,7 +406,11 @@ void Disassembler::EmitInstruction(const Instruction* inst) {
         },
         [&](const ir::Var* v) {
             EmitValue(v);
-            out_ << " = var " << v->address_space << " " << v->access;
+            out_ << " = var " << v->address_space << ", " << v->access;
+            if (v->initializer) {
+                out_ << ", ";
+                EmitValue(v->initializer);
+            }
         });
 }
 

src/tint/ir/var.h

@@ -42,6 +42,9 @@ class Var : public utils::Castable<Var, Instruction> {
 
     /// The variable access mode
     builtin::Access access = builtin::Access::kUndefined;
+
+    /// The optional initializer
+    Value* initializer = nullptr;
 };
 
 }  // namespace tint::ir