Convert @id to an expression.

This CL updates the @id AST nodes to store an expression instead
of a literal value. This is in anticipation of the parser updates
for ID expressions.

Bug: tint:1633
Change-Id: I4dd626007dd1f9093999a0236e220ffdbc9e62db
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/100760
Reviewed-by: Ben Clayton <bclayton@google.com>
Commit-Queue: Dan Sinclair <dsinclair@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>

src/tint/BUILD.gn

@@ -1043,7 +1043,6 @@ if (tint_build_unittests) {
       "ast/module_clone_test.cc",
       "ast/module_test.cc",
       "ast/multisampled_texture_test.cc",
-      "ast/override_test.cc",
       "ast/phony_expression_test.cc",
       "ast/pointer_test.cc",
       "ast/return_statement_test.cc",

src/tint/CMakeLists.txt

@@ -736,7 +736,6 @@ if(TINT_BUILD_TESTS)
     ast/module_clone_test.cc
     ast/module_test.cc
     ast/multisampled_texture_test.cc
-    ast/override_test.cc
     ast/phony_expression_test.cc
     ast/pointer_test.cc
     ast/return_statement_test.cc

src/tint/ast/id_attribute.cc

@@ -22,7 +22,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::ast::IdAttribute);
 
 namespace tint::ast {
 
-IdAttribute::IdAttribute(ProgramID pid, NodeID nid, const Source& src, uint32_t val)
+IdAttribute::IdAttribute(ProgramID pid, NodeID nid, const Source& src, const ast::Expression* val)
     : Base(pid, nid, src), value(val) {}
 
 IdAttribute::~IdAttribute() = default;
@@ -34,7 +34,8 @@ std::string IdAttribute::Name() const {
 const IdAttribute* IdAttribute::Clone(CloneContext* ctx) const {
     // Clone arguments outside of create() call to have deterministic ordering
     auto src = ctx->Clone(source);
-    return ctx->dst->create<IdAttribute>(src, value);
+    auto* value_ = ctx->Clone(value);
+    return ctx->dst->create<IdAttribute>(src, value_);
 }
 
 }  // namespace tint::ast

src/tint/ast/id_attribute.h

@@ -18,6 +18,7 @@
 #include <string>
 
 #include "src/tint/ast/attribute.h"
+#include "src/tint/ast/expression.h"
 
 namespace tint::ast {
 
@@ -28,8 +29,8 @@ class IdAttribute final : public Castable<IdAttribute, Attribute> {
     /// @param pid the identifier of the program that owns this node
     /// @param nid the unique node identifier
     /// @param src the source of this node
-    /// @param val the numeric id value
+    /// @param val the numeric id value expression
-    IdAttribute(ProgramID pid, NodeID nid, const Source& src, uint32_t val);
+    IdAttribute(ProgramID pid, NodeID nid, const Source& src, const ast::Expression* val);
     ~IdAttribute() override;
 
     /// @returns the WGSL name for the attribute
@@ -41,8 +42,8 @@ class IdAttribute final : public Castable<IdAttribute, Attribute> {
     /// @return the newly cloned node
     const IdAttribute* Clone(CloneContext* ctx) const override;
 
-    /// The id value
+    /// The id value expression
-    const uint32_t value;
+    const ast::Expression* const value;
 };
 
 }  // namespace tint::ast

src/tint/ast/id_attribute_test.cc

@@ -19,11 +19,12 @@
 namespace tint::ast {
 namespace {
 
+using namespace tint::number_suffixes;  // NOLINT
 using IdAttributeTest = TestHelper;
 
 TEST_F(IdAttributeTest, Creation) {
-    auto* d = create<IdAttribute>(12u);
+    auto* d = Id(12_a);
-    EXPECT_EQ(12u, d->value);
+    EXPECT_TRUE(d->value->Is<ast::IntLiteralExpression>());
 }
 
 }  // namespace

src/tint/ast/override.cc

@@ -48,11 +48,4 @@ const Override* Override::Clone(CloneContext* ctx) const {
     return ctx->dst->create<Override>(src, sym, ty, ctor, std::move(attrs));
 }
 
-std::string Override::Identifier(const SymbolTable& symbols) const {
-    if (auto* id = ast::GetAttribute<ast::IdAttribute>(attributes)) {
-        return std::to_string(id->value);
-    }
-    return symbols.NameFor(symbol);
-}
-
 }  // namespace tint::ast

src/tint/ast/override.h

@@ -62,12 +62,6 @@ class Override final : public Castable<Override, Variable> {
     /// @param ctx the clone context
     /// @return the newly cloned node
     const Override* Clone(CloneContext* ctx) const override;
-
-    /// @param symbols the symbol table to retrieve the name from
-    /// @returns the identifier string for the override. If the override has
-    /// an ID attribute, the string is the id-stringified. Otherwise, the ID
-    /// is the symbol.
-    std::string Identifier(const SymbolTable& symbols) const;
 };
 
 }  // namespace tint::ast

src/tint/ast/override_test.cc

@@ -1,35 +0,0 @@
-// Copyright 2022 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/ast/override.h"
-
-#include "src/tint/ast/test_helper.h"
-
-namespace tint::ast {
-namespace {
-
-using OverrideTest = TestHelper;
-
-TEST_F(OverrideTest, Identifier_NoId) {
-    auto* o = Override("o", Expr(f32(1.0)));
-    EXPECT_EQ(std::string("o"), o->Identifier(Symbols()));
-}
-
-TEST_F(OverrideTest, Identifier_WithId) {
-    auto* o = Override("o", Expr(f32(1.0)), Id(4u));
-    EXPECT_EQ(std::string("4"), o->Identifier(Symbols()));
-}
-
-}  // namespace
-}  // namespace tint::ast

src/tint/ast/variable_test.cc

@@ -93,7 +93,7 @@ TEST_F(VariableTest, Assert_DifferentProgramID_Constructor) {
 
 TEST_F(VariableTest, WithAttributes) {
     auto* var = Var("my_var", ty.i32(), StorageClass::kFunction, Location(1u),
-                    Builtin(BuiltinValue::kPosition), Id(1200u));
+                    Builtin(BuiltinValue::kPosition), Id(1200_u));
 
     auto& attributes = var->attributes;
     EXPECT_TRUE(ast::HasAttribute<ast::LocationAttribute>(attributes));

src/tint/inspector/inspector_test.cc

@@ -697,8 +697,8 @@ TEST_F(InspectorGetEntryPointTest, OverrideReferencedByCallee) {
 }
 
 TEST_F(InspectorGetEntryPointTest, OverrideSomeReferenced) {
-    Override("foo", ty.f32(), Id(1));
+    Override("foo", ty.f32(), Id(1_a));
-    Override("bar", ty.f32(), Id(2));
+    Override("bar", ty.f32(), Id(2_a));
     MakePlainGlobalReferenceBodyFunction("callee_func", "foo", ty.f32(), utils::Empty);
     MakeCallerBodyFunction("ep_func", utils::Vector{std::string("callee_func")},
                            utils::Vector{
@@ -789,7 +789,7 @@ TEST_F(InspectorGetEntryPointTest, OverrideUninitialized) {
 
 TEST_F(InspectorGetEntryPointTest, OverrideNumericIDSpecified) {
     Override("foo_no_id", ty.f32());
-    Override("foo_id", ty.f32(), Id(1234));
+    Override("foo_id", ty.f32(), Id(1234_a));
 
     MakePlainGlobalReferenceBodyFunction("no_id_func", "foo_no_id", ty.f32(), utils::Empty);
     MakePlainGlobalReferenceBodyFunction("id_func", "foo_id", ty.f32(), utils::Empty);
@@ -1225,9 +1225,9 @@ INSTANTIATE_TEST_SUITE_P(
             InterpolationType::kFlat, InterpolationSampling::kNone}));
 
 TEST_F(InspectorGetOverrideDefaultValuesTest, Bool) {
-    Override("foo", ty.bool_(), Id(1));
+    Override("foo", ty.bool_(), Id(1_a));
-    Override("bar", ty.bool_(), Expr(true), Id(20));
+    Override("bar", ty.bool_(), Expr(true), Id(20_a));
-    Override("baz", ty.bool_(), Expr(false), Id(300));
+    Override("baz", ty.bool_(), Expr(false), Id(300_a));
 
     Inspector& inspector = Build();
 
@@ -1247,8 +1247,8 @@ TEST_F(InspectorGetOverrideDefaultValuesTest, Bool) {
 }
 
 TEST_F(InspectorGetOverrideDefaultValuesTest, U32) {
-    Override("foo", ty.u32(), Id(1));
+    Override("foo", ty.u32(), Id(1_a));
-    Override("bar", ty.u32(), Expr(42_u), Id(20));
+    Override("bar", ty.u32(), Expr(42_u), Id(20_a));
 
     Inspector& inspector = Build();
 
@@ -1264,9 +1264,9 @@ TEST_F(InspectorGetOverrideDefaultValuesTest, U32) {
 }
 
 TEST_F(InspectorGetOverrideDefaultValuesTest, I32) {
-    Override("foo", ty.i32(), Id(1));
+    Override("foo", ty.i32(), Id(1_a));
-    Override("bar", ty.i32(), Expr(-42_i), Id(20));
+    Override("bar", ty.i32(), Expr(-42_i), Id(20_a));
-    Override("baz", ty.i32(), Expr(42_i), Id(300));
+    Override("baz", ty.i32(), Expr(42_i), Id(300_a));
 
     Inspector& inspector = Build();
 
@@ -1286,10 +1286,10 @@ TEST_F(InspectorGetOverrideDefaultValuesTest, I32) {
 }
 
 TEST_F(InspectorGetOverrideDefaultValuesTest, Float) {
-    Override("foo", ty.f32(), Id(1));
+    Override("foo", ty.f32(), Id(1_a));
-    Override("bar", ty.f32(), Expr(0_f), Id(20));
+    Override("bar", ty.f32(), Expr(0_f), Id(20_a));
-    Override("baz", ty.f32(), Expr(-10_f), Id(300));
+    Override("baz", ty.f32(), Expr(-10_f), Id(300_a));
-    Override("x", ty.f32(), Expr(15_f), Id(4000));
+    Override("x", ty.f32(), Expr(15_f), Id(4000_a));
 
     Inspector& inspector = Build();
 
@@ -1313,9 +1313,9 @@ TEST_F(InspectorGetOverrideDefaultValuesTest, Float) {
 }
 
 TEST_F(InspectorGetConstantNameToIdMapTest, WithAndWithoutIds) {
-    Override("v1", ty.f32(), Id(1));
+    Override("v1", ty.f32(), Id(1_a));
-    Override("v20", ty.f32(), Id(20));
+    Override("v20", ty.f32(), Id(20_a));
-    Override("v300", ty.f32(), Id(300));
+    Override("v300", ty.f32(), Id(300_a));
     auto* a = Override("a", ty.f32());
     auto* b = Override("b", ty.f32());
     auto* c = Override("c", ty.f32());

src/tint/program_builder.h

@@ -2946,26 +2946,32 @@ class ProgramBuilder {
     /// @param id the id value
     /// @returns the override attribute pointer
     const ast::IdAttribute* Id(const Source& source, OverrideId id) {
-        return create<ast::IdAttribute>(source, id.value);
+        return create<ast::IdAttribute>(source, Expr(AInt(id.value)));
     }
 
     /// Creates an ast::IdAttribute with an override identifier
     /// @param id the optional id value
     /// @returns the override attribute pointer
-    const ast::IdAttribute* Id(OverrideId id) { return Id(source_, id); }
+    const ast::IdAttribute* Id(OverrideId id) {
+        return create<ast::IdAttribute>(Expr(AInt(id.value)));
+    }
 
     /// Creates an ast::IdAttribute
     /// @param source the source information
-    /// @param id the id value
+    /// @param id the id value expression
     /// @returns the override attribute pointer
-    const ast::IdAttribute* Id(const Source& source, uint32_t id) {
+    template <typename EXPR>
-        return create<ast::IdAttribute>(source, id);
+    const ast::IdAttribute* Id(const Source& source, EXPR&& id) {
+        return create<ast::IdAttribute>(source, Expr(std::forward<EXPR>(id)));
     }
 
     /// Creates an ast::IdAttribute with an override identifier
-    /// @param id the optional id value
+    /// @param id the optional id value expression
     /// @returns the override attribute pointer
-    const ast::IdAttribute* Id(uint32_t id) { return Id(source_, id); }
+    template <typename EXPR>
+    const ast::IdAttribute* Id(EXPR&& id) {
+        return create<ast::IdAttribute>(Expr(std::forward<EXPR>(id)));
+    }
 
     /// Creates an ast::StageAttribute
     /// @param source the source information

src/tint/reader/spirv/parser_impl.cc

@@ -1366,7 +1366,7 @@ bool ParserImpl::EmitScalarSpecConstants() {
                                          "between 0 and 65535: ID %"
                                       << inst.result_id() << " has SpecId " << id;
                     }
-                    auto* cid = create<ast::IdAttribute>(Source{}, id);
+                    auto* cid = builder_.Id(Source{}, AInt(id));
                     spec_id_decos.Push(cid);
                     break;
                 }

src/tint/reader/wgsl/parser_impl.cc

@@ -3508,7 +3508,9 @@ Maybe<const ast::Attribute*> ParserImpl::attribute() {
             }
             match(Token::Type::kComma);
 
-            return create<ast::IdAttribute>(t.source(), val.value);
+            return create<ast::IdAttribute>(
+                t.source(), create<ast::IntLiteralExpression>(
+                                val.value, ast::IntLiteralExpression::Suffix::kNone));
         });
     }
 

src/tint/reader/wgsl/parser_impl_global_constant_decl_test.cc

@@ -201,7 +201,7 @@ TEST_F(ParserImplTest, GlobalOverrideDecl_WithId) {
 
     auto* override_attr = ast::GetAttribute<ast::IdAttribute>(override->attributes);
     ASSERT_NE(override_attr, nullptr);
-    EXPECT_EQ(override_attr->value, 7u);
+    EXPECT_TRUE(override_attr->value->Is<ast::IntLiteralExpression>());
 }
 
 TEST_F(ParserImplTest, GlobalOverrideDecl_WithId_TrailingComma) {
@@ -231,7 +231,7 @@ TEST_F(ParserImplTest, GlobalOverrideDecl_WithId_TrailingComma) {
 
     auto* override_attr = ast::GetAttribute<ast::IdAttribute>(override->attributes);
     ASSERT_NE(override_attr, nullptr);
-    EXPECT_EQ(override_attr->value, 7u);
+    EXPECT_TRUE(override_attr->value->Is<ast::IntLiteralExpression>());
 }
 
 TEST_F(ParserImplTest, GlobalOverrideDecl_WithoutId) {

src/tint/resolver/attribute_validation_test.cc

@@ -97,7 +97,7 @@ static utils::Vector<const ast::Attribute*, 2> createAttributes(const Source& so
         case AttributeKind::kGroup:
             return {builder.Group(source, 1_a)};
         case AttributeKind::kId:
-            return {builder.create<ast::IdAttribute>(source, 0u)};
+            return {builder.Id(source, 0_a)};
         case AttributeKind::kInterpolate:
             return {builder.Interpolate(source, ast::InterpolationType::kLinear,
                                         ast::InterpolationSampling::kCenter)};
@@ -786,8 +786,8 @@ INSTANTIATE_TEST_SUITE_P(ResolverAttributeValidationTest,
 TEST_F(ConstantAttributeTest, DuplicateAttribute) {
     GlobalConst("a", ty.f32(), Expr(1.23_f),
                 utils::Vector{
-                    create<ast::IdAttribute>(Source{{12, 34}}, 0u),
+                    Id(Source{{12, 34}}, 0_a),
-                    create<ast::IdAttribute>(Source{{56, 78}}, 1u),
+                    Id(Source{{56, 78}}, 1_a),
                 });
 
     EXPECT_FALSE(r()->Resolve());
@@ -829,8 +829,8 @@ INSTANTIATE_TEST_SUITE_P(ResolverAttributeValidationTest,
 TEST_F(OverrideAttributeTest, DuplicateAttribute) {
     Override("a", ty.f32(), Expr(1.23_f),
              utils::Vector{
-                 create<ast::IdAttribute>(Source{{12, 34}}, 0u),
+                 Id(Source{{12, 34}}, 0_a),
-                 create<ast::IdAttribute>(Source{{56, 78}}, 1u),
+                 Id(Source{{56, 78}}, 1_a),
              });
 
     EXPECT_FALSE(r()->Resolve());

src/tint/resolver/override_test.cc

@@ -50,7 +50,7 @@ TEST_F(ResolverOverrideTest, NonOverridable) {
 }
 
 TEST_F(ResolverOverrideTest, WithId) {
-    auto* a = Override("a", ty.f32(), Expr(1_f), Id(7u));
+    auto* a = Override("a", ty.f32(), Expr(1_f), Id(7_u));
 
     EXPECT_TRUE(r()->Resolve()) << r()->error();
 
@@ -69,10 +69,10 @@ TEST_F(ResolverOverrideTest, WithAndWithoutIds) {
     std::vector<ast::Variable*> variables;
     auto* a = Override("a", ty.f32(), Expr(1_f));
     auto* b = Override("b", ty.f32(), Expr(1_f));
-    auto* c = Override("c", ty.f32(), Expr(1_f), Id(2u));
+    auto* c = Override("c", ty.f32(), Expr(1_f), Id(2_u));
-    auto* d = Override("d", ty.f32(), Expr(1_f), Id(4u));
+    auto* d = Override("d", ty.f32(), Expr(1_f), Id(4_u));
     auto* e = Override("e", ty.f32(), Expr(1_f));
-    auto* f = Override("f", ty.f32(), Expr(1_f), Id(1u));
+    auto* f = Override("f", ty.f32(), Expr(1_f), Id(1_u));
 
     EXPECT_TRUE(r()->Resolve()) << r()->error();
 
@@ -86,8 +86,8 @@ TEST_F(ResolverOverrideTest, WithAndWithoutIds) {
 }
 
 TEST_F(ResolverOverrideTest, DuplicateIds) {
-    Override("a", ty.f32(), Expr(1_f), Id(Source{{12, 34}}, 7u));
+    Override("a", ty.f32(), Expr(1_f), Id(Source{{12, 34}}, 7_u));
-    Override("b", ty.f32(), Expr(1_f), Id(Source{{56, 78}}, 7u));
+    Override("b", ty.f32(), Expr(1_f), Id(Source{{56, 78}}, 7_u));
 
     EXPECT_FALSE(r()->Resolve());
 
@@ -96,7 +96,7 @@ TEST_F(ResolverOverrideTest, DuplicateIds) {
 }
 
 TEST_F(ResolverOverrideTest, IdTooLarge) {
-    Override("a", ty.f32(), Expr(1_f), Id(Source{{12, 34}}, 65536u));
+    Override("a", ty.f32(), Expr(1_f), Id(Source{{12, 34}}, 65536_u));
 
     EXPECT_FALSE(r()->Resolve());
 
@@ -106,7 +106,7 @@ TEST_F(ResolverOverrideTest, IdTooLarge) {
 TEST_F(ResolverOverrideTest, F16_TemporallyBan) {
     Enable(ast::Extension::kF16);
 
-    Override(Source{{12, 34}}, "a", ty.f16(), Expr(1_h), Id(1u));
+    Override(Source{{12, 34}}, "a", ty.f16(), Expr(1_h), Id(1_u));
 
     EXPECT_FALSE(r()->Resolve());
 

src/tint/resolver/resolver.cc

@@ -23,6 +23,7 @@
 #include "src/tint/ast/alias.h"
 #include "src/tint/ast/array.h"
 #include "src/tint/ast/assignment_statement.h"
+#include "src/tint/ast/attribute.h"
 #include "src/tint/ast/bitcast_expression.h"
 #include "src/tint/ast/break_statement.h"
 #include "src/tint/ast/call_statement.h"
@@ -437,12 +438,35 @@ sem::Variable* Resolver::Override(const ast::Override* v) {
     auto* sem = builder_->create<sem::GlobalVariable>(
         v, ty, sem::EvaluationStage::kOverride, ast::StorageClass::kNone, ast::Access::kUndefined,
         /* constant_value */ nullptr, sem::BindingPoint{});
+    sem->SetConstructor(rhs);
 
-    if (auto* id = ast::GetAttribute<ast::IdAttribute>(v->attributes)) {
+    if (auto* id_attr = ast::GetAttribute<ast::IdAttribute>(v->attributes)) {
-        sem->SetOverrideId(OverrideId{static_cast<decltype(OverrideId::value)>(id->value)});
+        auto* materialize = Materialize(Expression(id_attr->value));
+        if (!materialize) {
+            return nullptr;
+        }
+        auto* c = materialize->ConstantValue();
+        if (!c) {
+            // TODO(crbug.com/tint/1633): Handle invalid materialization when expressions
+            // are supported.
+            return nullptr;
+        }
+
+        auto value = c->As<uint32_t>();
+        if (value > std::numeric_limits<decltype(OverrideId::value)>::max()) {
+            AddError("override IDs must be between 0 and " +
+                         std::to_string(std::numeric_limits<decltype(OverrideId::value)>::max()),
+                     id_attr->source);
+            return nullptr;
+        }
+
+        auto o = OverrideId{static_cast<decltype(OverrideId::value)>(value)};
+        sem->SetOverrideId(o);
+
+        // Track the constant IDs that are specified in the shader.
+        override_ids_.emplace(o, sem);
     }
 
-    sem->SetConstructor(rhs);
     builder_->Sem().Add(v, sem);
     return sem;
 }
@@ -737,8 +761,8 @@ bool Resolver::AllocateOverridableConstantIds() {
         }
 
         OverrideId id;
-        if (auto* id_attr = ast::GetAttribute<ast::IdAttribute>(override->attributes)) {
+        if (ast::HasAttribute<ast::IdAttribute>(override->attributes)) {
-            id = OverrideId{static_cast<decltype(OverrideId::value)>(id_attr->value)};
+            id = builder_->Sem().Get<sem::GlobalVariable>(override)->OverrideId();
         } else {
             // No ID was specified, so allocate the next available ID.
             while (!ids_exhausted && override_ids_.count(next_id)) {
@@ -777,12 +801,6 @@ sem::GlobalVariable* Resolver::GlobalVariable(const ast::Variable* v) {
 
     for (auto* attr : v->attributes) {
         Mark(attr);
-
-        if (auto* id_attr = attr->As<ast::IdAttribute>()) {
-            // Track the constant IDs that are specified in the shader.
-            override_ids_.emplace(
-                OverrideId{static_cast<decltype(OverrideId::value)>(id_attr->value)}, sem);
-        }
     }
 
     if (!validator_.NoDuplicateAttributes(v->attributes)) {

src/tint/resolver/resolver_test.cc

@@ -1001,9 +1001,9 @@ TEST_F(ResolverTest, Function_WorkgroupSize_OverridableConsts) {
     // @id(2) override depth = 2i;
     // @compute @workgroup_size(width, height, depth)
     // fn main() {}
-    auto* width = Override("width", ty.i32(), Expr(16_i), Id(0));
+    auto* width = Override("width", ty.i32(), Expr(16_i), Id(0_a));
-    auto* height = Override("height", ty.i32(), Expr(8_i), Id(1));
+    auto* height = Override("height", ty.i32(), Expr(8_i), Id(1_a));
-    auto* depth = Override("depth", ty.i32(), Expr(2_i), Id(2));
+    auto* depth = Override("depth", ty.i32(), Expr(2_i), Id(2_a));
     auto* func = Func("main", utils::Empty, ty.void_(), utils::Empty,
                       utils::Vector{
                           Stage(ast::PipelineStage::kCompute),
@@ -1029,9 +1029,9 @@ TEST_F(ResolverTest, Function_WorkgroupSize_OverridableConsts_NoInit) {
     // @id(2) override depth : i32;
     // @compute @workgroup_size(width, height, depth)
     // fn main() {}
-    auto* width = Override("width", ty.i32(), Id(0));
+    auto* width = Override("width", ty.i32(), Id(0_a));
-    auto* height = Override("height", ty.i32(), Id(1));
+    auto* height = Override("height", ty.i32(), Id(1_a));
-    auto* depth = Override("depth", ty.i32(), Id(2));
+    auto* depth = Override("depth", ty.i32(), Id(2_a));
     auto* func = Func("main", utils::Empty, ty.void_(), utils::Empty,
                       utils::Vector{
                           Stage(ast::PipelineStage::kCompute),
@@ -1056,7 +1056,7 @@ TEST_F(ResolverTest, Function_WorkgroupSize_Mixed) {
     // const depth = 3i;
     // @compute @workgroup_size(8, height, depth)
     // fn main() {}
-    auto* height = Override("height", ty.i32(), Expr(2_i), Id(0));
+    auto* height = Override("height", ty.i32(), Expr(2_i), Id(0_a));
     GlobalConst("depth", ty.i32(), Expr(3_i));
     auto* func = Func("main", utils::Empty, ty.void_(), utils::Empty,
                       utils::Vector{

src/tint/resolver/type_validation_test.cc

@@ -75,7 +75,7 @@ TEST_F(ResolverTypeValidationTest, VariableDeclNoConstructor_Pass) {
 
 TEST_F(ResolverTypeValidationTest, GlobalOverrideNoConstructor_Pass) {
     // @id(0) override a :i32;
-    Override(Source{{12, 34}}, "a", ty.i32(), Id(0));
+    Override(Source{{12, 34}}, "a", ty.i32(), Id(0_u));
 
     EXPECT_TRUE(r()->Resolve()) << r()->error();
 }

src/tint/resolver/validator.cc

@@ -776,7 +776,7 @@ bool Validator::Let(const sem::Variable* v) const {
 }
 
 bool Validator::Override(
-    const sem::Variable* v,
+    const sem::GlobalVariable* v,
     const std::unordered_map<OverrideId, const sem::Variable*>& override_ids) const {
     auto* decl = v->Declaration();
     auto* storage_ty = v->Type()->UnwrapRef();
@@ -788,20 +788,11 @@ bool Validator::Override(
     }
 
     for (auto* attr : decl->attributes) {
-        if (auto* id_attr = attr->As<ast::IdAttribute>()) {
+        if (attr->Is<ast::IdAttribute>()) {
-            uint32_t id = id_attr->value;
+            auto id = v->OverrideId();
-            if (id > std::numeric_limits<decltype(OverrideId::value)>::max()) {
+            if (auto it = override_ids.find(id); it != override_ids.end() && it->second != v) {
-                AddError(
-                    "override IDs must be between 0 and " +
-                        std::to_string(std::numeric_limits<decltype(OverrideId::value)>::max()),
-                    attr->source);
-                return false;
-            }
-            if (auto it =
-                    override_ids.find(OverrideId{static_cast<decltype(OverrideId::value)>(id)});
-                it != override_ids.end() && it->second != v) {
                 AddError("override IDs must be unique", attr->source);
-                AddNote("a override with an ID of " + std::to_string(id) +
+                AddNote("a override with an ID of " + std::to_string(id.value) +
                             " was previously declared here:",
                         ast::GetAttribute<ast::IdAttribute>(it->second->Declaration()->attributes)
                             ->source);

src/tint/resolver/validator.h

@@ -376,7 +376,7 @@ class Validator {
     /// @param v the variable to validate
     /// @param override_id the set of override ids in the module
     /// @returns true on success, false otherwise.
-    bool Override(const sem::Variable* v,
+    bool Override(const sem::GlobalVariable* v,
                   const std::unordered_map<OverrideId, const sem::Variable*>& override_id) const;
 
     /// Validates a 'const' variable declaration

src/tint/resolver/variable_validation_test.cc

@@ -98,7 +98,7 @@ TEST_F(ResolverVariableValidationTest, OverrideExceedsIDLimit_LastReserved) {
     // ...
     // @id(N) override oN : i32;
     constexpr size_t kLimit = std::numeric_limits<decltype(OverrideId::value)>::max();
-    Override("reserved", ty.i32(), Id(kLimit));
+    Override("reserved", ty.i32(), Id(AInt(kLimit)));
     for (size_t i = 0; i < kLimit; i++) {
         Override("o" + std::to_string(i), ty.i32());
     }

src/tint/transform/array_length_from_uniform_test.cc

@@ -124,8 +124,9 @@ fn main() {
     auto got = Run<Unshadow, SimplifyPointers, ArrayLengthFromUniform>(src, data);
 
     EXPECT_EQ(expect, str(got));
-    EXPECT_EQ(std::unordered_set<uint32_t>({0}),
+    auto* val = got.data.Get<ArrayLengthFromUniform::Result>();
-              got.data.Get<ArrayLengthFromUniform::Result>()->used_size_indices);
+    ASSERT_NE(val, nullptr);
+    EXPECT_EQ(std::unordered_set<uint32_t>({0}), val->used_size_indices);
 }
 
 TEST_F(ArrayLengthFromUniformTest, BasicInStruct) {

src/tint/writer/glsl/generator_impl_function_test.cc

@@ -804,9 +804,9 @@ void main() {
 
 TEST_F(GlslGeneratorImplTest_Function,
        Emit_Attribute_EntryPoint_Compute_WithWorkgroup_OverridableConst) {
-    Override("width", ty.i32(), Construct(ty.i32(), 2_i), Id(7u));
+    Override("width", ty.i32(), Construct(ty.i32(), 2_i), Id(7_u));
-    Override("height", ty.i32(), Construct(ty.i32(), 3_i), Id(8u));
+    Override("height", ty.i32(), Construct(ty.i32(), 3_i), Id(8_u));
-    Override("depth", ty.i32(), Construct(ty.i32(), 4_i), Id(9u));
+    Override("depth", ty.i32(), Construct(ty.i32(), 4_i), Id(9_u));
     Func("main", utils::Empty, ty.void_(), {},
          utils::Vector{
              Stage(ast::PipelineStage::kCompute),

src/tint/writer/glsl/generator_impl_module_constant_test.cc

@@ -346,7 +346,7 @@ void f() {
 }
 
 TEST_F(GlslGeneratorImplTest_ModuleConstant, Emit_Override) {
-    auto* var = Override("pos", ty.f32(), Expr(3_f), Id(23));
+    auto* var = Override("pos", ty.f32(), Expr(3_f), Id(23_a));
 
     GeneratorImpl& gen = Build();
 
@@ -359,7 +359,7 @@ const float pos = WGSL_SPEC_CONSTANT_23;
 }
 
 TEST_F(GlslGeneratorImplTest_ModuleConstant, Emit_Override_NoConstructor) {
-    auto* var = Override("pos", ty.f32(), Id(23));
+    auto* var = Override("pos", ty.f32(), Id(23_a));
 
     GeneratorImpl& gen = Build();
 
@@ -372,7 +372,7 @@ const float pos = WGSL_SPEC_CONSTANT_23;
 }
 
 TEST_F(GlslGeneratorImplTest_ModuleConstant, Emit_Override_NoId) {
-    auto* a = Override("a", ty.f32(), Expr(3_f), Id(0));
+    auto* a = Override("a", ty.f32(), Expr(3_f), Id(0_a));
     auto* b = Override("b", ty.f32(), Expr(2_f));
 
     GeneratorImpl& gen = Build();

src/tint/writer/hlsl/generator_impl_function_test.cc

@@ -714,9 +714,9 @@ void main() {
 
 TEST_F(HlslGeneratorImplTest_Function,
        Emit_Attribute_EntryPoint_Compute_WithWorkgroup_OverridableConst) {
-    Override("width", ty.i32(), Construct(ty.i32(), 2_i), Id(7u));
+    Override("width", ty.i32(), Construct(ty.i32(), 2_i), Id(7_u));
-    Override("height", ty.i32(), Construct(ty.i32(), 3_i), Id(8u));
+    Override("height", ty.i32(), Construct(ty.i32(), 3_i), Id(8_u));
-    Override("depth", ty.i32(), Construct(ty.i32(), 4_i), Id(9u));
+    Override("depth", ty.i32(), Construct(ty.i32(), 4_i), Id(9_u));
     Func("main", utils::Empty, ty.void_(), utils::Empty,
          utils::Vector{
              Stage(ast::PipelineStage::kCompute),

src/tint/writer/hlsl/generator_impl_module_constant_test.cc

@@ -243,7 +243,7 @@ TEST_F(HlslGeneratorImplTest_ModuleConstant, Emit_GlobalConst_arr_vec2_bool) {
 }
 
 TEST_F(HlslGeneratorImplTest_ModuleConstant, Emit_Override) {
-    auto* var = Override("pos", ty.f32(), Expr(3_f), Id(23));
+    auto* var = Override("pos", ty.f32(), Expr(3_f), Id(23_a));
 
     GeneratorImpl& gen = Build();
 
@@ -256,7 +256,7 @@ static const float pos = WGSL_SPEC_CONSTANT_23;
 }
 
 TEST_F(HlslGeneratorImplTest_ModuleConstant, Emit_Override_NoConstructor) {
-    auto* var = Override("pos", ty.f32(), Id(23));
+    auto* var = Override("pos", ty.f32(), Id(23_a));
 
     GeneratorImpl& gen = Build();
 
@@ -269,7 +269,7 @@ static const float pos = WGSL_SPEC_CONSTANT_23;
 }
 
 TEST_F(HlslGeneratorImplTest_ModuleConstant, Emit_Override_NoId) {
-    auto* a = Override("a", ty.f32(), Expr(3_f), Id(0));
+    auto* a = Override("a", ty.f32(), Expr(3_f), Id(0_a));
     auto* b = Override("b", ty.f32(), Expr(2_f));
 
     GeneratorImpl& gen = Build();

src/tint/writer/msl/generator_impl_module_constant_test.cc

@@ -329,7 +329,7 @@ void f() {
 }
 
 TEST_F(MslGeneratorImplTest, Emit_Override) {
-    auto* var = Override("pos", ty.f32(), Expr(3_f), Id(23));
+    auto* var = Override("pos", ty.f32(), Expr(3_f), Id(23_a));
 
     GeneratorImpl& gen = Build();
 
@@ -338,7 +338,7 @@ TEST_F(MslGeneratorImplTest, Emit_Override) {
 }
 
 TEST_F(MslGeneratorImplTest, Emit_Override_NoId) {
-    auto* var_a = Override("a", ty.f32(), Id(0));
+    auto* var_a = Override("a", ty.f32(), Id(0_a));
     auto* var_b = Override("b", ty.f32());
 
     GeneratorImpl& gen = Build();

src/tint/writer/spirv/builder_function_attribute_test.cc

@@ -150,9 +150,9 @@ TEST_F(BuilderTest, Decoration_ExecutionMode_WorkgroupSize_Const) {
 }
 
 TEST_F(BuilderTest, Decoration_ExecutionMode_WorkgroupSize_OverridableConst) {
-    Override("width", ty.i32(), Construct(ty.i32(), 2_i), Id(7u));
+    Override("width", ty.i32(), Construct(ty.i32(), 2_i), Id(7_u));
-    Override("height", ty.i32(), Construct(ty.i32(), 3_i), Id(8u));
+    Override("height", ty.i32(), Construct(ty.i32(), 3_i), Id(8_u));
-    Override("depth", ty.i32(), Construct(ty.i32(), 4_i), Id(9u));
+    Override("depth", ty.i32(), Construct(ty.i32(), 4_i), Id(9_u));
     auto* func = Func("main", utils::Empty, ty.void_(), utils::Empty,
                       utils::Vector{
                           WorkgroupSize("width", "height", "depth"),
@@ -180,7 +180,7 @@ OpDecorate %3 BuiltIn WorkgroupSize
 }
 
 TEST_F(BuilderTest, Decoration_ExecutionMode_WorkgroupSize_LiteralAndConst) {
-    Override("height", ty.i32(), Construct(ty.i32(), 2_i), Id(7u));
+    Override("height", ty.i32(), Construct(ty.i32(), 2_i), Id(7_u));
     GlobalConst("depth", ty.i32(), Construct(ty.i32(), 3_i));
     auto* func = Func("main", utils::Empty, ty.void_(), utils::Empty,
                       utils::Vector{

src/tint/writer/spirv/builder_global_variable_test.cc

@@ -250,7 +250,7 @@ OpDecorate %1 DescriptorSet 3
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Bool) {
-    auto* v = Override("var", ty.bool_(), Expr(true), Id(1200));
+    auto* v = Override("var", ty.bool_(), Expr(true), Id(1200_a));
 
     spirv::Builder& b = Build();
 
@@ -265,7 +265,7 @@ TEST_F(BuilderTest, GlobalVar_Override_Bool) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Bool_ZeroValue) {
-    auto* v = Override("var", ty.bool_(), Construct<bool>(), Id(1200));
+    auto* v = Override("var", ty.bool_(), Construct<bool>(), Id(1200_a));
 
     spirv::Builder& b = Build();
 
@@ -280,7 +280,7 @@ TEST_F(BuilderTest, GlobalVar_Override_Bool_ZeroValue) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Bool_NoConstructor) {
-    auto* v = Override("var", ty.bool_(), Id(1200));
+    auto* v = Override("var", ty.bool_(), Id(1200_a));
 
     spirv::Builder& b = Build();
 
@@ -295,7 +295,7 @@ TEST_F(BuilderTest, GlobalVar_Override_Bool_NoConstructor) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar) {
-    auto* v = Override("var", ty.f32(), Expr(2_f), Id(0));
+    auto* v = Override("var", ty.f32(), Expr(2_f), Id(0_a));
 
     spirv::Builder& b = Build();
 
@@ -310,7 +310,7 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar_ZeroValue) {
-    auto* v = Override("var", ty.f32(), Construct<f32>(), Id(0));
+    auto* v = Override("var", ty.f32(), Construct<f32>(), Id(0_a));
 
     spirv::Builder& b = Build();
 
@@ -325,7 +325,7 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar_ZeroValue) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar_F32_NoConstructor) {
-    auto* v = Override("var", ty.f32(), Id(0));
+    auto* v = Override("var", ty.f32(), Id(0_a));
 
     spirv::Builder& b = Build();
 
@@ -340,7 +340,7 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar_F32_NoConstructor) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar_I32_NoConstructor) {
-    auto* v = Override("var", ty.i32(), Id(0));
+    auto* v = Override("var", ty.i32(), Id(0_a));
 
     spirv::Builder& b = Build();
 
@@ -355,7 +355,7 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar_I32_NoConstructor) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar_U32_NoConstructor) {
-    auto* v = Override("var", ty.u32(), Id(0));
+    auto* v = Override("var", ty.u32(), Id(0_a));
 
     spirv::Builder& b = Build();
 
@@ -370,7 +370,7 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar_U32_NoConstructor) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_NoId) {
-    auto* var_a = Override("a", ty.bool_(), Expr(true), Id(0));
+    auto* var_a = Override("a", ty.bool_(), Expr(true), Id(0_a));
     auto* var_b = Override("b", ty.bool_(), Expr(false));
 
     spirv::Builder& b = Build();

src/tint/writer/wgsl/generator_impl.cc

@@ -776,7 +776,11 @@ bool GeneratorImpl::EmitAttributes(std::ostream& out,
                 return true;
             },
             [&](const ast::IdAttribute* override_deco) {
-                out << "id(" << override_deco->value << ")";
+                out << "id(";
+                if (!EmitExpression(out, override_deco->value)) {
+                    return false;
+                }
+                out << ")";
                 return true;
             },
             [&](const ast::StructMemberSizeAttribute* size) {

src/tint/writer/wgsl/generator_impl_global_decl_test.cc

@@ -144,7 +144,7 @@ TEST_F(WgslGeneratorImplTest, Emit_GlobalConst) {
 
 TEST_F(WgslGeneratorImplTest, Emit_OverridableConstants) {
     Override("a", ty.f32());
-    Override("b", ty.f32(), Id(7u));
+    Override("b", ty.f32(), Id(7_a));
 
     GeneratorImpl& gen = Build();