From 88417687fac18974ee168271f3e8e526c86e4ffa Mon Sep 17 00:00:00 2001
From: Ben Clayton <bclayton@google.com>
Date: Wed, 10 May 2023 09:44:10 +0000
Subject: [PATCH] tint/ir: Add name metadata to Module

Register the name of 'var' and 'let' variables when building the IR.
Use the names of these variables when printing the disassembly.

Change-Id: I56ee24a29333f1bc8f97459bc1cfca5c3a59e79d
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/131741
Auto-Submit: Ben Clayton <bclayton@google.com>
Kokoro: Kokoro <noreply+kokoro@google.com>
Reviewed-by: James Price <jrprice@google.com>
Commit-Queue: Ben Clayton <bclayton@google.com>
---
 src/tint/BUILD.gn                       |  1 +
 src/tint/CMakeLists.txt                 |  1 +
 src/tint/ir/builder_impl.cc             |  6 ++
 src/tint/ir/builder_impl_binary_test.cc | 80 ++++++++++++-------------
 src/tint/ir/builder_impl_call_test.cc   | 16 ++---
 src/tint/ir/builder_impl_store_test.cc  |  4 +-
 src/tint/ir/builder_impl_unary_test.cc  | 10 ++--
 src/tint/ir/builder_impl_var_test.cc    | 12 ++--
 src/tint/ir/disassembler.cc             |  7 ++-
 src/tint/ir/module.cc                   | 31 ++++++++++
 src/tint/ir/module.h                    | 21 +++++--
 src/tint/ir/module_test.cc              | 66 ++++++++++++++++++++
 src/tint/symbol.h                       |  3 +
 13 files changed, 192 insertions(+), 66 deletions(-)
 create mode 100644 src/tint/ir/module_test.cc

diff --git a/src/tint/BUILD.gn b/src/tint/BUILD.gn
index 92edf1239d..0a6f186773 100644
--- a/src/tint/BUILD.gn
+++ b/src/tint/BUILD.gn
@@ -2161,6 +2161,7 @@ if (tint_build_unittests) {
       "ir/builder_impl_var_test.cc",
       "ir/constant_test.cc",
       "ir/discard_test.cc",
+      "ir/module_test.cc",
       "ir/store_test.cc",
       "ir/test_helper.h",
       "ir/unary_test.cc",
diff --git a/src/tint/CMakeLists.txt b/src/tint/CMakeLists.txt
index 60b9885456..0b06c666ef 100644
--- a/src/tint/CMakeLists.txt
+++ b/src/tint/CMakeLists.txt
@@ -1457,6 +1457,7 @@ if(TINT_BUILD_TESTS)
       ir/builder_impl_var_test.cc
       ir/constant_test.cc
       ir/discard_test.cc
+      ir/module_test.cc
       ir/store_test.cc
       ir/test_helper.h
       ir/unary_test.cc
diff --git a/src/tint/ir/builder_impl.cc b/src/tint/ir/builder_impl.cc
index ef4a94ec37..ecf7a30029 100644
--- a/src/tint/ir/builder_impl.cc
+++ b/src/tint/ir/builder_impl.cc
@@ -781,6 +781,9 @@ void BuilderImpl::EmitVariable(const ast::Variable* var) {
             }
             // Store the declaration so we can get the instruction to store too
             scopes_.Set(v->name->symbol, val);
+
+            // Record the original name of the var
+            builder.ir.SetName(val, v->name->symbol.Name());
         },
         [&](const ast::Let* l) {
             // A `let` doesn't exist as a standalone item in the IR, it's just the result of the
@@ -792,6 +795,9 @@ void BuilderImpl::EmitVariable(const ast::Variable* var) {
 
             // Store the results of the initialization
             scopes_.Set(l->name->symbol, init.Get());
+
+            // Record the original name of the let
+            builder.ir.SetName(init.Get(), l->name->symbol.Name());
         },
         [&](const ast::Override*) {
             add_error(var->source,
diff --git a/src/tint/ir/builder_impl_binary_test.cc b/src/tint/ir/builder_impl_binary_test.cc
index ed131595d9..20c5cca96f 100644
--- a/src/tint/ir/builder_impl_binary_test.cc
+++ b/src/tint/ir/builder_impl_binary_test.cc
@@ -54,14 +54,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundAdd) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, u32, read_write> = add %1:ref<private, u32, read_write>, 1u
-  store %1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
+  %2:ref<private, u32, read_write> = add %v1:ref<private, u32, read_write>, 1u
+  store %v1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
   ret
 func_end
 
@@ -96,14 +96,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundSubtract) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, u32, read_write> = sub %1:ref<private, u32, read_write>, 1u
-  store %1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
+  %2:ref<private, u32, read_write> = sub %v1:ref<private, u32, read_write>, 1u
+  store %v1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
   ret
 func_end
 
@@ -138,14 +138,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundMultiply) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, u32, read_write> = mul %1:ref<private, u32, read_write>, 1u
-  store %1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
+  %2:ref<private, u32, read_write> = mul %v1:ref<private, u32, read_write>, 1u
+  store %v1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
   ret
 func_end
 
