[ir] Drop address space and access from ir::Var.
The address space and access mode both exist on the type of the Var,
storing the information into Var is redundant. This CL removes them from
var in favour of the type fields.
Bug: tint:1718
Change-Id: I0198a8794d6359834592562251b2788e8a0347ca
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/133163
Commit-Queue: Dan Sinclair <dsinclair@chromium.org>
Reviewed-by: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
diff --git a/src/tint/ir/builder.cc b/src/tint/ir/builder.cc
index d62bb43..74b6f64 100644
--- a/src/tint/ir/builder.cc
+++ b/src/tint/ir/builder.cc
@@ -239,10 +239,8 @@
return ir.instructions.Create<ir::Store>(to, from);
}
-ir::Var* Builder::Declare(const type::Type* type,
- builtin::AddressSpace address_space,
- builtin::Access access) {
- return ir.instructions.Create<ir::Var>(type, address_space, access);
+ir::Var* Builder::Declare(const type::Type* type) {
+ return ir.instructions.Create<ir::Var>(type);
}
} // namespace tint::ir
diff --git a/src/tint/ir/builder.h b/src/tint/ir/builder.h
index b4a2a81..060bbfd 100644
--- a/src/tint/ir/builder.h
+++ b/src/tint/ir/builder.h
@@ -358,12 +358,8 @@
/// Creates a new `var` declaration
/// @param type the var type
- /// @param address_space the address space
- /// @param access the access mode
/// @returns the instruction
- ir::Var* Declare(const type::Type* type,
- builtin::AddressSpace address_space,
- builtin::Access access);
+ ir::Var* Declare(const type::Type* type);
/// Retrieves the root block for the module, creating if necessary
/// @returns the root block
diff --git a/src/tint/ir/disassembler.cc b/src/tint/ir/disassembler.cc
index d4ee3a7..8a07d54 100644
--- a/src/tint/ir/disassembler.cc
+++ b/src/tint/ir/disassembler.cc
@@ -417,7 +417,7 @@
},
[&](const ir::Var* v) {
EmitValue(v);
- out_ << " = var " << v->address_space << ", " << v->access;
+ out_ << " = var";
if (v->initializer) {
out_ << ", ";
EmitValue(v->initializer);
diff --git a/src/tint/ir/from_program.cc b/src/tint/ir/from_program.cc
index f7729f3..15b92c1 100644
--- a/src/tint/ir/from_program.cc
+++ b/src/tint/ir/from_program.cc
@@ -80,6 +80,7 @@
#include "src/tint/sem/value_expression.h"
#include "src/tint/sem/variable.h"
#include "src/tint/switch.h"
+#include "src/tint/type/reference.h"
#include "src/tint/type/void.h"
#include "src/tint/utils/defer.h"
#include "src/tint/utils/result.h"
@@ -843,7 +844,7 @@
var,
[&](const ast::Var* v) {
auto* ty = sem->Type()->Clone(clone_ctx_.type_ctx);
- auto* val = builder_.Declare(ty, sem->AddressSpace(), sem->Access());
+ auto* val = builder_.Declare(ty);
current_flow_block_->instructions.Push(val);
if (v->initializer) {
@@ -944,9 +945,10 @@
}
// Generate a variable to store the short-circut into
- auto* ty = builder_.ir.types.Get<type::Bool>();
- auto* result_var =
- builder_.Declare(ty, builtin::AddressSpace::kFunction, builtin::Access::kReadWrite);
+ auto* ty = builder_.ir.types.Get<type::Reference>(builder_.ir.types.Get<type::Bool>(),
+ builtin::AddressSpace::kFunction,
+ builtin::Access::kReadWrite);
+ auto* result_var = builder_.Declare(ty);
current_flow_block_->instructions.Push(result_var);
auto* lhs_store = builder_.Store(result_var, lhs.Get());
diff --git a/src/tint/ir/from_program_binary_test.cc b/src/tint/ir/from_program_binary_test.cc
index 33266cb..ab8b5c6 100644
--- a/src/tint/ir/from_program_binary_test.cc
+++ b/src/tint/ir/from_program_binary_test.cc
@@ -58,7 +58,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -81,7 +81,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -127,7 +127,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, i32, read_write> = var private, read_write
+ %v1:ref<private, i32, read_write> = var
}
@@ -150,7 +150,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -196,7 +196,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -242,7 +242,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -288,7 +288,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -334,7 +334,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, bool, read_write> = var private, read_write
+ %v1:ref<private, bool, read_write> = var
}
@@ -380,7 +380,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, bool, read_write> = var private, read_write
+ %v1:ref<private, bool, read_write> = var
}
@@ -426,7 +426,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -456,14 +456,14 @@
%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
- store %tint_symbol:bool, %1:bool
+ %tint_symbol:ref<function, bool, read_write> = var
+ store %tint_symbol:ref<function, bool, read_write>, %1:bool
} -> %fn5 # branch
%fn5 = if %1:bool [t: %fn6, f: %fn7, m: %fn8]
# true branch
%fn6 = block {
- store %tint_symbol:bool, false
+ store %tint_symbol:ref<function, bool, read_write>, false
} -> %fn8 # branch
# if merge
@@ -490,15 +490,15 @@
%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
- store %tint_symbol:bool, %1:bool
+ %tint_symbol:ref<function, bool, read_write> = var
+ store %tint_symbol:ref<function, bool, read_write>, %1:bool
} -> %fn5 # branch
%fn5 = if %1:bool [t: %fn6, f: %fn7, m: %fn8]
# true branch
# false branch
%fn7 = block {
- store %tint_symbol:bool, true
+ store %tint_symbol:ref<function, bool, read_write>, true
} -> %fn8 # branch
# if merge
@@ -679,7 +679,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -725,7 +725,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, u32, read_write> = var private, read_write
+ %v1:ref<private, u32, read_write> = var
}
@@ -758,8 +758,8 @@
%fn4 = block {
%1:f32 = call my_func
%2:bool = lt %1:f32, 2.0f
- %tint_symbol:bool = var function, read_write
- store %tint_symbol:bool, %2:bool
+ %tint_symbol:ref<function, bool, read_write> = var
+ store %tint_symbol:ref<function, bool, read_write>, %2:bool
} -> %fn5 # branch
%fn5 = if %2:bool [t: %fn6, f: %fn7, m: %fn8]
@@ -770,7 +770,7 @@
%6:f32 = mul 2.29999995231628417969f, %5:f32
%7:f32 = div %4:f32, %6:f32
%8:bool = gt 2.5f, %7:f32
- store %tint_symbol:bool, %8:bool
+ store %tint_symbol:ref<function, bool, read_write>, %8:bool
} -> %fn8 # branch
# if merge
diff --git a/src/tint/ir/from_program_call_test.cc b/src/tint/ir/from_program_call_test.cc
index ae217d1..18fe3b1 100644
--- a/src/tint/ir/from_program_call_test.cc
+++ b/src/tint/ir/from_program_call_test.cc
@@ -100,7 +100,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %i:ref<private, i32, read_write> = var private, read_write, 1i
+ %i:ref<private, i32, read_write> = var, 1i
}
@@ -121,7 +121,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %i:ref<private, vec3<f32>, read_write> = var private, read_write, vec3<f32> 0.0f
+ %i:ref<private, vec3<f32>, read_write> = var, vec3<f32> 0.0f
}
@@ -137,7 +137,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %i:ref<private, f32, read_write> = var private, read_write, 1.0f
+ %i:ref<private, f32, read_write> = var, 1.0f
}
diff --git a/src/tint/ir/from_program_store_test.cc b/src/tint/ir/from_program_store_test.cc
index 4e89dbe..58fbe4a 100644
--- a/src/tint/ir/from_program_store_test.cc
+++ b/src/tint/ir/from_program_store_test.cc
@@ -36,7 +36,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %a:ref<private, u32, read_write> = var private, read_write
+ %a:ref<private, u32, read_write> = var
}
diff --git a/src/tint/ir/from_program_unary_test.cc b/src/tint/ir/from_program_unary_test.cc
index b2ba2d7..3bcc5b1 100644
--- a/src/tint/ir/from_program_unary_test.cc
+++ b/src/tint/ir/from_program_unary_test.cc
@@ -105,7 +105,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, i32, read_write> = var private, read_write
+ %v1:ref<private, i32, read_write> = var
}
@@ -130,7 +130,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %v1:ref<private, i32, read_write> = var private, read_write
+ %v1:ref<private, i32, read_write> = var
}
diff --git a/src/tint/ir/from_program_var_test.cc b/src/tint/ir/from_program_var_test.cc
index b5379ac..0cc9cbb 100644
--- a/src/tint/ir/from_program_var_test.cc
+++ b/src/tint/ir/from_program_var_test.cc
@@ -33,7 +33,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %a:ref<private, u32, read_write> = var private, read_write
+ %a:ref<private, u32, read_write> = var
}
@@ -48,7 +48,7 @@
ASSERT_TRUE(m) << (!m ? m.Failure() : "");
EXPECT_EQ(Disassemble(m.Get()), R"(%fn1 = block {
- %a:ref<private, u32, read_write> = var private, read_write, 2u
+ %a:ref<private, u32, read_write> = var, 2u
}
@@ -65,7 +65,7 @@
EXPECT_EQ(Disassemble(m.Get()),
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
} -> %func_end # return
} %func_end
@@ -83,7 +83,7 @@
EXPECT_EQ(Disassemble(m.Get()),
R"(%fn1 = func test_function():void [@compute @workgroup_size(1, 1, 1)] {
%fn2 = block {
- %a:ref<function, u32, read_write> = var function, read_write, 2u
+ %a:ref<function, u32, read_write> = var, 2u
} -> %func_end # return
} %func_end
diff --git a/src/tint/ir/module_test.cc b/src/tint/ir/module_test.cc
index c9b1150..9cd36ae 100644
--- a/src/tint/ir/module_test.cc
+++ b/src/tint/ir/module_test.cc
@@ -25,23 +25,20 @@
TEST_F(IR_ModuleTest, NameOfUnnamed) {
Module mod;
- auto* v = mod.values.Create<ir::Var>(
- mod.types.Get<type::I32>(), builtin::AddressSpace::kUndefined, builtin::Access::kUndefined);
+ auto* v = mod.values.Create<ir::Var>(mod.types.Get<type::I32>());
EXPECT_FALSE(mod.NameOf(v).IsValid());
}
TEST_F(IR_ModuleTest, SetName) {
Module mod;
- auto* v = mod.values.Create<ir::Var>(
- mod.types.Get<type::I32>(), builtin::AddressSpace::kUndefined, builtin::Access::kUndefined);
+ auto* v = mod.values.Create<ir::Var>(mod.types.Get<type::I32>());
EXPECT_EQ(mod.SetName(v, "a").Name(), "a");
EXPECT_EQ(mod.NameOf(v).Name(), "a");
}
TEST_F(IR_ModuleTest, SetNameRename) {
Module mod;
- auto* v = mod.values.Create<ir::Var>(
- mod.types.Get<type::I32>(), builtin::AddressSpace::kUndefined, builtin::Access::kUndefined);
+ auto* v = mod.values.Create<ir::Var>(mod.types.Get<type::I32>());
EXPECT_EQ(mod.SetName(v, "a").Name(), "a");
EXPECT_EQ(mod.SetName(v, "b").Name(), "b");
EXPECT_EQ(mod.NameOf(v).Name(), "b");
@@ -49,12 +46,9 @@
TEST_F(IR_ModuleTest, SetNameCollision) {
Module mod;
- auto* a = mod.values.Create<ir::Var>(
- mod.types.Get<type::I32>(), builtin::AddressSpace::kUndefined, builtin::Access::kUndefined);
- auto* b = mod.values.Create<ir::Var>(
- mod.types.Get<type::I32>(), builtin::AddressSpace::kUndefined, builtin::Access::kUndefined);
- auto* c = mod.values.Create<ir::Var>(
- mod.types.Get<type::I32>(), builtin::AddressSpace::kUndefined, builtin::Access::kUndefined);
+ auto* a = mod.values.Create<ir::Var>(mod.types.Get<type::I32>());
+ auto* b = mod.values.Create<ir::Var>(mod.types.Get<type::I32>());
+ auto* c = mod.values.Create<ir::Var>(mod.types.Get<type::I32>());
EXPECT_EQ(mod.SetName(a, "x").Name(), "x");
EXPECT_EQ(mod.SetName(b, "x_1").Name(), "x_1");
EXPECT_EQ(mod.SetName(c, "x").Name(), "x_2");
diff --git a/src/tint/ir/to_program.cc b/src/tint/ir/to_program.cc
index 94b0f59..4d38015 100644
--- a/src/tint/ir/to_program.cc
+++ b/src/tint/ir/to_program.cc
@@ -14,6 +14,7 @@
#include "src/tint/ir/to_program.h"
+#include <string>
#include <utility>
#include "src/tint/ir/block.h"
@@ -271,7 +272,12 @@
const ast::VariableDeclStatement* Var(const ir::Var* var) {
Symbol name = NameOf(var);
- auto ty = Type(var->Type());
+ auto* ptr = var->Type()->As<type::Reference>();
+ if (!ptr) {
+ Err("Incorrect type for var");
+ return nullptr;
+ }
+ auto ty = Type(ptr);
const ast::Expression* init = nullptr;
if (var->initializer) {
init = Expr(var->initializer);
@@ -279,13 +285,13 @@
return nullptr;
}
}
- switch (var->address_space) {
+ switch (ptr->AddressSpace()) {
case builtin::AddressSpace::kFunction:
return b.Decl(b.Var(name, ty.Get(), init));
case builtin::AddressSpace::kStorage:
- return b.Decl(b.Var(name, ty.Get(), init, var->access, var->address_space));
+ return b.Decl(b.Var(name, ty.Get(), init, ptr->Access(), ptr->AddressSpace()));
default:
- return b.Decl(b.Var(name, ty.Get(), init, var->address_space));
+ return b.Decl(b.Var(name, ty.Get(), init, ptr->AddressSpace()));
}
}
@@ -445,7 +451,7 @@
Symbol Sym(const Symbol& s) { return b.Symbols().Register(s.NameView()); }
- // void Err(std::string str) { b.Diagnostics().add_error(diag::System::IR, std::move(str)); }
+ void Err(std::string str) { b.Diagnostics().add_error(diag::System::IR, std::move(str)); }
};
} // namespace
diff --git a/src/tint/ir/var.cc b/src/tint/ir/var.cc
index e33da54..9bf4b8d 100644
--- a/src/tint/ir/var.cc
+++ b/src/tint/ir/var.cc
@@ -19,8 +19,7 @@
namespace tint::ir {
-Var::Var(const type::Type* ty, builtin::AddressSpace addr_space, builtin::Access acc)
- : type(ty), address_space(addr_space), access(acc) {}
+Var::Var(const type::Type* ty) : type(ty) {}
Var::~Var() = default;
diff --git a/src/tint/ir/var.h b/src/tint/ir/var.h
index 5a61104..c874a62 100644
--- a/src/tint/ir/var.h
+++ b/src/tint/ir/var.h
@@ -27,9 +27,7 @@
public:
/// Constructor
/// @param type the type of the var
- /// @param address_space the address space of the var
- /// @param access the access mode of the var
- Var(const type::Type* type, builtin::AddressSpace address_space, builtin::Access access);
+ explicit Var(const type::Type* type);
Var(const Var& inst) = delete;
Var(Var&& inst) = delete;
~Var() override;
@@ -43,12 +41,6 @@
/// the result type of the instruction
const type::Type* type = nullptr;
- /// The variable address space
- builtin::AddressSpace address_space = builtin::AddressSpace::kUndefined;
-
- /// The variable access mode
- builtin::Access access = builtin::Access::kUndefined;
-
/// The optional initializer
Value* initializer = nullptr;
};