@@ -180,14 +180,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundDiv) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, u32, read_write> = div %1:ref<private, u32, read_write>, 1u
-  store %1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
+  %2:ref<private, u32, read_write> = div %v1:ref<private, u32, read_write>, 1u
+  store %v1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
   ret
 func_end
 
@@ -222,14 +222,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundModulo) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, u32, read_write> = mod %1:ref<private, u32, read_write>, 1u
-  store %1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
+  %2:ref<private, u32, read_write> = mod %v1:ref<private, u32, read_write>, 1u
+  store %v1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
   ret
 func_end
 
@@ -264,14 +264,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundAnd) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, bool, read_write> = var private read_write
+%v1:ref<private, bool, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, bool, read_write> = and %1:ref<private, bool, read_write>, false
-  store %1:ref<private, bool, read_write>, %2:ref<private, bool, read_write>
+  %2:ref<private, bool, read_write> = and %v1:ref<private, bool, read_write>, false
+  store %v1:ref<private, bool, read_write>, %2:ref<private, bool, read_write>
   ret
 func_end
 
@@ -306,14 +306,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundOr) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, bool, read_write> = var private read_write
+%v1:ref<private, bool, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, bool, read_write> = or %1:ref<private, bool, read_write>, false
-  store %1:ref<private, bool, read_write>, %2:ref<private, bool, read_write>
+  %2:ref<private, bool, read_write> = or %v1:ref<private, bool, read_write>, false
+  store %v1:ref<private, bool, read_write>, %2:ref<private, bool, read_write>
   ret
 func_end
 
@@ -348,14 +348,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundXor) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, u32, read_write> = xor %1:ref<private, u32, read_write>, 1u
-  store %1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
+  %2:ref<private, u32, read_write> = xor %v1:ref<private, u32, read_write>, 1u
+  store %v1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
   ret
 func_end
 
@@ -379,14 +379,14 @@ func_end
 %fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn4 = block
   %1:bool = call my_func
-  %2:bool = var function read_write
-  store %2:bool, %1:bool
+  %tint_symbol:bool = var function read_write
+  store %tint_symbol:bool, %1:bool
   branch %fn5
 
   %fn5 = if %1:bool [t: %fn6, f: %fn7, m: %fn8]
     # true branch
     %fn6 = block
-    store %2:bool, false
+    store %tint_symbol:bool, false
     branch %fn8
 
   # if merge
@@ -414,15 +414,15 @@ func_end
 %fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn4 = block
   %1:bool = call my_func
-  %2:bool = var function read_write
-  store %2:bool, %1:bool
+  %tint_symbol:bool = var function read_write
+  store %tint_symbol:bool, %1:bool
   branch %fn5
 
   %fn5 = if %1:bool [t: %fn6, f: %fn7, m: %fn8]
     # true branch
     # false branch
     %fn7 = block
-    store %2:bool, true
+    store %tint_symbol:bool, true
     branch %fn8
 
   # if merge
@@ -569,14 +569,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundShiftLeft) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, u32, read_write> = shiftl %1:ref<private, u32, read_write>, 1u
-  store %1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
+  %2:ref<private, u32, read_write> = shiftl %v1:ref<private, u32, read_write>, 1u
+  store %v1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
   ret
 func_end
 
@@ -611,14 +611,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Binary_CompoundShiftRight) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%v1:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ref<private, u32, read_write> = shiftr %1:ref<private, u32, read_write>, 1u
-  store %1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
+  %2:ref<private, u32, read_write> = shiftr %v1:ref<private, u32, read_write>, 1u
+  store %v1:ref<private, u32, read_write>, %2:ref<private, u32, read_write>
   ret
 func_end
 
@@ -645,8 +645,8 @@ func_end
   %fn4 = block
   %1:f32 = call my_func
   %2:bool = lt %1:f32, 2.0f
-  %3:bool = var function read_write
-  store %3:bool, %2:bool
+  %tint_symbol:bool = var function read_write
+  store %tint_symbol:bool, %2:bool
   branch %fn5
 
   %fn5 = if %2:bool [t: %fn6, f: %fn7, m: %fn8]
@@ -657,7 +657,7 @@ func_end
     %6:f32 = mul 2.29999995231628417969f, %5:f32
     %7:f32 = div %4:f32, %6:f32
     %8:bool = gt 2.5f, %7:f32
-    store %3:bool, %8:bool
+    store %tint_symbol:bool, %8:bool
     branch %fn8
 
   # if merge
@@ -685,7 +685,7 @@ func_end
 
 %fn3 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn4 = block
-  %1:bool = call my_func, false
+  %tint_symbol:bool = call my_func, false
   ret
 func_end
 
diff --git a/src/tint/ir/builder_impl_call_test.cc b/src/tint/ir/builder_impl_call_test.cc
index 6339626109..2e63a212ec 100644
--- a/src/tint/ir/builder_impl_call_test.cc
+++ b/src/tint/ir/builder_impl_call_test.cc
@@ -92,14 +92,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Convert) {
     ASSERT_TRUE(r);
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, i32, read_write> = var private read_write
-store %1:ref<private, i32, read_write>, 1i
+%i:ref<private, i32, read_write> = var private read_write
+store %i:ref<private, i32, read_write>, 1i
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:f32 = convert i32, %1:ref<private, i32, read_write>
+  %tint_symbol:f32 = convert i32, %i:ref<private, i32, read_write>
   ret
 func_end
 
@@ -116,8 +116,8 @@ TEST_F(IR_BuilderImplTest, EmitExpression_ConstructEmpty) {
     ASSERT_TRUE(r);
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, vec3<f32>, read_write> = var private read_write
-store %1:ref<private, vec3<f32>, read_write>, vec3<f32> 0.0f
+%i:ref<private, vec3<f32>, read_write> = var private read_write
+store %i:ref<private, vec3<f32>, read_write>, vec3<f32> 0.0f
 
 
 
@@ -135,14 +135,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Construct) {
     ASSERT_TRUE(r);
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, f32, read_write> = var private read_write
-store %1:ref<private, f32, read_write>, 1.0f
+%i:ref<private, f32, read_write> = var private read_write
+store %i:ref<private, f32, read_write>, 1.0f
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:vec3<f32> = construct 2.0f, 3.0f, %1:ref<private, f32, read_write>
+  %tint_symbol:vec3<f32> = construct 2.0f, 3.0f, %i:ref<private, f32, read_write>
   ret
 func_end
 
diff --git a/src/tint/ir/builder_impl_store_test.cc b/src/tint/ir/builder_impl_store_test.cc
index 7cd074c96c..75bb7c0416 100644
--- a/src/tint/ir/builder_impl_store_test.cc
+++ b/src/tint/ir/builder_impl_store_test.cc
@@ -37,13 +37,13 @@ TEST_F(IR_BuilderImplTest, EmitStatement_Assign) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
+%a:ref<private, u32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  store %1:ref<private, u32, read_write>, 4u
+  store %a:ref<private, u32, read_write>, 4u
   ret
 func_end
 
diff --git a/src/tint/ir/builder_impl_unary_test.cc b/src/tint/ir/builder_impl_unary_test.cc
index 091493d72f..16ec6f9dda 100644
--- a/src/tint/ir/builder_impl_unary_test.cc
+++ b/src/tint/ir/builder_impl_unary_test.cc
@@ -91,13 +91,13 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_AddressOf) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, i32, read_write> = var private read_write
+%v1:ref<private, i32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ptr<private, i32, read_write> = addr_of %1:ref<private, i32, read_write>
+  %v2:ptr<private, i32, read_write> = addr_of %v1:ref<private, i32, read_write>
   ret
 func_end
 
@@ -117,14 +117,14 @@ TEST_F(IR_BuilderImplTest, EmitExpression_Unary_Indirection) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, i32, read_write> = var private read_write
+%v1:ref<private, i32, read_write> = var private read_write
 
 
 
 %fn2 = func test_function():void [@compute @workgroup_size(1, 1, 1)]
   %fn3 = block
-  %2:ptr<private, i32, read_write> = addr_of %1:ref<private, i32, read_write>
-  %3:i32 = indirection %2:ptr<private, i32, read_write>
+  %v3:ptr<private, i32, read_write> = addr_of %v1:ref<private, i32, read_write>
+  %v2:i32 = indirection %v3:ptr<private, i32, read_write>
   ret
 func_end
 
diff --git a/src/tint/ir/builder_impl_var_test.cc b/src/tint/ir/builder_impl_var_test.cc
index d68ce8a60b..5d96a279e7 100644
--- a/src/tint/ir/builder_impl_var_test.cc
+++ b/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
-%1:ref<private, u32, read_write> = var private read_write
+%a:ref<private, u32, read_write> = var private read_write
 
 
 
@@ -50,8 +50,8 @@ TEST_F(IR_BuilderImplTest, Emit_GlobalVar_Init) {
     auto m = r.Move();
 
     EXPECT_EQ(Disassemble(m), R"(%fn1 = block
-%1:ref<private, u32, read_write> = var private read_write
-store %1:ref<private, u32, read_write>, 2u
+%a:ref<private, u32, read_write> = var private read_write
+store %a:ref<private, u32, read_write>, 2u
 
 
 
@@ -69,7 +69,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
-  %1:ref<function, u32, read_write> = var function read_write
+  %a:ref<function, u32, read_write> = var function read_write
   ret
 func_end
 
@@ -88,8 +88,8 @@ 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
-  %1:ref<function, u32, read_write> = var function read_write
-  store %1:ref<function, u32, read_write>, 2u
+  %a:ref<function, u32, read_write> = var function read_write
+  store %a:ref<function, u32, read_write>, 2u
   ret
 func_end
 
diff --git a/src/tint/ir/disassembler.cc b/src/tint/ir/disassembler.cc
index 50d11cc522..2d8d401fc4 100644
--- a/src/tint/ir/disassembler.cc
+++ b/src/tint/ir/disassembler.cc
@@ -94,7 +94,12 @@ size_t Disassembler::IdOf(const FlowNode* node) {
 
 std::string_view Disassembler::IdOf(const Value* value) {
     TINT_ASSERT(IR, value);
-    return value_ids_.GetOrCreate(value, [&] { return std::to_string(value_ids_.Count()); });
+    return value_ids_.GetOrCreate(value, [&] {
+        if (auto sym = mod_.NameOf(value)) {
+            return sym.Name();
+        }
+        return std::to_string(value_ids_.Count());
+    });
 }
 
 void Disassembler::Walk(const FlowNode* node) {
diff --git a/src/tint/ir/module.cc b/src/tint/ir/module.cc
index 133d364727..acee4aee99 100644
--- a/src/tint/ir/module.cc
+++ b/src/tint/ir/module.cc
@@ -14,6 +14,8 @@
 
 #include "src/tint/ir/module.h"
 
+#include <limits>
+
 namespace tint::ir {
 
 Module::Module() = default;
@@ -24,4 +26,33 @@ Module::~Module() = default;
 
 Module& Module::operator=(Module&&) = default;
 
+Symbol Module::NameOf(const Value* value) const {
+    return value_to_id_.Get(value).value_or(Symbol{});
+}
+
+Symbol Module::SetName(const Value* value, std::string_view name) {
+    TINT_ASSERT(IR, !name.empty());
+
+    if (auto old = value_to_id_.Get(value)) {
+        value_to_id_.Remove(value);
+        id_to_value_.Remove(old.value());
+    }
+
+    auto sym = symbols.Register(name);
+    if (id_to_value_.Add(sym, value)) {
+        value_to_id_.Add(value, sym);
+        return sym;
+    }
+    auto prefix = std::string(name) + "_";
+    for (uint64_t suffix = 1; suffix != std::numeric_limits<uint64_t>::max(); suffix++) {
+        sym = symbols.Register(prefix + std::to_string(suffix));
+        if (id_to_value_.Add(sym, value)) {
+            value_to_id_.Add(value, sym);
+            return sym;
+        }
+    }
+    TINT_ASSERT(IR, false);  // !
+    return Symbol{};
+}
+
 }  // namespace tint::ir
diff --git a/src/tint/ir/module.h b/src/tint/ir/module.h
index ebf9209133..0f89d86b74 100644
--- a/src/tint/ir/module.h
+++ b/src/tint/ir/module.h
@@ -32,6 +32,15 @@ namespace tint::ir {
 
 /// Main module class for the IR.
 class Module {
+    /// Program Id required to create other components
+    ProgramID prog_id_;
+
+    /// Map of value to pre-declared identifier
+    utils::Hashmap<const Value*, Symbol, 32> value_to_id_;
+
+    /// Map of pre-declared identifier to value
+    utils::Hashmap<Symbol, const Value*, 32> id_to_value_;
+
   public:
     /// Constructor
     Module();
@@ -46,11 +55,15 @@ class Module {
     /// @returns a reference to this module
     Module& operator=(Module&& o);
 
-  private:
-    /// Program Id required to create other components
-    ProgramID prog_id_;
+    /// @param value the value
+    /// @return the name of the given value, or an invalid symbol if the value is not named.
+    Symbol NameOf(const Value* value) const;
+
+    /// @param value the value to name.
+    /// @param name the desired name of the value. May be suffixed on collision.
+    /// @return the unique symbol of the given value.
+    Symbol SetName(const Value* value, std::string_view name);
 
-  public:
     /// The flow node allocator
     utils::BlockAllocator<FlowNode> flow_nodes;
     /// The constant allocator
diff --git a/src/tint/ir/module_test.cc b/src/tint/ir/module_test.cc
new file mode 100644
index 0000000000..15d5cb18d4
--- /dev/null
+++ b/src/tint/ir/module_test.cc
@@ -0,0 +1,66 @@
+// Copyright 2023 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+#include "src/tint/ir/module.h"
+#include "src/tint/ir/test_helper.h"
+
+namespace tint::ir {
+namespace {
+
+using namespace tint::number_suffixes;  // NOLINT
+
+using IR_ModuleTest = TestHelper;
+
+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);
+    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);
+    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);
+    EXPECT_EQ(mod.SetName(v, "a").Name(), "a");
+    EXPECT_EQ(mod.SetName(v, "b").Name(), "b");
+    EXPECT_EQ(mod.NameOf(v).Name(), "b");
+}
+
+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);
+    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");
+    EXPECT_EQ(mod.NameOf(a).Name(), "x");
+    EXPECT_EQ(mod.NameOf(b).Name(), "x_1");
+    EXPECT_EQ(mod.NameOf(c).Name(), "x_2");
+}
+
+}  // namespace
+}  // namespace tint::ir
diff --git a/src/tint/symbol.h b/src/tint/symbol.h
index ae2955926a..ed6fe72d88 100644
--- a/src/tint/symbol.h
+++ b/src/tint/symbol.h
@@ -100,6 +100,9 @@ class Symbol {
     /// @returns true if the symbol is valid
     bool IsValid() const { return val_ != static_cast<uint32_t>(-1); }
 
+    /// @returns true if the symbol is valid
+    operator bool() const { return IsValid(); }
+
     /// @returns the value for the symbol
     uint32_t value() const { return val_; }