Add support for override keyword and @id attribute

This replaces the @override keyword that was previously used for
pipeline-overridable constants.

Support for pipeline-overridable constants in Dawn is hidden behind
the "disallow-unsafe-APIs" toggle, so we can make this change directly
instead of deprecating and continuing to support the old syntax.

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

docs/origin-trial-changes.md

@@ -11,6 +11,7 @@
 ### New Features
 
 * Module-scope declarations can now be declared in any order. [tint:1266](crbug.com/tint/1266)
+* The `override` keyword and `@id()` attribute for pipeline-overridable constants are now supported, replacing the `@override` attribute. [tint:1403](crbug.com/tint/1403)
 
 ## Changes for M99
 

src/BUILD.gn

@@ -240,6 +240,8 @@ libtint_source_set("libtint_core_all_src") {
     "ast/group_attribute.h",
     "ast/i32.cc",
     "ast/i32.h",
+    "ast/id_attribute.cc",
+    "ast/id_attribute.h",
     "ast/identifier_expression.cc",
     "ast/identifier_expression.h",
     "ast/if_statement.cc",
@@ -270,8 +272,6 @@ libtint_source_set("libtint_core_all_src") {
     "ast/multisampled_texture.h",
     "ast/node.cc",
     "ast/node.h",
-    "ast/override_attribute.cc",
-    "ast/override_attribute.h",
     "ast/phony_expression.cc",
     "ast/phony_expression.h",
     "ast/pipeline_stage.cc",

src/CMakeLists.txt

@@ -123,6 +123,8 @@ set(TINT_LIB_SRCS
   ast/group_attribute.h
   ast/i32.cc
   ast/i32.h
+  ast/id_attribute.cc
+  ast/id_attribute.h
   ast/identifier_expression.cc
   ast/identifier_expression.h
   ast/if_statement.cc
@@ -151,8 +153,6 @@ set(TINT_LIB_SRCS
   ast/multisampled_texture.h
   ast/node.cc
   ast/node.h
-  ast/override_attribute.cc
-  ast/override_attribute.h
   ast/phony_expression.cc
   ast/phony_expression.h
   ast/pipeline_stage.cc
@@ -659,6 +659,7 @@ if(TINT_BUILD_TESTS)
     ast/function_test.cc
     ast/group_attribute_test.cc
     ast/i32_test.cc
+    ast/id_attribute_test.cc
     ast/identifier_expression_test.cc
     ast/if_statement_test.cc
     ast/index_accessor_expression_test.cc
@@ -672,7 +673,6 @@ if(TINT_BUILD_TESTS)
     ast/module_clone_test.cc
     ast/module_test.cc
     ast/multisampled_texture_test.cc
-    ast/override_attribute_test.cc
     ast/phony_expression_test.cc
     ast/pointer_test.cc
     ast/return_statement_test.cc

src/ast/group_attribute_test.cc

@@ -12,7 +12,6 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
 #include "src/ast/test_helper.h"
 
 namespace tint {

src/ast/id_attribute.cc

@@ -0,0 +1,42 @@
+// 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/ast/id_attribute.h"
+
+#include <string>
+
+#include "src/program_builder.h"
+
+TINT_INSTANTIATE_TYPEINFO(tint::ast::IdAttribute);
+
+namespace tint {
+namespace ast {
+
+IdAttribute::IdAttribute(ProgramID pid, const Source& src, uint32_t val)
+    : Base(pid, src), value(val) {}
+
+IdAttribute::~IdAttribute() = default;
+
+std::string IdAttribute::Name() const {
+  return "id";
+}
+
+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);
+}
+
+}  // namespace ast
+}  // namespace tint

src/ast/id_attribute.h

@@ -1,4 +1,4 @@
-// Copyright 2020 The Tint Authors.
+// 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.
@@ -12,8 +12,8 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#ifndef SRC_AST_OVERRIDE_ATTRIBUTE_H_
+#ifndef SRC_AST_ID_ATTRIBUTE_H_
-#define SRC_AST_OVERRIDE_ATTRIBUTE_H_
+#define SRC_AST_ID_ATTRIBUTE_H_
 
 #include <string>
 
@@ -22,19 +22,15 @@
 namespace tint {
 namespace ast {
 
-/// An override attribute
+/// An id attribute for pipeline-overridable constants
-class OverrideAttribute : public Castable<OverrideAttribute, Attribute> {
+class IdAttribute : public Castable<IdAttribute, Attribute> {
  public:
-  /// Create an override attribute with no specified id.
+  /// Create an id attribute.
   /// @param pid the identifier of the program that owns this node
   /// @param src the source of this node
-  OverrideAttribute(ProgramID pid, const Source& src);
+  /// @param val the numeric id value
-  /// Create an override attribute with a specific id value.
+  IdAttribute(ProgramID pid, const Source& src, uint32_t val);
-  /// @param pid the identifier of the program that owns this node
+  ~IdAttribute() override;
-  /// @param src the source of this node
-  /// @param val the override value
-  OverrideAttribute(ProgramID pid, const Source& src, uint32_t val);
-  ~OverrideAttribute() override;
 
   /// @returns the WGSL name for the attribute
   std::string Name() const override;
@@ -43,16 +39,13 @@ class OverrideAttribute : public Castable<OverrideAttribute, Attribute> {
   /// `ctx`.
   /// @param ctx the clone context
   /// @return the newly cloned node
-  const OverrideAttribute* Clone(CloneContext* ctx) const override;
+  const IdAttribute* Clone(CloneContext* ctx) const override;
 
-  /// True if an override id was specified
+  /// The id value
-  const bool has_value;
-
-  /// The override id value
   const uint32_t value;
 };
 
 }  // namespace ast
 }  // namespace tint
 
-#endif  // SRC_AST_OVERRIDE_ATTRIBUTE_H_
+#endif  // SRC_AST_ID_ATTRIBUTE_H_

src/ast/id_attribute_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 
 #include "src/ast/test_helper.h"
 
@@ -20,19 +20,13 @@ namespace tint {
 namespace ast {
 namespace {
 
-using OverrideAttributeTest = TestHelper;
+using IdAttributeTest = TestHelper;
 
-TEST_F(OverrideAttributeTest, Creation_WithValue) {
+TEST_F(IdAttributeTest, Creation) {
-  auto* d = create<OverrideAttribute>(12);
+  auto* d = create<IdAttribute>(12);
-  EXPECT_TRUE(d->has_value);
   EXPECT_EQ(12u, d->value);
 }
 
-TEST_F(OverrideAttributeTest, Creation_WithoutValue) {
-  auto* d = create<OverrideAttribute>();
-  EXPECT_FALSE(d->has_value);
-}
-
 }  // namespace
 }  // namespace ast
 }  // namespace tint

src/ast/location_attribute_test.cc

@@ -12,7 +12,6 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
 #include "src/ast/test_helper.h"
 
 namespace tint {

src/ast/override_attribute.cc

@@ -1,51 +0,0 @@
-// Copyright 2020 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/ast/override_attribute.h"
-
-#include <string>
-
-#include "src/program_builder.h"
-
-TINT_INSTANTIATE_TYPEINFO(tint::ast::OverrideAttribute);
-
-namespace tint {
-namespace ast {
-
-OverrideAttribute::OverrideAttribute(ProgramID pid, const Source& src)
-    : Base(pid, src), has_value(false), value(0) {}
-
-OverrideAttribute::OverrideAttribute(ProgramID pid,
-                                     const Source& src,
-                                     uint32_t val)
-    : Base(pid, src), has_value(true), value(val) {}
-
-OverrideAttribute::~OverrideAttribute() = default;
-
-std::string OverrideAttribute::Name() const {
-  return "override";
-}
-
-const OverrideAttribute* OverrideAttribute::Clone(CloneContext* ctx) const {
-  // Clone arguments outside of create() call to have deterministic ordering
-  auto src = ctx->Clone(source);
-  if (has_value) {
-    return ctx->dst->create<OverrideAttribute>(src, value);
-  } else {
-    return ctx->dst->create<OverrideAttribute>(src);
-  }
-}
-
-}  // namespace ast
-}  // namespace tint

src/ast/variable.cc

@@ -14,7 +14,6 @@
 
 #include "src/ast/variable.h"
 
-#include "src/ast/override_attribute.h"
 #include "src/program_builder.h"
 #include "src/sem/variable.h"
 
@@ -30,17 +29,20 @@ Variable::Variable(ProgramID pid,
                    Access da,
                    const ast::Type* ty,
                    bool constant,
+                   bool overridable,
                    const Expression* ctor,
                    AttributeList attrs)
     : Base(pid, src),
       symbol(sym),
       type(ty),
       is_const(constant),
+      is_overridable(overridable),
       constructor(ctor),
       attributes(std::move(attrs)),
       declared_storage_class(dsc),
       declared_access(da) {
   TINT_ASSERT(AST, symbol.IsValid());
+  TINT_ASSERT(AST, is_overridable ? is_const : true);
   TINT_ASSERT_PROGRAM_IDS_EQUAL_IF_VALID(AST, symbol, program_id);
   TINT_ASSERT_PROGRAM_IDS_EQUAL_IF_VALID(AST, constructor, program_id);
 }
@@ -69,7 +71,8 @@ const Variable* Variable::Clone(CloneContext* ctx) const {
   auto* ctor = ctx->Clone(constructor);
   auto attrs = ctx->Clone(attributes);
   return ctx->dst->create<Variable>(src, sym, declared_storage_class,
-                                    declared_access, ty, is_const, ctor, attrs);
+                                    declared_access, ty, is_const,
+                                    is_overridable, ctor, attrs);
 }
 
 }  // namespace ast

src/ast/variable.h

@@ -46,8 +46,9 @@ struct VariableBindingPoint {
 
 /// A Variable statement.
 ///
-/// An instance of this class represents one of three constructs in WGSL: "var"
+/// An instance of this class represents one of four constructs in WGSL: "var"
-/// declaration, "let" declaration, or formal parameter to a function.
+/// declaration, "let" declaration, "override" declaration, or formal parameter
+/// to a function.
 ///
 /// 1. A "var" declaration is a name for typed storage.  Examples:
 ///
@@ -65,7 +66,14 @@ struct VariableBindingPoint {
 ///
 ///       let twice_depth : i32 = width + width;  // Must have initializer
 ///
-/// 3. A formal parameter to a function is a name for a typed value to
+/// 3. An "override" declaration is a name for a pipeline-overridable constant.
+/// Examples:
+///
+///       override radius : i32 = 2;       // Can be overridden by name.
+///       @id(5) override width : i32 = 2; // Can be overridden by ID.
+///       override scale : f32;            // No default - must be overridden.
+///
+/// 4. A formal parameter to a function is a name for a typed value to
 ///    be passed into a function.  Example:
 ///
 ///       fn twice(a: i32) -> i32 {  // "a:i32" is the formal parameter
@@ -84,14 +92,22 @@ struct VariableBindingPoint {
 ///
 /// This class uses the term "type" to refer to:
 ///     the value type of a "let",
+///     the value type of an "override",
 ///     the value type of the formal parameter,
 ///     or the store type of the "var".
 //
 /// Setting is_const:
 ///   - "var" gets false
 ///   - "let" gets true
+///   - "override" gets true
 ///   - formal parameter gets true
 ///
+/// Setting is_overrideable:
+///   - "var" gets false
+///   - "let" gets false
+///   - "override" gets true
+///   - formal parameter gets false
+///
 /// Setting storage class:
 ///   - "var" is StorageClass::kNone when using the
 ///     defaulting syntax for a "var" declared inside a function.
@@ -107,6 +123,7 @@ class Variable : public Castable<Variable, Node> {
   /// @param declared_access the declared access control
   /// @param type the declared variable type
   /// @param is_const true if the variable is const
+  /// @param is_overridable true if the variable is pipeline-overridable
   /// @param constructor the constructor expression
   /// @param attributes the variable attributes
   Variable(ProgramID program_id,
@@ -116,6 +133,7 @@ class Variable : public Castable<Variable, Node> {
            Access declared_access,
            const ast::Type* type,
            bool is_const,
+           bool is_overridable,
            const Expression* constructor,
            AttributeList attributes);
   /// Move constructor
@@ -143,6 +161,9 @@ class Variable : public Castable<Variable, Node> {
   /// True if this is a constant, false otherwise
   const bool is_const;
 
+  /// True if this is a pipeline-overridable constant, false otherwise
+  const bool is_overridable;
+
   /// The constructor expression or nullptr if none set
   const Expression* const constructor;
 

src/ast/variable_test.cc

@@ -13,7 +13,7 @@
 // limitations under the License.
 
 #include "gtest/gtest-spi.h"
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/test_helper.h"
 
 namespace tint {
@@ -96,13 +96,13 @@ TEST_F(VariableTest, WithAttributes) {
                   AttributeList{
                       create<LocationAttribute>(1),
                       create<BuiltinAttribute>(Builtin::kPosition),
-                      create<OverrideAttribute>(1200),
+                      create<IdAttribute>(1200),
                   });
 
   auto& attributes = var->attributes;
   EXPECT_TRUE(ast::HasAttribute<ast::LocationAttribute>(attributes));
   EXPECT_TRUE(ast::HasAttribute<ast::BuiltinAttribute>(attributes));
-  EXPECT_TRUE(ast::HasAttribute<ast::OverrideAttribute>(attributes));
+  EXPECT_TRUE(ast::HasAttribute<ast::IdAttribute>(attributes));
 
   auto* location = ast::GetAttribute<ast::LocationAttribute>(attributes);
   ASSERT_NE(nullptr, location);

src/inspector/inspector.cc

@@ -20,10 +20,10 @@
 #include "src/ast/bool_literal_expression.h"
 #include "src/ast/call_expression.h"
 #include "src/ast/float_literal_expression.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/interpolate_attribute.h"
 #include "src/ast/location_attribute.h"
 #include "src/ast/module.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/sint_literal_expression.h"
 #include "src/ast/uint_literal_expression.h"
 #include "src/sem/array.h"
@@ -213,10 +213,9 @@ std::vector<EntryPoint> Inspector::GetEntryPoints() {
 
         overridable_constant.is_initialized =
             global->Declaration()->constructor;
-        auto* override_attr = ast::GetAttribute<ast::OverrideAttribute>(
-            global->Declaration()->attributes);
         overridable_constant.is_numeric_id_specified =
-            override_attr ? override_attr->has_value : false;
+            ast::HasAttribute<ast::IdAttribute>(
+                global->Declaration()->attributes);
 
         entry_point.overridable_constants.push_back(overridable_constant);
       }

src/inspector/inspector_test.cc

@@ -15,7 +15,6 @@
 #include "gtest/gtest.h"
 #include "src/ast/call_statement.h"
 #include "src/ast/disable_validation_attribute.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/stage_attribute.h"
 #include "src/ast/struct_block_attribute.h"
 #include "src/ast/workgroup_attribute.h"

src/inspector/test_inspector_builder.h

@@ -22,7 +22,7 @@
 
 #include "src/ast/call_statement.h"
 #include "src/ast/disable_validation_attribute.h"
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/stage_attribute.h"
 #include "src/ast/struct_block_attribute.h"
 #include "src/ast/workgroup_attribute.h"
@@ -109,10 +109,7 @@ class InspectorBuilder : public ProgramBuilder {
       uint32_t id,
       const ast::Type* type,
       const ast::Expression* constructor) {
-    return GlobalConst(name, type, constructor,
+    return Override(name, type, constructor, {Id(id)});
-                       ast::AttributeList{
-                           Override(id),
-                       });
   }
 
   /// Add a pipeline constant to the global variables, without a specific ID.
@@ -125,10 +122,7 @@ class InspectorBuilder : public ProgramBuilder {
       std::string name,
       const ast::Type* type,
       const ast::Expression* constructor) {
-    return GlobalConst(name, type, constructor,
+    return Override(name, type, constructor);
-                       ast::AttributeList{
-                           Override(),
-                       });
   }
 
   /// Generates a function that references module-scoped, plain-typed constant

src/program_builder.h

@@ -43,6 +43,7 @@
 #include "src/ast/float_literal_expression.h"
 #include "src/ast/for_loop_statement.h"
 #include "src/ast/i32.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/if_statement.h"
 #include "src/ast/index_accessor_expression.h"
 #include "src/ast/interpolate_attribute.h"
@@ -52,7 +53,6 @@
 #include "src/ast/member_accessor_expression.h"
 #include "src/ast/module.h"
 #include "src/ast/multisampled_texture.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/phony_expression.h"
 #include "src/ast/pointer.h"
 #include "src/ast/return_statement.h"
@@ -1333,7 +1333,8 @@ class ProgramBuilder {
                            OPTIONAL&&... optional) {
     VarOptionals opts(std::forward<OPTIONAL>(optional)...);
     return create<ast::Variable>(Sym(std::forward<NAME>(name)), opts.storage,
-                                 opts.access, type, false, opts.constructor,
+                                 opts.access, type, false /* is_const */,
+                                 false /* is_overridable */, opts.constructor,
                                  std::move(opts.attributes));
   }
 
@@ -1355,9 +1356,10 @@ class ProgramBuilder {
                            const ast::Type* type,
                            OPTIONAL&&... optional) {
     VarOptionals opts(std::forward<OPTIONAL>(optional)...);
-    return create<ast::Variable>(source, Sym(std::forward<NAME>(name)),
+    return create<ast::Variable>(
-                                 opts.storage, opts.access, type, false,
+        source, Sym(std::forward<NAME>(name)), opts.storage, opts.access, type,
-                                 opts.constructor, std::move(opts.attributes));
+        false /* is_const */, false /* is_overridable */, opts.constructor,
+        std::move(opts.attributes));
   }
 
   /// @param name the variable name
@@ -1372,7 +1374,8 @@ class ProgramBuilder {
                              ast::AttributeList attributes = {}) {
     return create<ast::Variable>(
         Sym(std::forward<NAME>(name)), ast::StorageClass::kNone,
-        ast::Access::kUndefined, type, true, constructor, attributes);
+        ast::Access::kUndefined, type, true /* is_const */,
+        false /* is_overridable */, constructor, attributes);
   }
 
   /// @param source the variable source
@@ -1389,7 +1392,8 @@ class ProgramBuilder {
                              ast::AttributeList attributes = {}) {
     return create<ast::Variable>(
         source, Sym(std::forward<NAME>(name)), ast::StorageClass::kNone,
-        ast::Access::kUndefined, type, true, constructor, attributes);
+        ast::Access::kUndefined, type, true /* is_const */,
+        false /* is_overridable */, constructor, attributes);
   }
 
   /// @param name the parameter name
@@ -1402,7 +1406,8 @@ class ProgramBuilder {
                              ast::AttributeList attributes = {}) {
     return create<ast::Variable>(
         Sym(std::forward<NAME>(name)), ast::StorageClass::kNone,
-        ast::Access::kUndefined, type, true, nullptr, attributes);
+        ast::Access::kUndefined, type, true /* is_const */,
+        false /* is_overridable */, nullptr, attributes);
   }
 
   /// @param source the parameter source
@@ -1417,7 +1422,8 @@ class ProgramBuilder {
                              ast::AttributeList attributes = {}) {
     return create<ast::Variable>(
         source, Sym(std::forward<NAME>(name)), ast::StorageClass::kNone,
-        ast::Access::kUndefined, type, true, nullptr, attributes);
+        ast::Access::kUndefined, type, true /* is_const */,
+        false /* is_overridable */, nullptr, attributes);
   }
 
   /// @param name the variable name
@@ -1506,6 +1512,47 @@ class ProgramBuilder {
     return var;
   }
 
+  /// @param name the variable name
+  /// @param type the variable type
+  /// @param constructor optional constructor expression
+  /// @param attributes optional variable attributes
+  /// @returns an overridable const `ast::Variable` which is automatically
+  /// registered as a global variable with the ast::Module.
+  template <typename NAME>
+  const ast::Variable* Override(NAME&& name,
+                                const ast::Type* type,
+                                const ast::Expression* constructor,
+                                ast::AttributeList attributes = {}) {
+    auto* var = create<ast::Variable>(
+        source_, Sym(std::forward<NAME>(name)), ast::StorageClass::kNone,
+        ast::Access::kUndefined, type, true /* is_const */,
+        true /* is_overridable */, constructor, std::move(attributes));
+    AST().AddGlobalVariable(var);
+    return var;
+  }
+
+  /// @param source the variable source
+  /// @param name the variable name
+  /// @param type the variable type
+  /// @param constructor constructor expression
+  /// @param attributes optional variable attributes
+  /// @returns a const `ast::Variable` constructed by calling Var() with the
+  /// arguments of `args`, which is automatically registered as a global
+  /// variable with the ast::Module.
+  template <typename NAME>
+  const ast::Variable* Override(const Source& source,
+                                NAME&& name,
+                                const ast::Type* type,
+                                const ast::Expression* constructor,
+                                ast::AttributeList attributes = {}) {
+    auto* var = create<ast::Variable>(
+        source, Sym(std::forward<NAME>(name)), ast::StorageClass::kNone,
+        ast::Access::kUndefined, type, true /* is_const */,
+        true /* is_overridable */, constructor, std::move(attributes));
+    AST().AddGlobalVariable(var);
+    return var;
+  }
+
   /// @param source the source information
   /// @param expr the expression to take the address of
   /// @return an ast::UnaryOpExpression that takes the address of `expr`
@@ -2382,31 +2429,18 @@ class ProgramBuilder {
     return create<ast::LocationAttribute>(source_, location);
   }
 
-  /// Creates an ast::OverrideAttribute with a specific constant ID
+  /// Creates an ast::IdAttribute
   /// @param source the source information
   /// @param id the id value
   /// @returns the override attribute pointer
-  const ast::OverrideAttribute* Override(const Source& source, uint32_t id) {
+  const ast::IdAttribute* Id(const Source& source, uint32_t id) {
-    return create<ast::OverrideAttribute>(source, id);
+    return create<ast::IdAttribute>(source, id);
   }
 
-  /// Creates an ast::OverrideAttribute with a specific constant ID
+  /// Creates an ast::IdAttribute with a constant ID
   /// @param id the optional id value
   /// @returns the override attribute pointer
-  const ast::OverrideAttribute* Override(uint32_t id) {
+  const ast::IdAttribute* Id(uint32_t id) { return Id(source_, id); }
-    return Override(source_, id);
-  }
-
-  /// Creates an ast::OverrideAttribute without a constant ID
-  /// @param source the source information
-  /// @returns the override attribute pointer
-  const ast::OverrideAttribute* Override(const Source& source) {
-    return create<ast::OverrideAttribute>(source);
-  }
-
-  /// Creates an ast::OverrideAttribute without a constant ID
-  /// @returns the override attribute pointer
-  const ast::OverrideAttribute* Override() { return Override(source_); }
 
   /// Creates an ast::StageAttribute
   /// @param source the source information

src/reader/spirv/function.cc

@@ -1440,8 +1440,8 @@ bool FunctionEmitter::ParseFunctionDeclaration(FunctionDeclaration* decl) {
         auto* type = parser_impl_.ConvertType(param->type_id());
         if (type != nullptr) {
           auto* ast_param = parser_impl_.MakeVariable(
-              param->result_id(), ast::StorageClass::kNone, type, true, nullptr,
+              param->result_id(), ast::StorageClass::kNone, type, true, false,
-              ast::AttributeList{});
+              nullptr, ast::AttributeList{});
           // Parameters are treated as const declarations.
           ast_params.emplace_back(ast_param);
           // The value is accessible by name.
@@ -2542,7 +2542,7 @@ bool FunctionEmitter::EmitFunctionVariables() {
     }
     auto* var = parser_impl_.MakeVariable(
         inst.result_id(), ast::StorageClass::kNone, var_store_type, false,
-        constructor, ast::AttributeList{});
+        false, constructor, ast::AttributeList{});
     auto* var_decl_stmt = create<ast::VariableDeclStatement>(Source{}, var);
     AddStatement(var_decl_stmt);
     auto* var_type = ty_.Reference(var_store_type, ast::StorageClass::kNone);
@@ -3444,9 +3444,9 @@ bool FunctionEmitter::EmitStatementsInBasicBlock(const BlockInfo& block_info,
     auto* storage_type =
         RemapStorageClass(parser_impl_.ConvertType(def_inst->type_id()), id);
     AddStatement(create<ast::VariableDeclStatement>(
-        Source{},
+        Source{}, parser_impl_.MakeVariable(id, ast::StorageClass::kNone,
-        parser_impl_.MakeVariable(id, ast::StorageClass::kNone, storage_type,
+                                            storage_type, false, false, nullptr,
-                                  false, nullptr, ast::AttributeList{})));
+                                            ast::AttributeList{})));
     auto* type = ty_.Reference(storage_type, ast::StorageClass::kNone);
     identifier_types_.emplace(id, type);
   }
@@ -3518,8 +3518,8 @@ bool FunctionEmitter::EmitConstDefinition(
 
   expr = AddressOfIfNeeded(expr, &inst);
   auto* ast_const = parser_impl_.MakeVariable(
-      inst.result_id(), ast::StorageClass::kNone, expr.type, true, expr.expr,
+      inst.result_id(), ast::StorageClass::kNone, expr.type, true, false,
-      ast::AttributeList{});
+      expr.expr, ast::AttributeList{});
   if (!ast_const) {
     return false;
   }

src/reader/spirv/parser_impl.cc

@@ -22,8 +22,8 @@
 #include "source/opt/build_module.h"
 #include "src/ast/bitcast_expression.h"
 #include "src/ast/disable_validation_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/interpolate_attribute.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/type_name.h"
 #include "src/ast/unary_op_expression.h"
 #include "src/reader/spirv/function.h"
@@ -1395,14 +1395,14 @@ bool ParserImpl::EmitScalarSpecConstants() {
                              "between 0 and 65535: ID %"
                           << inst.result_id() << " has SpecId " << id;
           }
-          auto* cid = create<ast::OverrideAttribute>(Source{}, id);
+          auto* cid = create<ast::IdAttribute>(Source{}, id);
           spec_id_decos.push_back(cid);
           break;
         }
       }
       auto* ast_var =
           MakeVariable(inst.result_id(), ast::StorageClass::kNone, ast_type,
-                       true, ast_expr, std::move(spec_id_decos));
+                       true, true, ast_expr, std::move(spec_id_decos));
       if (ast_var) {
         builder_.AST().AddGlobalVariable(ast_var);
         scalar_spec_constants_.insert(inst.result_id());
@@ -1526,7 +1526,7 @@ bool ParserImpl::EmitModuleScopeVariables() {
     }
     auto* ast_var =
         MakeVariable(var.result_id(), ast_storage_class, ast_store_type, false,
-                     ast_constructor, ast::AttributeList{});
+                     false, ast_constructor, ast::AttributeList{});
     // TODO(dneto): initializers (a.k.a. constructor expression)
     if (ast_var) {
       builder_.AST().AddGlobalVariable(ast_var);
@@ -1561,7 +1561,7 @@ bool ParserImpl::EmitModuleScopeVariables() {
     auto* ast_var = MakeVariable(
         builtin_position_.per_vertex_var_id,
         enum_converter_.ToStorageClass(builtin_position_.storage_class),
-        ConvertType(builtin_position_.position_member_type_id), false,
+        ConvertType(builtin_position_.position_member_type_id), false, false,
         ast_constructor, {});
 
     builder_.AST().AddGlobalVariable(ast_var);
@@ -1598,6 +1598,7 @@ ast::Variable* ParserImpl::MakeVariable(uint32_t id,
                                         ast::StorageClass sc,
                                         const Type* storage_type,
                                         bool is_const,
+                                        bool is_overridable,
                                         const ast::Expression* constructor,
                                         ast::AttributeList decorations) {
   if (storage_type == nullptr) {
@@ -1630,12 +1631,12 @@ ast::Variable* ParserImpl::MakeVariable(uint32_t id,
   std::string name = namer_.Name(id);
 
   // Note: we're constructing the variable here with the *storage* type,
-  // regardless of whether this is a `let` or `var` declaration.
+  // regardless of whether this is a `let`, `override`, or `var` declaration.
   // `var` declarations will have a resolved type of ref<storage>, but at the
-  // AST level both `var` and `let` are declared with the same type.
+  // AST level all three are declared with the same type.
   return create<ast::Variable>(Source{}, builder_.Symbols().Register(name), sc,
                                access, storage_type->Build(builder_), is_const,
-                               constructor, decorations);
+                               is_overridable, constructor, decorations);
 }
 
 bool ParserImpl::ConvertDecorationsForVariable(uint32_t id,

src/reader/spirv/parser_impl.h

@@ -421,6 +421,7 @@ class ParserImpl : Reader {
   /// @param sc the storage class, which cannot be ast::StorageClass::kNone
   /// @param storage_type the storage type of the variable
   /// @param is_const if true, the variable is const
+  /// @param is_overridable if true, the variable is pipeline-overridable
   /// @param constructor the variable constructor
   /// @param decorations the variable decorations
   /// @returns a new Variable node, or null in the ignorable variable case and
@@ -429,6 +430,7 @@ class ParserImpl : Reader {
                               ast::StorageClass sc,
                               const Type* storage_type,
                               bool is_const,
+                              bool is_overridable,
                               const ast::Expression* constructor,
                               ast::AttributeList decorations);
 

src/reader/spirv/parser_impl_module_var_test.cc

@@ -1655,7 +1655,7 @@ TEST_F(SpvModuleScopeVarParserTest, ScalarSpecConstant_DeclareConst_True) {
   ASSERT_TRUE(p->BuildAndParseInternalModuleExceptFunctions()) << p->error();
   EXPECT_TRUE(p->error().empty());
   const auto module_str = test::ToString(p->program());
-  EXPECT_THAT(module_str, HasSubstr("@override(12) let myconst : bool = true;"))
+  EXPECT_THAT(module_str, HasSubstr("@id(12) override myconst : bool = true;"))
       << module_str;
 }
 
@@ -1671,8 +1671,7 @@ TEST_F(SpvModuleScopeVarParserTest, ScalarSpecConstant_DeclareConst_False) {
   ASSERT_TRUE(p->BuildAndParseInternalModuleExceptFunctions()) << p->error();
   EXPECT_TRUE(p->error().empty());
   const auto module_str = test::ToString(p->program());
-  EXPECT_THAT(module_str,
+  EXPECT_THAT(module_str, HasSubstr("@id(12) override myconst : bool = false;"))
-              HasSubstr("@override(12) let myconst : bool = false;"))
       << module_str;
 }
 
@@ -1688,7 +1687,7 @@ TEST_F(SpvModuleScopeVarParserTest, ScalarSpecConstant_DeclareConst_U32) {
   ASSERT_TRUE(p->BuildAndParseInternalModuleExceptFunctions()) << p->error();
   EXPECT_TRUE(p->error().empty());
   const auto module_str = test::ToString(p->program());
-  EXPECT_THAT(module_str, HasSubstr("@override(12) let myconst : u32 = 42u;"))
+  EXPECT_THAT(module_str, HasSubstr("@id(12) override myconst : u32 = 42u;"))
       << module_str;
 }
 
@@ -1704,7 +1703,7 @@ TEST_F(SpvModuleScopeVarParserTest, ScalarSpecConstant_DeclareConst_I32) {
   ASSERT_TRUE(p->BuildAndParseInternalModuleExceptFunctions()) << p->error();
   EXPECT_TRUE(p->error().empty());
   const auto module_str = test::ToString(p->program());
-  EXPECT_THAT(module_str, HasSubstr("@override(12) let myconst : i32 = 42;"))
+  EXPECT_THAT(module_str, HasSubstr("@id(12) override myconst : i32 = 42;"))
       << module_str;
 }
 
@@ -1720,7 +1719,7 @@ TEST_F(SpvModuleScopeVarParserTest, ScalarSpecConstant_DeclareConst_F32) {
   ASSERT_TRUE(p->BuildAndParseInternalModuleExceptFunctions()) << p->error();
   EXPECT_TRUE(p->error().empty());
   const auto module_str = test::ToString(p->program());
-  EXPECT_THAT(module_str, HasSubstr("@override(12) let myconst : f32 = 2.5;"))
+  EXPECT_THAT(module_str, HasSubstr("@id(12) override myconst : f32 = 2.5;"))
       << module_str;
 }
 
@@ -1737,7 +1736,8 @@ TEST_F(SpvModuleScopeVarParserTest,
   ASSERT_TRUE(p->BuildAndParseInternalModuleExceptFunctions()) << p->error();
   EXPECT_TRUE(p->error().empty());
   const auto module_str = test::ToString(p->program());
-  EXPECT_THAT(module_str, HasSubstr("let myconst : f32 = 2.5;")) << module_str;
+  EXPECT_THAT(module_str, HasSubstr("override myconst : f32 = 2.5;"))
+      << module_str;
 }
 
 TEST_F(SpvModuleScopeVarParserTest, ScalarSpecConstant_UsedInFunction) {

src/reader/wgsl/lexer.cc

@@ -996,6 +996,8 @@ Token Lexer::check_keyword(const Source& source, std::string_view str) {
     return {Token::Type::kMat4x3, source, "mat4x3"};
   if (str == "mat4x4")
     return {Token::Type::kMat4x4, source, "mat4x4"};
+  if (str == "override")
+    return {Token::Type::kOverride, source, "override"};
   if (str == "private")
     return {Token::Type::kPrivate, source, "private"};
   if (str == "ptr")

src/reader/wgsl/lexer_test.cc

@@ -750,6 +750,7 @@ INSTANTIATE_TEST_SUITE_P(
         TokenData{"mat4x2", Token::Type::kMat4x2},
         TokenData{"mat4x3", Token::Type::kMat4x3},
         TokenData{"mat4x4", Token::Type::kMat4x4},
+        TokenData{"override", Token::Type::kOverride},
         TokenData{"private", Token::Type::kPrivate},
         TokenData{"ptr", Token::Type::kPtr},
         TokenData{"return", Token::Type::kReturn},

src/reader/wgsl/parser_impl.cc

@@ -23,10 +23,10 @@
 #include "src/ast/discard_statement.h"
 #include "src/ast/external_texture.h"
 #include "src/ast/fallthrough_statement.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/if_statement.h"
 #include "src/ast/invariant_attribute.h"
 #include "src/ast/loop_statement.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/return_statement.h"
 #include "src/ast/stage_attribute.h"
 #include "src/ast/struct_block_attribute.h"
@@ -114,10 +114,10 @@ const char kBindingAttribute[] = "binding";
 const char kBlockAttribute[] = "block";
 const char kBuiltinAttribute[] = "builtin";
 const char kGroupAttribute[] = "group";
+const char kIdAttribute[] = "id";
 const char kInterpolateAttribute[] = "interpolate";
 const char kInvariantAttribute[] = "invariant";
 const char kLocationAttribute[] = "location";
-const char kOverrideAttribute[] = "override";
 const char kSizeAttribute[] = "size";
 const char kAlignAttribute[] = "align";
 const char kStageAttribute[] = "stage";
@@ -127,8 +127,8 @@ const char kWorkgroupSizeAttribute[] = "workgroup_size";
 bool is_attribute(Token t) {
   return t == kAlignAttribute || t == kBindingAttribute ||
          t == kBlockAttribute || t == kBuiltinAttribute ||
-         t == kGroupAttribute || t == kInterpolateAttribute ||
+         t == kGroupAttribute || t == kIdAttribute ||
-         t == kLocationAttribute || t == kOverrideAttribute ||
+         t == kInterpolateAttribute || t == kLocationAttribute ||
          t == kSizeAttribute || t == kStageAttribute || t == kStrideAttribute ||
          t == kWorkgroupSizeAttribute;
 }
@@ -485,22 +485,29 @@ Maybe<const ast::Variable*> ParserImpl::global_variable_decl(
       decl->access,                             // access control
       decl->type,                               // type
       false,                                    // is_const
+      false,                                    // is_overridable
       constructor,                              // constructor
       std::move(attrs));                        // attributes
 }
 
-// global_constant_decl
+// global_constant_decl :
-//  : attribute_list* LET variable_ident_decl global_const_initializer?
+//  | LET (ident | variable_ident_decl) global_const_initializer
+//  | attribute* override (ident | variable_ident_decl) (equal expression)?
 // global_const_initializer
 //  : EQUAL const_expr
 Maybe<const ast::Variable*> ParserImpl::global_constant_decl(
     ast::AttributeList& attrs) {
-  if (!match(Token::Type::kLet)) {
+  bool is_overridable = false;
+  const char* use = nullptr;
+  if (match(Token::Type::kLet)) {
+    use = "let declaration";
+  } else if (match(Token::Type::kOverride)) {
+    use = "override declaration";
+    is_overridable = true;
+  } else {
     return Failure::kNoMatch;
   }
 
-  const char* use = "let declaration";
-
   auto decl = expect_variable_ident_decl(use, /* allow_inferred = */ true);
   if (decl.errored)
     return Failure::kErrored;
@@ -521,6 +528,7 @@ Maybe<const ast::Variable*> ParserImpl::global_constant_decl(
       ast::Access::kUndefined,                  // access control
       decl->type,                               // type
       true,                                     // is_const
+      is_overridable,                           // is_overridable
       initializer,                              // constructor
       std::move(attrs));                        // attributes
 }
@@ -1402,6 +1410,7 @@ Expect<ast::Variable*> ParserImpl::expect_param() {
                             ast::Access::kUndefined,   // access control
                             decl->type,                // type
                             true,                      // is_const
+                            false,                     // is_overridable
                             nullptr,                   // constructor
                             std::move(attrs.value));   // attributes
   // Formal parameters are treated like a const declaration where the
@@ -1660,6 +1669,7 @@ Maybe<const ast::VariableDeclStatement*> ParserImpl::variable_stmt() {
         ast::Access::kUndefined,                  // access control
         decl->type,                               // type
         true,                                     // is_const
+        false,                                    // is_overridable
         constructor.value,                        // constructor
         ast::AttributeList{});                    // attributes
 
@@ -1690,6 +1700,7 @@ Maybe<const ast::VariableDeclStatement*> ParserImpl::variable_stmt() {
                             decl->access,           // access control
                             decl->type,             // type
                             false,                  // is_const
+                            false,                  // is_overridable
                             constructor,            // constructor
                             ast::AttributeList{});  // attributes
 
@@ -3081,22 +3092,15 @@ Maybe<const ast::Attribute*> ParserImpl::attribute() {
     });
   }
 
-  if (t == kOverrideAttribute) {
+  if (t == kIdAttribute) {
-    const char* use = "override attribute";
+    const char* use = "id attribute";
+    return expect_paren_block(use, [&]() -> Result {
+      auto val = expect_positive_sint(use);
+      if (val.errored)
+        return Failure::kErrored;
 
-    if (peek_is(Token::Type::kParenLeft)) {
+      return create<ast::IdAttribute>(t.source(), val.value);
-      // @override(x)
+    });
-      return expect_paren_block(use, [&]() -> Result {
-        auto val = expect_positive_sint(use);
-        if (val.errored)
-          return Failure::kErrored;
-
-        return create<ast::OverrideAttribute>(t.source(), val.value);
-      });
-    } else {
-      // [[override]]
-      return create<ast::OverrideAttribute>(t.source());
-    }
   }
 
   return Failure::kNoMatch;

src/reader/wgsl/parser_impl_error_msg_test.cc

@@ -54,24 +54,24 @@ fn f() { return 1 & >; }
 }
 
 TEST_F(ParserImplErrorTest, AliasDeclInvalidAttribute) {
-  EXPECT("@override type e=u32;",
+  EXPECT("@invariant type e=u32;",
          R"(test.wgsl:1:2 error: unexpected attributes
-@override type e=u32;
+@invariant type e=u32;
- ^^^^^^^^
+ ^^^^^^^^^
 )");
 }
 
 // TODO(crbug.com/tint/1382): Remove
 TEST_F(ParserImplErrorTest, DEPRECATED_AliasDeclInvalidAttribute) {
   EXPECT(
-      "[[override]]type e=u32;",
+      "[[invariant]]type e=u32;",
       R"(test.wgsl:1:1 warning: use of deprecated language feature: [[attribute]] style attributes have been replaced with @attribute style
-[[override]]type e=u32;
+[[invariant]]type e=u32;
 ^^
 
 test.wgsl:1:3 error: unexpected attributes
-[[override]]type e=u32;
+[[invariant]]type e=u32;
-  ^^^^^^^^
+  ^^^^^^^^^
 )");
 }
 

src/reader/wgsl/parser_impl_global_constant_decl_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/reader/wgsl/parser_impl_test_helper.h"
 
 namespace tint {
@@ -32,6 +32,7 @@ TEST_F(ParserImplTest, GlobalConstantDecl) {
   ASSERT_NE(e.value, nullptr);
 
   EXPECT_TRUE(e->is_const);
+  EXPECT_FALSE(e->is_overridable);
   EXPECT_EQ(e->symbol, p->builder().Symbols().Get("a"));
   ASSERT_NE(e->type, nullptr);
   EXPECT_TRUE(e->type->Is<ast::F32>());
@@ -43,8 +44,6 @@ TEST_F(ParserImplTest, GlobalConstantDecl) {
 
   ASSERT_NE(e->constructor, nullptr);
   EXPECT_TRUE(e->constructor->Is<ast::LiteralExpression>());
-
-  EXPECT_FALSE(ast::HasAttribute<ast::OverrideAttribute>(e.value->attributes));
 }
 
 TEST_F(ParserImplTest, GlobalConstantDecl_Inferred) {
@@ -59,6 +58,7 @@ TEST_F(ParserImplTest, GlobalConstantDecl_Inferred) {
   ASSERT_NE(e.value, nullptr);
 
   EXPECT_TRUE(e->is_const);
+  EXPECT_FALSE(e->is_overridable);
   EXPECT_EQ(e->symbol, p->builder().Symbols().Get("a"));
   EXPECT_EQ(e->type, nullptr);
 
@@ -69,8 +69,6 @@ TEST_F(ParserImplTest, GlobalConstantDecl_Inferred) {
 
   ASSERT_NE(e->constructor, nullptr);
   EXPECT_TRUE(e->constructor->Is<ast::LiteralExpression>());
-
-  EXPECT_FALSE(ast::HasAttribute<ast::OverrideAttribute>(e.value->attributes));
 }
 
 TEST_F(ParserImplTest, GlobalConstantDecl_InvalidExpression) {
@@ -100,7 +98,7 @@ TEST_F(ParserImplTest, GlobalConstantDecl_MissingExpression) {
 }
 
 TEST_F(ParserImplTest, GlobalConstantDec_Override_WithId) {
-  auto p = parser("@override(7) let a : f32 = 1.");
+  auto p = parser("@id(7) override a : f32 = 1.");
   auto attrs = p->attribute_list();
   EXPECT_FALSE(attrs.errored);
   EXPECT_TRUE(attrs.matched);
@@ -112,30 +110,30 @@ TEST_F(ParserImplTest, GlobalConstantDec_Override_WithId) {
   ASSERT_NE(e.value, nullptr);
 
   EXPECT_TRUE(e->is_const);
+  EXPECT_TRUE(e->is_overridable);
   EXPECT_EQ(e->symbol, p->builder().Symbols().Get("a"));
   ASSERT_NE(e->type, nullptr);
   EXPECT_TRUE(e->type->Is<ast::F32>());
 
   EXPECT_EQ(e->source.range.begin.line, 1u);
-  EXPECT_EQ(e->source.range.begin.column, 18u);
+  EXPECT_EQ(e->source.range.begin.column, 17u);
   EXPECT_EQ(e->source.range.end.line, 1u);
-  EXPECT_EQ(e->source.range.end.column, 19u);
+  EXPECT_EQ(e->source.range.end.column, 18u);
 
   ASSERT_NE(e->constructor, nullptr);
   EXPECT_TRUE(e->constructor->Is<ast::LiteralExpression>());
 
   auto* override_attr =
-      ast::GetAttribute<ast::OverrideAttribute>(e.value->attributes);
+      ast::GetAttribute<ast::IdAttribute>(e.value->attributes);
   ASSERT_NE(override_attr, nullptr);
-  EXPECT_TRUE(override_attr->has_value);
   EXPECT_EQ(override_attr->value, 7u);
 }
 
 TEST_F(ParserImplTest, GlobalConstantDec_Override_WithoutId) {
-  auto p = parser("[[override]] let a : f32 = 1.");
+  auto p = parser("override a : f32 = 1.");
   auto attrs = p->attribute_list();
   EXPECT_FALSE(attrs.errored);
-  EXPECT_TRUE(attrs.matched);
+  EXPECT_FALSE(attrs.matched);
 
   auto e = p->global_constant_decl(attrs.value);
   EXPECT_FALSE(p->has_error()) << p->error();
@@ -144,26 +142,25 @@ TEST_F(ParserImplTest, GlobalConstantDec_Override_WithoutId) {
   ASSERT_NE(e.value, nullptr);
 
   EXPECT_TRUE(e->is_const);
+  EXPECT_TRUE(e->is_overridable);
   EXPECT_EQ(e->symbol, p->builder().Symbols().Get("a"));
   ASSERT_NE(e->type, nullptr);
   EXPECT_TRUE(e->type->Is<ast::F32>());
 
   EXPECT_EQ(e->source.range.begin.line, 1u);
-  EXPECT_EQ(e->source.range.begin.column, 18u);
+  EXPECT_EQ(e->source.range.begin.column, 10u);
   EXPECT_EQ(e->source.range.end.line, 1u);
-  EXPECT_EQ(e->source.range.end.column, 19u);
+  EXPECT_EQ(e->source.range.end.column, 11u);
 
   ASSERT_NE(e->constructor, nullptr);
   EXPECT_TRUE(e->constructor->Is<ast::LiteralExpression>());
 
-  auto* override_attr =
+  auto* id_attr = ast::GetAttribute<ast::IdAttribute>(e.value->attributes);
-      ast::GetAttribute<ast::OverrideAttribute>(e.value->attributes);
+  ASSERT_EQ(id_attr, nullptr);
-  ASSERT_NE(override_attr, nullptr);
-  EXPECT_FALSE(override_attr->has_value);
 }
 
 TEST_F(ParserImplTest, GlobalConstantDec_Override_MissingId) {
-  auto p = parser("@override() let a : f32 = 1.");
+  auto p = parser("@id() override a : f32 = 1.");
   auto attrs = p->attribute_list();
   EXPECT_TRUE(attrs.errored);
   EXPECT_FALSE(attrs.matched);
@@ -175,11 +172,11 @@ TEST_F(ParserImplTest, GlobalConstantDec_Override_MissingId) {
 
   EXPECT_TRUE(p->has_error());
   EXPECT_EQ(p->error(),
-            "1:11: expected signed integer literal for override attribute");
+            "1:5: expected signed integer literal for id attribute");
 }
 
 TEST_F(ParserImplTest, GlobalConstantDec_Override_InvalidId) {
-  auto p = parser("@override(-7) let a : f32 = 1.");
+  auto p = parser("@id(-7) override a : f32 = 1.");
   auto attrs = p->attribute_list();
   EXPECT_TRUE(attrs.errored);
   EXPECT_FALSE(attrs.matched);
@@ -190,7 +187,7 @@ TEST_F(ParserImplTest, GlobalConstantDec_Override_InvalidId) {
   ASSERT_NE(e.value, nullptr);
 
   EXPECT_TRUE(p->has_error());
-  EXPECT_EQ(p->error(), "1:11: override attribute must be positive");
+  EXPECT_EQ(p->error(), "1:5: id attribute must be positive");
 }
 
 }  // namespace

src/reader/wgsl/parser_impl_struct_attribute_decl_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
+#include "src/ast/invariant_attribute.h"
 #include "src/reader/wgsl/parser_impl_test_helper.h"
 
 namespace tint {
@@ -21,14 +21,14 @@ namespace wgsl {
 namespace {
 
 TEST_F(ParserImplTest, AttributeDecl_Parses) {
-  auto p = parser("@override");
+  auto p = parser("@invariant");
   auto attrs = p->attribute_list();
   EXPECT_FALSE(p->has_error());
   EXPECT_FALSE(attrs.errored);
   EXPECT_TRUE(attrs.matched);
   ASSERT_EQ(attrs.value.size(), 1u);
-  auto* override_attr = attrs.value[0]->As<ast::Attribute>();
+  auto* invariant = attrs.value[0]->As<ast::Attribute>();
-  EXPECT_TRUE(override_attr->Is<ast::OverrideAttribute>());
+  EXPECT_TRUE(invariant->Is<ast::InvariantAttribute>());
 }
 
 TEST_F(ParserImplTest, AttributeDecl_MissingParenLeft) {
@@ -73,19 +73,19 @@ TEST_F(ParserImplTest, AttributeDecl_Invalidattribute) {
 
 // TODO(crbug.com/tint/1382): Remove
 TEST_F(ParserImplTest, DEPRECATED_attributeDecl_Parses) {
-  auto p = parser("[[override]]");
+  auto p = parser("[[invariant]]");
   auto attrs = p->attribute_list();
   EXPECT_FALSE(p->has_error());
   EXPECT_FALSE(attrs.errored);
   EXPECT_TRUE(attrs.matched);
   ASSERT_EQ(attrs.value.size(), 1u);
-  auto* override_attr = attrs.value[0]->As<ast::Attribute>();
+  auto* invariant_attr = attrs.value[0]->As<ast::Attribute>();
-  EXPECT_TRUE(override_attr->Is<ast::OverrideAttribute>());
+  EXPECT_TRUE(invariant_attr->Is<ast::InvariantAttribute>());
 }
 
 // TODO(crbug.com/tint/1382): Remove
 TEST_F(ParserImplTest, DEPRECATED_attributeDecl_MissingAttrRight) {
-  auto p = parser("[[override");
+  auto p = parser("[[invariant");
   auto attrs = p->attribute_list();
   EXPECT_TRUE(p->has_error());
   EXPECT_TRUE(attrs.errored);
@@ -94,7 +94,7 @@ TEST_F(ParserImplTest, DEPRECATED_attributeDecl_MissingAttrRight) {
   EXPECT_EQ(
       p->error(),
       R"(1:1: use of deprecated language feature: [[attribute]] style attributes have been replaced with @attribute style
-1:11: expected ']]' for attribute list)");
+1:12: expected ']]' for attribute list)");
 }
 
 // TODO(crbug.com/tint/1382): Remove

src/reader/wgsl/token.cc

@@ -177,6 +177,8 @@ std::string_view Token::TypeToName(Type type) {
       return "mat4x3";
     case Token::Type::kMat4x4:
       return "mat4x4";
+    case Token::Type::kOverride:
+      return "override";
     case Token::Type::kPrivate:
       return "private";
     case Token::Type::kPtr:

src/reader/wgsl/token.h

@@ -187,6 +187,8 @@ class Token {
     kMat4x3,
     /// A 'mat4x4'
     kMat4x4,
+    /// A 'override'
+    kOverride,
     /// A 'private'
     kPrivate,
     /// A 'ptr'

src/resolver/attribute_validation_test.cc

@@ -55,10 +55,10 @@ enum class AttributeKind {
   kBinding,
   kBuiltin,
   kGroup,
+  kId,
   kInterpolate,
   kInvariant,
   kLocation,
-  kOverride,
   kOffset,
   kSize,
   kStage,
@@ -98,6 +98,8 @@ static ast::AttributeList createAttributes(const Source& source,
       return {builder.Builtin(source, ast::Builtin::kPosition)};
     case AttributeKind::kGroup:
       return {builder.create<ast::GroupAttribute>(source, 1u)};
+    case AttributeKind::kId:
+      return {builder.create<ast::IdAttribute>(source, 0u)};
     case AttributeKind::kInterpolate:
       return {builder.Interpolate(source, ast::InterpolationType::kLinear,
                                   ast::InterpolationSampling::kCenter)};
@@ -105,8 +107,6 @@ static ast::AttributeList createAttributes(const Source& source,
       return {builder.Invariant(source)};
     case AttributeKind::kLocation:
       return {builder.Location(source, 1)};
-    case AttributeKind::kOverride:
-      return {builder.create<ast::OverrideAttribute>(source, 0u)};
     case AttributeKind::kOffset:
       return {builder.create<ast::StructMemberOffsetAttribute>(source, 4u)};
     case AttributeKind::kSize:
@@ -152,10 +152,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, false},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -187,10 +187,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, false},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -236,10 +236,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, false},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -275,10 +275,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, true},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     // kInterpolate tested separately (requires [[location]])
                     TestParams{AttributeKind::kInvariant, true},
                     TestParams{AttributeKind::kLocation, true},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -325,10 +325,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, true},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, true},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -373,10 +373,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, false},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -423,10 +423,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, true},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -469,10 +469,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, true},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     // kInterpolate tested separately (requires [[location]])
                     TestParams{AttributeKind::kInvariant, true},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -560,10 +560,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, false},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -617,10 +617,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, true},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     // kInterpolate tested separately (requires [[location]])
                     // kInvariant tested separately (requires position builtin)
                     TestParams{AttributeKind::kLocation, true},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, true},
                     TestParams{AttributeKind::kSize, true},
                     TestParams{AttributeKind::kStage, false},
@@ -701,10 +701,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, false},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -745,10 +745,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, false},
                     TestParams{AttributeKind::kInterpolate, false},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, false},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -810,10 +810,10 @@ INSTANTIATE_TEST_SUITE_P(
                     TestParams{AttributeKind::kBinding, false},
                     TestParams{AttributeKind::kBuiltin, false},
                     TestParams{AttributeKind::kGroup, false},
+                    TestParams{AttributeKind::kId, true},
                     TestParams{AttributeKind::kInterpolate, false},
                     TestParams{AttributeKind::kInvariant, false},
                     TestParams{AttributeKind::kLocation, false},
-                    TestParams{AttributeKind::kOverride, true},
                     TestParams{AttributeKind::kOffset, false},
                     TestParams{AttributeKind::kSize, false},
                     TestParams{AttributeKind::kStage, false},
@@ -825,15 +825,15 @@ INSTANTIATE_TEST_SUITE_P(
 TEST_F(ConstantAttributeTest, DuplicateAttribute) {
   GlobalConst("a", ty.f32(), Expr(1.23f),
               ast::AttributeList{
-                  create<ast::OverrideAttribute>(Source{{12, 34}}),
+                  create<ast::IdAttribute>(Source{{12, 34}}, 0),
-                  create<ast::OverrideAttribute>(Source{{56, 78}}, 1),
+                  create<ast::IdAttribute>(Source{{56, 78}}, 1),
               });
 
   WrapInFunction();
 
   EXPECT_FALSE(r()->Resolve());
   EXPECT_EQ(r()->error(),
-            R"(56:78 error: duplicate override attribute
+            R"(56:78 error: duplicate id attribute
 12:34 note: first attribute declared here)");
 }
 

src/resolver/dependency_graph.cc

@@ -404,9 +404,9 @@ class DependencyScanner {
     }
     if (attr->IsAnyOf<
             ast::BindingAttribute, ast::BuiltinAttribute, ast::GroupAttribute,
-            ast::InternalAttribute, ast::InterpolateAttribute,
+            ast::IdAttribute, ast::InternalAttribute, ast::InterpolateAttribute,
             ast::InvariantAttribute, ast::LocationAttribute,
-            ast::OverrideAttribute, ast::StageAttribute, ast::StrideAttribute,
+            ast::StageAttribute, ast::StrideAttribute,
             ast::StructBlockAttribute, ast::StructMemberAlignAttribute,
             ast::StructMemberOffsetAttribute,
             ast::StructMemberSizeAttribute>()) {

src/resolver/pipeline_overridable_constant_test.cc

@@ -49,7 +49,7 @@ TEST_F(ResolverPipelineOverridableConstantTest, NonOverridable) {
 }
 
 TEST_F(ResolverPipelineOverridableConstantTest, WithId) {
-  auto* a = GlobalConst("a", ty.f32(), Expr(1.f), {Override(7u)});
+  auto* a = Override("a", ty.f32(), Expr(1.f), {Id(7u)});
 
   EXPECT_TRUE(r()->Resolve()) << r()->error();
 
@@ -57,7 +57,7 @@ TEST_F(ResolverPipelineOverridableConstantTest, WithId) {
 }
 
 TEST_F(ResolverPipelineOverridableConstantTest, WithoutId) {
-  auto* a = GlobalConst("a", ty.f32(), Expr(1.f), {Override()});
+  auto* a = Override("a", ty.f32(), Expr(1.f));
 
   EXPECT_TRUE(r()->Resolve()) << r()->error();
 
@@ -66,12 +66,12 @@ TEST_F(ResolverPipelineOverridableConstantTest, WithoutId) {
 
 TEST_F(ResolverPipelineOverridableConstantTest, WithAndWithoutIds) {
   std::vector<ast::Variable*> variables;
-  auto* a = GlobalConst("a", ty.f32(), Expr(1.f), {Override()});
+  auto* a = Override("a", ty.f32(), Expr(1.f));
-  auto* b = GlobalConst("b", ty.f32(), Expr(1.f), {Override()});
+  auto* b = Override("b", ty.f32(), Expr(1.f));
-  auto* c = GlobalConst("c", ty.f32(), Expr(1.f), {Override(2u)});
+  auto* c = Override("c", ty.f32(), Expr(1.f), {Id(2u)});
-  auto* d = GlobalConst("d", ty.f32(), Expr(1.f), {Override(4u)});
+  auto* d = Override("d", ty.f32(), Expr(1.f), {Id(4u)});
-  auto* e = GlobalConst("e", ty.f32(), Expr(1.f), {Override()});
+  auto* e = Override("e", ty.f32(), Expr(1.f));
-  auto* f = GlobalConst("f", ty.f32(), Expr(1.f), {Override(1u)});
+  auto* f = Override("f", ty.f32(), Expr(1.f), {Id(1u)});
 
   EXPECT_TRUE(r()->Resolve()) << r()->error();
 
@@ -85,8 +85,8 @@ TEST_F(ResolverPipelineOverridableConstantTest, WithAndWithoutIds) {
 }
 
 TEST_F(ResolverPipelineOverridableConstantTest, DuplicateIds) {
-  GlobalConst("a", ty.f32(), Expr(1.f), {Override(Source{{12, 34}}, 7u)});
+  Override("a", ty.f32(), Expr(1.f), {Id(Source{{12, 34}}, 7u)});
-  GlobalConst("b", ty.f32(), Expr(1.f), {Override(Source{{56, 78}}, 7u)});
+  Override("b", ty.f32(), Expr(1.f), {Id(Source{{56, 78}}, 7u)});
 
   EXPECT_FALSE(r()->Resolve());
 
@@ -95,7 +95,7 @@ TEST_F(ResolverPipelineOverridableConstantTest, DuplicateIds) {
 }
 
 TEST_F(ResolverPipelineOverridableConstantTest, IdTooLarge) {
-  GlobalConst("a", ty.f32(), Expr(1.f), {Override(Source{{12, 34}}, 65536u)});
+  Override("a", ty.f32(), Expr(1.f), {Id(Source{{12, 34}}, 65536u)});
 
   EXPECT_FALSE(r()->Resolve());
 

src/resolver/resolver.cc

@@ -32,12 +32,12 @@
 #include "src/ast/discard_statement.h"
 #include "src/ast/fallthrough_statement.h"
 #include "src/ast/for_loop_statement.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/if_statement.h"
 #include "src/ast/internal_attribute.h"
 #include "src/ast/interpolate_attribute.h"
 #include "src/ast/loop_statement.h"
 #include "src/ast/matrix.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/pointer.h"
 #include "src/ast/return_statement.h"
 #include "src/ast/sampled_texture.h"
@@ -333,8 +333,8 @@ sem::Variable* Resolver::Variable(const ast::Variable* var,
 
       storage_ty = rhs->Type()->UnwrapRef();  // Implicit load of RHS
     }
-  } else if (var->is_const && kind != VariableKind::kParameter &&
+  } else if (var->is_const && !var->is_overridable &&
-             !ast::HasAttribute<ast::OverrideAttribute>(var->attributes)) {
+             kind != VariableKind::kParameter) {
     AddError("let declaration must have an initializer", var->source);
     return nullptr;
   } else if (!var->type) {
@@ -426,19 +426,16 @@ sem::Variable* Resolver::Variable(const ast::Variable* var,
         binding_point = {bp.group->value, bp.binding->value};
       }
 
-      auto* override =
+      bool has_const_val = rhs && var->is_const && !var->is_overridable;
-          ast::GetAttribute<ast::OverrideAttribute>(var->attributes);
-      bool has_const_val = rhs && var->is_const && !override;
-
       auto* global = builder_->create<sem::GlobalVariable>(
           var, var_ty, storage_class, access,
           has_const_val ? rhs->ConstantValue() : sem::Constant{},
           binding_point);
 
-      if (override) {
+      if (var->is_overridable) {
         global->SetIsOverridable();
-        if (override->has_value) {
+        if (auto* id = ast::GetAttribute<ast::IdAttribute>(var->attributes)) {
-          global->SetConstantId(static_cast<uint16_t>(override->value));
+          global->SetConstantId(static_cast<uint16_t>(id->value));
         }
       }
 
@@ -492,18 +489,13 @@ void Resolver::AllocateOverridableConstantIds() {
   // unused constant, the allocation may change on the next Resolver pass.
   for (auto* decl : builder_->AST().GlobalDeclarations()) {
     auto* var = decl->As<ast::Variable>();
-    if (!var) {
+    if (!var || !var->is_overridable) {
-      continue;
-    }
-    auto* override_attr =
-        ast::GetAttribute<ast::OverrideAttribute>(var->attributes);
-    if (!override_attr) {
       continue;
     }
 
     uint16_t constant_id;
-    if (override_attr->has_value) {
+    if (auto* id_attr = ast::GetAttribute<ast::IdAttribute>(var->attributes)) {
-      constant_id = static_cast<uint16_t>(override_attr->value);
+      constant_id = static_cast<uint16_t>(id_attr->value);
     } else {
       // No ID was specified, so allocate the next available ID.
       constant_id = next_constant_id;
@@ -551,11 +543,9 @@ sem::GlobalVariable* Resolver::GlobalVariable(const ast::Variable* var) {
   for (auto* attr : var->attributes) {
     Mark(attr);
 
-    if (auto* override_attr = attr->As<ast::OverrideAttribute>()) {
+    if (auto* id_attr = attr->As<ast::IdAttribute>()) {
       // Track the constant IDs that are specified in the shader.
-      if (override_attr->has_value) {
+      constant_ids_.emplace(id_attr->value, sem);
-        constant_ids_.emplace(override_attr->value, sem);
-      }
     }
   }
 
@@ -791,8 +781,8 @@ bool Resolver::WorkgroupSize(const ast::Function* func) {
         AddError(kErrBadType, expr->source);
         return false;
       }
-      // Capture the constant if an [[override]] attribute is present.
+      // Capture the constant if it is pipeline-overridable.
-      if (ast::HasAttribute<ast::OverrideAttribute>(decl->attributes)) {
+      if (decl->is_overridable) {
         ws[i].overridable_const = decl;
       }
 
@@ -2275,15 +2265,13 @@ sem::Array* Resolver::Array(const ast::Array* arr) {
 
     if (auto* ident = count_expr->As<ast::IdentifierExpression>()) {
       // Make sure the identifier is a non-overridable module-scope constant.
-      auto* var = ResolvedSymbol<sem::Variable>(ident);
+      auto* var = ResolvedSymbol<sem::GlobalVariable>(ident);
-      if (!var || !var->Is<sem::GlobalVariable>() ||
+      if (!var || !var->Declaration()->is_const) {
-          !var->Declaration()->is_const) {
         AddError("array size identifier must be a module-scope constant",
                  size_source);
         return nullptr;
       }
-      if (ast::HasAttribute<ast::OverrideAttribute>(
+      if (var->IsOverridable()) {
-              var->Declaration()->attributes)) {
         AddError("array size expression must not be pipeline-overridable",
                  size_source);
         return nullptr;

src/resolver/resolver_test.cc

@@ -25,9 +25,9 @@
 #include "src/ast/call_statement.h"
 #include "src/ast/continue_statement.h"
 #include "src/ast/float_literal_expression.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/if_statement.h"
 #include "src/ast/loop_statement.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/return_statement.h"
 #include "src/ast/stage_attribute.h"
 #include "src/ast/struct_block_attribute.h"
@@ -1011,14 +1011,14 @@ TEST_F(ResolverTest, Function_WorkgroupSize_Consts_NestedInitializer) {
 }
 
 TEST_F(ResolverTest, Function_WorkgroupSize_OverridableConsts) {
-  // @override(0) let width = 16;
+  // @id(0) override width = 16;
-  // @override(1) let height = 8;
+  // @id(1) override height = 8;
-  // @override(2) let depth = 2;
+  // @id(2) override depth = 2;
   // @stage(compute) @workgroup_size(width, height, depth)
   // fn main() {}
-  auto* width = GlobalConst("width", ty.i32(), Expr(16), {Override(0)});
+  auto* width = Override("width", ty.i32(), Expr(16), {Id(0)});
-  auto* height = GlobalConst("height", ty.i32(), Expr(8), {Override(1)});
+  auto* height = Override("height", ty.i32(), Expr(8), {Id(1)});
-  auto* depth = GlobalConst("depth", ty.i32(), Expr(2), {Override(2)});
+  auto* depth = Override("depth", ty.i32(), Expr(2), {Id(2)});
   auto* func = Func("main", ast::VariableList{}, ty.void_(), {},
                     {Stage(ast::PipelineStage::kCompute),
                      WorkgroupSize("width", "height", "depth")});
@@ -1037,14 +1037,14 @@ TEST_F(ResolverTest, Function_WorkgroupSize_OverridableConsts) {
 }
 
 TEST_F(ResolverTest, Function_WorkgroupSize_OverridableConsts_NoInit) {
-  // @override(0) let width : i32;
+  // @id(0) override width : i32;
-  // @override(1) let height : i32;
+  // @id(1) override height : i32;
-  // @override(2) let depth : i32;
+  // @id(2) override depth : i32;
   // @stage(compute) @workgroup_size(width, height, depth)
   // fn main() {}
-  auto* width = GlobalConst("width", ty.i32(), nullptr, {Override(0)});
+  auto* width = Override("width", ty.i32(), nullptr, {Id(0)});
-  auto* height = GlobalConst("height", ty.i32(), nullptr, {Override(1)});
+  auto* height = Override("height", ty.i32(), nullptr, {Id(1)});
-  auto* depth = GlobalConst("depth", ty.i32(), nullptr, {Override(2)});
+  auto* depth = Override("depth", ty.i32(), nullptr, {Id(2)});
   auto* func = Func("main", ast::VariableList{}, ty.void_(), {},
                     {Stage(ast::PipelineStage::kCompute),
                      WorkgroupSize("width", "height", "depth")});
@@ -1063,11 +1063,11 @@ TEST_F(ResolverTest, Function_WorkgroupSize_OverridableConsts_NoInit) {
 }
 
 TEST_F(ResolverTest, Function_WorkgroupSize_Mixed) {
-  // @override(1) let height = 2;
+  // @id(1) override height = 2;
   // let depth = 3;
   // @stage(compute) @workgroup_size(8, height, depth)
   // fn main() {}
-  auto* height = GlobalConst("height", ty.i32(), Expr(2), {Override(0)});
+  auto* height = Override("height", ty.i32(), Expr(2), {Id(0)});
   GlobalConst("depth", ty.i32(), Expr(3));
   auto* func = Func("main", ast::VariableList{}, ty.void_(), {},
                     {Stage(ast::PipelineStage::kCompute),

src/resolver/resolver_validation.cc

@@ -30,12 +30,12 @@
 #include "src/ast/discard_statement.h"
 #include "src/ast/fallthrough_statement.h"
 #include "src/ast/for_loop_statement.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/if_statement.h"
 #include "src/ast/internal_attribute.h"
 #include "src/ast/interpolate_attribute.h"
 #include "src/ast/loop_statement.h"
 #include "src/ast/matrix.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/pointer.h"
 #include "src/ast/return_statement.h"
 #include "src/ast/sampled_texture.h"
@@ -411,25 +411,23 @@ bool Resolver::ValidateGlobalVariable(const sem::Variable* var) {
 
   for (auto* attr : decl->attributes) {
     if (decl->is_const) {
-      if (auto* override_attr = attr->As<ast::OverrideAttribute>()) {
+      if (auto* id_attr = attr->As<ast::IdAttribute>()) {
-        if (override_attr->has_value) {
+        uint32_t id = id_attr->value;
-          uint32_t id = override_attr->value;
+        auto it = constant_ids_.find(id);
-          auto it = constant_ids_.find(id);
+        if (it != constant_ids_.end() && it->second != var) {
-          if (it != constant_ids_.end() && it->second != var) {
+          AddError("pipeline constant IDs must be unique", attr->source);
-            AddError("pipeline constant IDs must be unique", attr->source);
+          AddNote("a pipeline constant with an ID of " + std::to_string(id) +
-            AddNote("a pipeline constant with an ID of " + std::to_string(id) +
+                      " was previously declared "
-                        " was previously declared "
+                      "here:",
-                        "here:",
+                  ast::GetAttribute<ast::IdAttribute>(
-                    ast::GetAttribute<ast::OverrideAttribute>(
+                      it->second->Declaration()->attributes)
-                        it->second->Declaration()->attributes)
+                      ->source);
-                        ->source);
+          return false;
-            return false;
+        }
-          }
+        if (id > 65535) {
-          if (id > 65535) {
+          AddError("pipeline constant IDs must be between 0 and 65535",
-            AddError("pipeline constant IDs must be between 0 and 65535",
+                   attr->source);
-                     attr->source);
+          return false;
-            return false;
-          }
         }
       } else {
         AddError("attribute is not valid for constants", attr->source);

src/resolver/type_validation_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/return_statement.h"
 #include "src/ast/stage_attribute.h"
 #include "src/ast/struct_block_attribute.h"
@@ -79,9 +79,8 @@ TEST_F(ResolverTypeValidationTest, VariableDeclNoConstructor_Pass) {
 }
 
 TEST_F(ResolverTypeValidationTest, GlobalConstantNoConstructor_Pass) {
-  // @override(0) let a :i32;
+  // @id(0) override a :i32;
-  GlobalConst(Source{{12, 34}}, "a", ty.i32(), nullptr,
+  Override(Source{{12, 34}}, "a", ty.i32(), nullptr, ast::AttributeList{Id(0)});
-              ast::AttributeList{create<ast::OverrideAttribute>(0)});
 
   EXPECT_TRUE(r()->Resolve()) << r()->error();
 }
@@ -98,7 +97,7 @@ TEST_F(ResolverTypeValidationTest, GlobalConstantWithStorageClass_Fail) {
   AST().AddGlobalVariable(create<ast::Variable>(
       Source{{12, 34}}, Symbols().Register("global_var"),
       ast::StorageClass::kPrivate, ast::Access::kUndefined, ty.f32(), true,
-      Expr(1.23f), ast::AttributeList{}));
+      false, Expr(1.23f), ast::AttributeList{}));
 
   EXPECT_FALSE(r()->Resolve());
   EXPECT_EQ(r()->error(),
@@ -346,9 +345,9 @@ TEST_F(ResolverTypeValidationTest, ArraySize_TooBig_ExplicitStride) {
 }
 
 TEST_F(ResolverTypeValidationTest, ArraySize_OverridableConstant) {
-  // [[override]] let size = 10;
+  // override size = 10;
   // var<private> a : array<f32, size>;
-  GlobalConst("size", nullptr, Expr(10), {Override()});
+  Override("size", nullptr, Expr(10));
   Global("a", ty.array(ty.f32(), Expr(Source{{12, 34}}, "size")),
          ast::StorageClass::kPrivate);
   EXPECT_FALSE(r()->Resolve());

src/transform/binding_remapper.cc

@@ -142,7 +142,7 @@ void BindingRemapper::Run(CloneContext& ctx,
         const ast::Type* inner_ty = CreateASTTypeFor(ctx, ty);
         auto* new_var = ctx.dst->create<ast::Variable>(
             ctx.Clone(var->source), ctx.Clone(var->symbol),
-            var->declared_storage_class, ac, inner_ty, var->is_const,
+            var->declared_storage_class, ac, inner_ty, false, false,
             ctx.Clone(var->constructor), ctx.Clone(var->attributes));
         ctx.Replace(var, new_var);
       }

src/transform/calculate_array_length.cc

@@ -108,7 +108,7 @@ void CalculateArrayLength::Run(CloneContext& ctx,
               // in order for HLSL to emit this as a ByteAddressBuffer.
               ctx.dst->create<ast::Variable>(
                   ctx.dst->Sym("buffer"), ast::StorageClass::kStorage,
-                  ast::Access::kUndefined, type, true, nullptr,
+                  ast::Access::kUndefined, type, true, false, nullptr,
                   ast::AttributeList{disable_validation}),
               ctx.dst->Param("result",
                              ctx.dst->ty.pointer(ctx.dst->ty.u32(),

src/transform/decompose_memory_access.cc

@@ -457,7 +457,7 @@ struct DecomposeMemoryAccess::State {
               // array.
               b.create<ast::Variable>(b.Sym("buffer"), storage_class,
                                       var_user->Variable()->Access(),
-                                      buf_ast_ty, true, nullptr,
+                                      buf_ast_ty, true, false, nullptr,
                                       ast::AttributeList{disable_validation}),
               b.Param("offset", b.ty.u32()),
           };
@@ -556,7 +556,7 @@ struct DecomposeMemoryAccess::State {
 
               b.create<ast::Variable>(b.Sym("buffer"), storage_class,
                                       var_user->Variable()->Access(),
-                                      buf_ast_ty, true, nullptr,
+                                      buf_ast_ty, true, false, nullptr,
                                       ast::AttributeList{disable_validation}),
               b.Param("offset", b.ty.u32()),
               b.Param("value", el_ast_ty),
@@ -656,7 +656,7 @@ struct DecomposeMemoryAccess::State {
           // order for HLSL to emit this as a ByteAddressBuffer.
           b.create<ast::Variable>(b.Sym("buffer"), ast::StorageClass::kStorage,
                                   var_user->Variable()->Access(), buf_ast_ty,
-                                  true, nullptr,
+                                  true, false, nullptr,
                                   ast::AttributeList{disable_validation}),
           b.Param("offset", b.ty.u32()),
       };

src/transform/external_texture_transform.cc

@@ -116,7 +116,7 @@ void ExternalTextureTransform::Run(CloneContext& ctx,
         auto clonedAttributes = ctx.Clone(var->attributes);
         auto* newVar = ctx.dst->create<ast::Variable>(
             clonedSrc, clonedSym, var->declared_storage_class,
-            var->declared_access, newType, var->is_const, clonedConstructor,
+            var->declared_access, newType, false, false, clonedConstructor,
             clonedAttributes);
 
         ctx.Replace(var, newVar);

src/transform/renamer_test.cc

@@ -1201,7 +1201,7 @@ INSTANTIATE_TEST_SUITE_P(
         "operator",
         "or",
         "or_eq",
-        "override",
+        // "override", // Also used in WGSL
         // "private",  // Also used in WGSL
         "protected",
         "public",

src/transform/robustness_test.cc

@@ -1056,7 +1056,7 @@ var<private> a : mat3x2<f32>;
 
 // TODO(dsinclair): Implement when constant_id exists
 TEST_F(RobustnessTest, DISABLED_Vector_Constant_Id_Clamps) {
-  // @override(1300) let idx : i32;
+  // @id(1300) override idx : i32;
   // var a : vec3<f32>
   // var b : f32 = a[idx]
   //
@@ -1065,7 +1065,7 @@ TEST_F(RobustnessTest, DISABLED_Vector_Constant_Id_Clamps) {
 
 // TODO(dsinclair): Implement when constant_id exists
 TEST_F(RobustnessTest, DISABLED_Array_Constant_Id_Clamps) {
-  // @override(1300) let idx : i32;
+  // @id(1300) override idx : i32;
   // var a : array<f32, 4>
   // var b : f32 = a[idx]
   //
@@ -1074,7 +1074,7 @@ TEST_F(RobustnessTest, DISABLED_Array_Constant_Id_Clamps) {
 
 // TODO(dsinclair): Implement when constant_id exists
 TEST_F(RobustnessTest, DISABLED_Matrix_Column_Constant_Id_Clamps) {
-  // @override(1300) let idx : i32;
+  // @id(1300) override idx : i32;
   // var a : mat3x2<f32>
   // var b : f32 = a[idx][1]
   //
@@ -1083,7 +1083,7 @@ TEST_F(RobustnessTest, DISABLED_Matrix_Column_Constant_Id_Clamps) {
 
 // TODO(dsinclair): Implement when constant_id exists
 TEST_F(RobustnessTest, DISABLED_Matrix_Row_Constant_Id_Clamps) {
-  // @override(1300) let idx : i32;
+  // @id(1300) override idx : i32;
   // var a : mat3x2<f32>
   // var b : f32 = a[1][idx]
   //

src/transform/single_entry_point.cc

@@ -77,19 +77,15 @@ void SingleEntryPoint::Run(CloneContext& ctx,
       ctx.dst->AST().AddTypeDecl(ctx.Clone(ty));
     } else if (auto* var = decl->As<ast::Variable>()) {
       if (referenced_vars.count(var)) {
-        if (var->is_const) {
+        if (var->is_overridable) {
-          if (auto* attr =
+          // It is an overridable constant
-                  ast::GetAttribute<ast::OverrideAttribute>(var->attributes)) {
+          if (!ast::HasAttribute<ast::IdAttribute>(var->attributes)) {
-            // It is an overridable constant
+            // If the constant doesn't already have an @id() attribute, add one
-            if (!attr->has_value) {
+            // so that its allocated ID so that it won't be affected by other
-              // If the attribute doesn't have numeric ID specified explicitly
+            // stripped away constants
-              // Make their ids explicitly assigned in the attribute so that
+            auto* global = sem.Get(var)->As<sem::GlobalVariable>();
-              // they won't be affected by other stripped away constants
+            const auto* id = ctx.dst->Id(global->ConstantId());
-              auto* global = sem.Get(var)->As<sem::GlobalVariable>();
+            ctx.InsertFront(var->attributes, id);
-              const auto* new_deco =
-                  ctx.dst->Override(attr->source, global->ConstantId());
-              ctx.Replace(attr, new_deco);
-            }
           }
         }
         ctx.dst->AST().AddGlobalVariable(ctx.Clone(var));

src/transform/single_entry_point_test.cc

@@ -258,10 +258,10 @@ fn main() {
 
 TEST_F(SingleEntryPointTest, OverridableConstants) {
   auto* src = R"(
-@override(1001) let c1 : u32 = 1u;
+@id(1001) override c1 : u32 = 1u;
-[[override]] let c2 : u32 = 1u;
+          override c2 : u32 = 1u;
-@override(0) let c3 : u32 = 1u;
+@id(0)    override c3 : u32 = 1u;
-@override(9999) let c4 : u32 = 1u;
+@id(9999) override c4 : u32 = 1u;
 
 @stage(compute) @workgroup_size(1)
 fn comp_main1() {
@@ -292,7 +292,7 @@ fn comp_main5() {
   {
     SingleEntryPoint::Config cfg("comp_main1");
     auto* expect = R"(
-@override(1001) let c1 : u32 = 1u;
+@id(1001) override c1 : u32 = 1u;
 
 @stage(compute) @workgroup_size(1)
 fn comp_main1() {
@@ -310,7 +310,7 @@ fn comp_main1() {
     // The decorator is replaced with the one with explicit id
     // And should not be affected by other constants stripped away
     auto* expect = R"(
-@override(1) let c2 : u32 = 1u;
+@id(1) override c2 : u32 = 1u;
 
 @stage(compute) @workgroup_size(1)
 fn comp_main2() {
@@ -326,7 +326,7 @@ fn comp_main2() {
   {
     SingleEntryPoint::Config cfg("comp_main3");
     auto* expect = R"(
-@override(0) let c3 : u32 = 1u;
+@id(0) override c3 : u32 = 1u;
 
 @stage(compute) @workgroup_size(1)
 fn comp_main3() {
@@ -342,7 +342,7 @@ fn comp_main3() {
   {
     SingleEntryPoint::Config cfg("comp_main4");
     auto* expect = R"(
-@override(9999) let c4 : u32 = 1u;
+@id(9999) override c4 : u32 = 1u;
 
 @stage(compute) @workgroup_size(1)
 fn comp_main4() {

src/transform/unshadow.cc

@@ -57,7 +57,7 @@ struct Unshadow::State {
       auto attributes = ctx.Clone(decl->attributes);
       return ctx.dst->create<ast::Variable>(
           source, symbol, decl->declared_storage_class, decl->declared_access,
-          type, decl->is_const, constructor, attributes);
+          type, decl->is_const, decl->is_overridable, constructor, attributes);
     };
 
     ctx.ReplaceAll([&](const ast::Variable* var) -> const ast::Variable* {

src/transform/zero_init_workgroup_memory_test.cc

@@ -506,7 +506,7 @@ var<workgroup> b : S;
 
 var<workgroup> c : array<S, 32>;
 
-@override(1) let X : i32;
+@id(1) override X : i32;
 
 @stage(compute) @workgroup_size(2, 3, X)
 fn f(@builtin(local_invocation_index) local_idx : u32) {
@@ -528,7 +528,7 @@ var<workgroup> b : S;
 
 var<workgroup> c : array<S, 32>;
 
-@override(1) let X : i32;
+@id(1) override X : i32;
 
 @stage(compute) @workgroup_size(2, 3, X)
 fn f(@builtin(local_invocation_index) local_idx : u32) {
@@ -576,7 +576,7 @@ var<workgroup> b : S;
 
 var<workgroup> c : array<S, 32>;
 
-@override(1) let X : u32;
+@id(1) override X : u32;
 
 @stage(compute) @workgroup_size(5u, X, 10u)
 fn f(@builtin(local_invocation_index) local_idx : u32) {
@@ -599,7 +599,7 @@ var<workgroup> b : S;
 
 var<workgroup> c : array<S, 32>;
 
-@override(1) let X : u32;
+@id(1) override X : u32;
 
 @stage(compute) @workgroup_size(5u, X, 10u)
 fn f(@builtin(local_invocation_index) local_idx : u32) {

src/writer/glsl/generator_impl.cc

@@ -23,9 +23,9 @@
 
 #include "src/ast/call_statement.h"
 #include "src/ast/fallthrough_statement.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/internal_attribute.h"
 #include "src/ast/interpolate_attribute.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/variable_decl_statement.h"
 #include "src/debug.h"
 #include "src/sem/array.h"
@@ -2607,7 +2607,7 @@ bool GeneratorImpl::EmitVariable(const ast::Variable* var) {
 
 bool GeneratorImpl::EmitProgramConstVariable(const ast::Variable* var) {
   for (auto* d : var->attributes) {
-    if (!d->Is<ast::OverrideAttribute>()) {
+    if (!d->Is<ast::IdAttribute>()) {
       diagnostics_.add_error(diag::System::Writer,
                              "Decorated const values not valid");
       return false;

src/writer/glsl/generator_impl_function_test.cc

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

src/writer/glsl/generator_impl_module_constant_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/writer/glsl/test_helper.h"
 
 namespace tint {
@@ -33,10 +33,10 @@ TEST_F(GlslGeneratorImplTest_ModuleConstant, Emit_ModuleConstant) {
 }
 
 TEST_F(GlslGeneratorImplTest_ModuleConstant, Emit_SpecConstant) {
-  auto* var = GlobalConst("pos", ty.f32(), Expr(3.0f),
+  auto* var = Override("pos", ty.f32(), Expr(3.0f),
-                          ast::AttributeList{
+                       ast::AttributeList{
-                              create<ast::OverrideAttribute>(23),
+                           Id(23),
-                          });
+                       });
 
   GeneratorImpl& gen = Build();
 
@@ -49,10 +49,10 @@ const float pos = WGSL_SPEC_CONSTANT_23;
 }
 
 TEST_F(GlslGeneratorImplTest_ModuleConstant, Emit_SpecConstant_NoConstructor) {
-  auto* var = GlobalConst("pos", ty.f32(), nullptr,
+  auto* var = Override("pos", ty.f32(), nullptr,
-                          ast::AttributeList{
+                       ast::AttributeList{
-                              create<ast::OverrideAttribute>(23),
+                           Id(23),
-                          });
+                       });
 
   GeneratorImpl& gen = Build();
 
@@ -65,14 +65,11 @@ const float pos = WGSL_SPEC_CONSTANT_23;
 }
 
 TEST_F(GlslGeneratorImplTest_ModuleConstant, Emit_SpecConstant_NoId) {
-  auto* a = GlobalConst("a", ty.f32(), Expr(3.0f),
+  auto* a = Override("a", ty.f32(), Expr(3.0f),
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(0),
+                         Id(0),
-                        });
+                     });
-  auto* b = GlobalConst("b", ty.f32(), Expr(2.0f),
+  auto* b = Override("b", ty.f32(), Expr(2.0f));
-                        ast::AttributeList{
-                            create<ast::OverrideAttribute>(),
-                        });
 
   GeneratorImpl& gen = Build();
 

src/writer/glsl/generator_impl_workgroup_var_test.cc

@@ -13,7 +13,7 @@
 // limitations under the License.
 
 #include "gmock/gmock.h"
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/stage_attribute.h"
 #include "src/writer/glsl/test_helper.h"
 

src/writer/hlsl/generator_impl.cc

@@ -24,9 +24,9 @@
 
 #include "src/ast/call_statement.h"
 #include "src/ast/fallthrough_statement.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/internal_attribute.h"
 #include "src/ast/interpolate_attribute.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/variable_decl_statement.h"
 #include "src/debug.h"
 #include "src/sem/array.h"
@@ -3898,7 +3898,7 @@ bool GeneratorImpl::EmitVariable(const ast::Variable* var) {
 
 bool GeneratorImpl::EmitProgramConstVariable(const ast::Variable* var) {
   for (auto* d : var->attributes) {
-    if (!d->Is<ast::OverrideAttribute>()) {
+    if (!d->Is<ast::IdAttribute>()) {
       diagnostics_.add_error(diag::System::Writer,
                              "Decorated const values not valid");
       return false;

src/writer/hlsl/generator_impl_function_test.cc

@@ -745,9 +745,9 @@ void main() {
 
 TEST_F(HlslGeneratorImplTest_Function,
        Emit_Attribute_EntryPoint_Compute_WithWorkgroup_OverridableConst) {
-  GlobalConst("width", ty.i32(), Construct(ty.i32(), 2), {Override(7u)});
+  Override("width", ty.i32(), Construct(ty.i32(), 2), {Id(7u)});
-  GlobalConst("height", ty.i32(), Construct(ty.i32(), 3), {Override(8u)});
+  Override("height", ty.i32(), Construct(ty.i32(), 3), {Id(8u)});
-  GlobalConst("depth", ty.i32(), Construct(ty.i32(), 4), {Override(9u)});
+  Override("depth", ty.i32(), Construct(ty.i32(), 4), {Id(9u)});
   Func("main", ast::VariableList{}, ty.void_(), {},
        {
            Stage(ast::PipelineStage::kCompute),

src/writer/hlsl/generator_impl_module_constant_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/writer/hlsl/test_helper.h"
 
 namespace tint {
@@ -33,10 +33,10 @@ TEST_F(HlslGeneratorImplTest_ModuleConstant, Emit_ModuleConstant) {
 }
 
 TEST_F(HlslGeneratorImplTest_ModuleConstant, Emit_SpecConstant) {
-  auto* var = GlobalConst("pos", ty.f32(), Expr(3.0f),
+  auto* var = Override("pos", ty.f32(), Expr(3.0f),
-                          ast::AttributeList{
+                       ast::AttributeList{
-                              create<ast::OverrideAttribute>(23),
+                           Id(23),
-                          });
+                       });
 
   GeneratorImpl& gen = Build();
 
@@ -49,10 +49,10 @@ static const float pos = WGSL_SPEC_CONSTANT_23;
 }
 
 TEST_F(HlslGeneratorImplTest_ModuleConstant, Emit_SpecConstant_NoConstructor) {
-  auto* var = GlobalConst("pos", ty.f32(), nullptr,
+  auto* var = Override("pos", ty.f32(), nullptr,
-                          ast::AttributeList{
+                       ast::AttributeList{
-                              create<ast::OverrideAttribute>(23),
+                           Id(23),
-                          });
+                       });
 
   GeneratorImpl& gen = Build();
 
@@ -65,14 +65,11 @@ static const float pos = WGSL_SPEC_CONSTANT_23;
 }
 
 TEST_F(HlslGeneratorImplTest_ModuleConstant, Emit_SpecConstant_NoId) {
-  auto* a = GlobalConst("a", ty.f32(), Expr(3.0f),
+  auto* a = Override("a", ty.f32(), Expr(3.0f),
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(0),
+                         Id(0),
-                        });
+                     });
-  auto* b = GlobalConst("b", ty.f32(), Expr(2.0f),
+  auto* b = Override("b", ty.f32(), Expr(2.0f));
-                        ast::AttributeList{
-                            create<ast::OverrideAttribute>(),
-                        });
 
   GeneratorImpl& gen = Build();
 

src/writer/hlsl/generator_impl_workgroup_var_test.cc

@@ -13,7 +13,7 @@
 // limitations under the License.
 
 #include "gmock/gmock.h"
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/stage_attribute.h"
 #include "src/writer/hlsl/test_helper.h"
 

src/writer/msl/generator_impl.cc

@@ -27,9 +27,9 @@
 #include "src/ast/disable_validation_attribute.h"
 #include "src/ast/fallthrough_statement.h"
 #include "src/ast/float_literal_expression.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/interpolate_attribute.h"
 #include "src/ast/module.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/sint_literal_expression.h"
 #include "src/ast/uint_literal_expression.h"
 #include "src/ast/variable_decl_statement.h"
@@ -2823,7 +2823,7 @@ bool GeneratorImpl::EmitVariable(const sem::Variable* var) {
 
 bool GeneratorImpl::EmitProgramConstVariable(const ast::Variable* var) {
   for (auto* d : var->attributes) {
-    if (!d->Is<ast::OverrideAttribute>()) {
+    if (!d->Is<ast::IdAttribute>()) {
       diagnostics_.add_error(diag::System::Writer,
                              "Decorated const values not valid");
       return false;

src/writer/msl/generator_impl_module_constant_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/writer/msl/test_helper.h"
 
 namespace tint {
@@ -33,10 +33,10 @@ TEST_F(MslGeneratorImplTest, Emit_ModuleConstant) {
 }
 
 TEST_F(MslGeneratorImplTest, Emit_SpecConstant) {
-  auto* var = GlobalConst("pos", ty.f32(), Expr(3.f),
+  auto* var = Override("pos", ty.f32(), Expr(3.f),
-                          ast::AttributeList{
+                       ast::AttributeList{
-                              create<ast::OverrideAttribute>(23),
+                           Id(23),
-                          });
+                       });
 
   GeneratorImpl& gen = Build();
 
@@ -45,14 +45,11 @@ TEST_F(MslGeneratorImplTest, Emit_SpecConstant) {
 }
 
 TEST_F(MslGeneratorImplTest, Emit_SpecConstant_NoId) {
-  auto* var_a = GlobalConst("a", ty.f32(), nullptr,
+  auto* var_a = Override("a", ty.f32(), nullptr,
-                            ast::AttributeList{
+                         ast::AttributeList{
-                                create<ast::OverrideAttribute>(0),
+                             Id(0),
-                            });
+                         });
-  auto* var_b = GlobalConst("b", ty.f32(), nullptr,
+  auto* var_b = Override("b", ty.f32(), nullptr);
-                            ast::AttributeList{
-                                create<ast::OverrideAttribute>(),
-                            });
 
   GeneratorImpl& gen = Build();
 

src/writer/spirv/builder.cc

@@ -20,8 +20,8 @@
 #include "spirv/unified1/GLSL.std.450.h"
 #include "src/ast/call_statement.h"
 #include "src/ast/fallthrough_statement.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/internal_attribute.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/traverse_expressions.h"
 #include "src/sem/array.h"
 #include "src/sem/atomic_type.h"
@@ -770,9 +770,8 @@ bool Builder::GenerateGlobalVariable(const ast::Variable* var) {
 
   if (var->is_const) {
     if (!var->constructor) {
-      // Constants must have an initializer unless they have an override
+      // Constants must have an initializer unless they are overridable.
-      // attribute.
+      if (!var->is_overridable) {
-      if (!ast::HasAttribute<ast::OverrideAttribute>(var->attributes)) {
         error_ = "missing constructor for constant";
         return false;
       }
@@ -905,7 +904,7 @@ bool Builder::GenerateGlobalVariable(const ast::Variable* var) {
                Operand::Int(group->value)});
           return true;
         },
-        [&](const ast::OverrideAttribute*) {
+        [&](const ast::IdAttribute*) {
           return true;  // Spec constants are handled elsewhere
         },
         [&](const ast::InternalAttribute*) {

src/writer/spirv/builder_function_attribute_test.cc

@@ -155,9 +155,9 @@ TEST_F(BuilderTest, Decoration_ExecutionMode_WorkgroupSize_Const) {
 }
 
 TEST_F(BuilderTest, Decoration_ExecutionMode_WorkgroupSize_OverridableConst) {
-  GlobalConst("width", ty.i32(), Construct(ty.i32(), 2), {Override(7u)});
+  Override("width", ty.i32(), Construct(ty.i32(), 2), {Id(7u)});
-  GlobalConst("height", ty.i32(), Construct(ty.i32(), 3), {Override(8u)});
+  Override("height", ty.i32(), Construct(ty.i32(), 3), {Id(8u)});
-  GlobalConst("depth", ty.i32(), Construct(ty.i32(), 4), {Override(9u)});
+  Override("depth", ty.i32(), Construct(ty.i32(), 4), {Id(9u)});
   auto* func = Func("main", {}, ty.void_(), ast::StatementList{},
                     ast::AttributeList{
                         WorkgroupSize("width", "height", "depth"),
@@ -185,7 +185,7 @@ OpDecorate %3 BuiltIn WorkgroupSize
 }
 
 TEST_F(BuilderTest, Decoration_ExecutionMode_WorkgroupSize_LiteralAndConst) {
-  GlobalConst("height", ty.i32(), Construct(ty.i32(), 2), {Override(7u)});
+  Override("height", ty.i32(), Construct(ty.i32(), 2), {Id(7u)});
   GlobalConst("depth", ty.i32(), Construct(ty.i32(), 3));
   auto* func = Func("main", {}, ty.void_(), ast::StatementList{},
                     ast::AttributeList{

src/writer/spirv/builder_global_variable_test.cc

@@ -12,7 +12,7 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/stage_attribute.h"
 #include "src/ast/struct_block_attribute.h"
 #include "src/writer/spirv/spv_dump.h"
@@ -150,10 +150,10 @@ OpDecorate %1 DescriptorSet 3
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Bool) {
-  auto* v = GlobalConst("var", ty.bool_(), Expr(true),
+  auto* v = Override("var", ty.bool_(), Expr(true),
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(1200),
+                         Id(1200),
-                        });
+                     });
 
   spirv::Builder& b = Build();
 
@@ -168,10 +168,10 @@ TEST_F(BuilderTest, GlobalVar_Override_Bool) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Bool_ZeroValue) {
-  auto* v = GlobalConst("var", ty.bool_(), Construct<bool>(),
+  auto* v = Override("var", ty.bool_(), Construct<bool>(),
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(1200),
+                         Id(1200),
-                        });
+                     });
 
   spirv::Builder& b = Build();
 
@@ -186,10 +186,10 @@ TEST_F(BuilderTest, GlobalVar_Override_Bool_ZeroValue) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Bool_NoConstructor) {
-  auto* v = GlobalConst("var", ty.bool_(), nullptr,
+  auto* v = Override("var", ty.bool_(), nullptr,
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(1200),
+                         Id(1200),
-                        });
+                     });
 
   spirv::Builder& b = Build();
 
@@ -204,10 +204,10 @@ TEST_F(BuilderTest, GlobalVar_Override_Bool_NoConstructor) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar) {
-  auto* v = GlobalConst("var", ty.f32(), Expr(2.f),
+  auto* v = Override("var", ty.f32(), Expr(2.f),
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(0),
+                         Id(0),
-                        });
+                     });
 
   spirv::Builder& b = Build();
 
@@ -222,10 +222,10 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar_ZeroValue) {
-  auto* v = GlobalConst("var", ty.f32(), Construct<f32>(),
+  auto* v = Override("var", ty.f32(), Construct<f32>(),
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(0),
+                         Id(0),
-                        });
+                     });
 
   spirv::Builder& b = Build();
 
@@ -240,10 +240,10 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar_ZeroValue) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar_F32_NoConstructor) {
-  auto* v = GlobalConst("var", ty.f32(), nullptr,
+  auto* v = Override("var", ty.f32(), nullptr,
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(0),
+                         Id(0),
-                        });
+                     });
 
   spirv::Builder& b = Build();
 
@@ -258,10 +258,10 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar_F32_NoConstructor) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar_I32_NoConstructor) {
-  auto* v = GlobalConst("var", ty.i32(), nullptr,
+  auto* v = Override("var", ty.i32(), nullptr,
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(0),
+                         Id(0),
-                        });
+                     });
 
   spirv::Builder& b = Build();
 
@@ -276,10 +276,10 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar_I32_NoConstructor) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_Scalar_U32_NoConstructor) {
-  auto* v = GlobalConst("var", ty.u32(), nullptr,
+  auto* v = Override("var", ty.u32(), nullptr,
-                        ast::AttributeList{
+                     ast::AttributeList{
-                            create<ast::OverrideAttribute>(0),
+                         Id(0),
-                        });
+                     });
 
   spirv::Builder& b = Build();
 
@@ -294,14 +294,11 @@ TEST_F(BuilderTest, GlobalVar_Override_Scalar_U32_NoConstructor) {
 }
 
 TEST_F(BuilderTest, GlobalVar_Override_NoId) {
-  auto* var_a = GlobalConst("a", ty.bool_(), Expr(true),
+  auto* var_a = Override("a", ty.bool_(), Expr(true),
-                            ast::AttributeList{
+                         ast::AttributeList{
-                                create<ast::OverrideAttribute>(0),
+                             Id(0),
-                            });
+                         });
-  auto* var_b = GlobalConst("b", ty.bool_(), Expr(false),
+  auto* var_b = Override("b", ty.bool_(), Expr(false));
-                            ast::AttributeList{
-                                create<ast::OverrideAttribute>(),
-                            });
 
   spirv::Builder& b = Build();
 

src/writer/wgsl/generator_impl.cc

@@ -28,13 +28,13 @@
 #include "src/ast/f32.h"
 #include "src/ast/float_literal_expression.h"
 #include "src/ast/i32.h"
+#include "src/ast/id_attribute.h"
 #include "src/ast/internal_attribute.h"
 #include "src/ast/interpolate_attribute.h"
 #include "src/ast/invariant_attribute.h"
 #include "src/ast/matrix.h"
 #include "src/ast/module.h"
 #include "src/ast/multisampled_texture.h"
-#include "src/ast/override_attribute.h"
 #include "src/ast/pointer.h"
 #include "src/ast/sampled_texture.h"
 #include "src/ast/sint_literal_expression.h"
@@ -646,7 +646,9 @@ bool GeneratorImpl::EmitVariable(std::ostream& out, const ast::Variable* var) {
     out << " ";
   }
 
-  if (var->is_const) {
+  if (var->is_overridable) {
+    out << "override";
+  } else if (var->is_const) {
     out << "let";
   } else {
     out << "var";
@@ -747,11 +749,8 @@ bool GeneratorImpl::EmitAttributes(std::ostream& out,
           out << "invariant";
           return true;
         },
-        [&](const ast::OverrideAttribute* override_deco) {
+        [&](const ast::IdAttribute* override_deco) {
-          out << "override";
+          out << "id(" << override_deco->value << ")";
-          if (override_deco->has_value) {
-            out << "(" << override_deco->value << ")";
-          }
           return true;
         },
         [&](const ast::StructMemberSizeAttribute* size) {

src/writer/wgsl/generator_impl_global_decl_test.cc

@@ -132,17 +132,17 @@ TEST_F(WgslGeneratorImplTest, Emit_Global_Texture) {
 }
 
 TEST_F(WgslGeneratorImplTest, Emit_OverridableConstants) {
-  GlobalConst("a", ty.f32(), nullptr, {Override()});
+  Override("a", ty.f32(), nullptr);
-  GlobalConst("b", ty.f32(), nullptr, {Override(7u)});
+  Override("b", ty.f32(), nullptr, {Id(7u)});
 
   GeneratorImpl& gen = Build();
 
   gen.increment_indent();
 
   ASSERT_TRUE(gen.Generate()) << gen.error();
-  EXPECT_EQ(gen.result(), R"(  @override let a : f32;
+  EXPECT_EQ(gen.result(), R"(  override a : f32;
 
-  @override(7) let b : f32;
+  @id(7) override b : f32;
 )");
 }
 

src/writer/wgsl/generator_impl_literal_test.cc

@@ -14,7 +14,6 @@
 
 #include <cstring>
 
-#include "src/ast/override_attribute.h"
 #include "src/writer/wgsl/test_helper.h"
 
 namespace tint {

src/writer/wgsl/generator_impl_variable_test.cc

@@ -12,7 +12,6 @@
 // See the License for the specific language governing permissions and
 // limitations under the License.
 
-#include "src/ast/override_attribute.h"
 #include "src/ast/struct_block_attribute.h"
 #include "src/writer/wgsl/test_helper.h"
 

test/BUILD.gn

@@ -171,6 +171,7 @@ tint_unittests_source_set("tint_unittests_ast_src") {
     "../src/ast/function_test.cc",
     "../src/ast/group_attribute_test.cc",
     "../src/ast/i32_test.cc",
+    "../src/ast/id_attribute_test.cc",
     "../src/ast/identifier_expression_test.cc",
     "../src/ast/if_statement_test.cc",
     "../src/ast/index_accessor_expression_test.cc",
@@ -184,7 +185,6 @@ tint_unittests_source_set("tint_unittests_ast_src") {
     "../src/ast/module_clone_test.cc",
     "../src/ast/module_test.cc",
     "../src/ast/multisampled_texture_test.cc",
-    "../src/ast/override_attribute_test.cc",
     "../src/ast/phony_expression_test.cc",
     "../src/ast/pointer_test.cc",
     "../src/ast/return_statement_test.cc",

test/var/override/named/no_init/bool.wgsl

@@ -1,4 +1,4 @@
-@override let o : bool;
+override o : bool;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/no_init/bool.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : bool;
+override o : bool;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/no_init/f32.wgsl

@@ -1,4 +1,4 @@
-@override let o : f32;
+override o : f32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/no_init/f32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : f32;
+override o : f32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/no_init/i32.wgsl

@@ -1,4 +1,4 @@
-@override let o : i32;
+override o : i32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/no_init/i32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : i32;
+override o : i32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/no_init/u32.wgsl

@@ -1,4 +1,4 @@
-@override let o : u32;
+override o : u32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/no_init/u32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : u32;
+override o : u32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/val_init/bool.wgsl

@@ -1,4 +1,4 @@
-@override let o : bool = true;
+override o : bool = true;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/val_init/bool.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : bool = true;
+override o : bool = true;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/val_init/f32.wgsl

@@ -1,4 +1,4 @@
-@override let o : f32 = 1.0;
+override o : f32 = 1.0;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/val_init/f32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : f32 = 1.0;
+override o : f32 = 1.0;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/val_init/i32.wgsl

@@ -1,4 +1,4 @@
-@override let o : i32 = 1;
+override o : i32 = 1;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/val_init/i32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : i32 = 1;
+override o : i32 = 1;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/val_init/u32.wgsl

@@ -1,4 +1,4 @@
-@override let o : u32 = 1u;
+override o : u32 = 1u;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/val_init/u32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : u32 = 1u;
+override o : u32 = 1u;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/zero_init/bool.wgsl

@@ -1,4 +1,4 @@
-@override let o : bool = bool();
+override o : bool = bool();
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/zero_init/bool.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : bool = bool();
+override o : bool = bool();
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/zero_init/f32.wgsl

@@ -1,4 +1,4 @@
-@override let o : f32 = f32();
+override o : f32 = f32();
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/zero_init/f32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : f32 = f32();
+override o : f32 = f32();
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/zero_init/i32.wgsl

@@ -1,4 +1,4 @@
-@override let o : i32 = i32();
+override o : i32 = i32();
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/zero_init/i32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : i32 = i32();
+override o : i32 = i32();
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/zero_init/u32.wgsl

@@ -1,4 +1,4 @@
-@override let o : u32 = u32();
+override o : u32 = u32();
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/named/zero_init/u32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override let o : u32 = u32();
+override o : u32 = u32();
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/no_init/bool.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : bool;
+@id(1234) override o : bool;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/no_init/bool.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : bool;
+@id(1234) override o : bool;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/no_init/f32.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : f32;
+@id(1234) override o : f32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/no_init/f32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : f32;
+@id(1234) override o : f32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/no_init/i32.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : i32;
+@id(1234) override o : i32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/no_init/i32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : i32;
+@id(1234) override o : i32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/no_init/u32.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : u32;
+@id(1234) override o : u32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/no_init/u32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : u32;
+@id(1234) override o : u32;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/val_init/bool.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : bool = true;
+@id(1234) override o : bool = true;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/val_init/bool.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : bool = true;
+@id(1234) override o : bool = true;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/val_init/f32.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : f32 = 1.0;
+@id(1234) override o : f32 = 1.0;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/val_init/f32.wgsl.expected.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : f32 = 1.0;
+@id(1234) override o : f32 = 1.0;
 
 @stage(compute) @workgroup_size(1)
 fn main() {

test/var/override/numbered/val_init/i32.wgsl

@@ -1,4 +1,4 @@
-@override(1234) let o : i32 = 1;
+@id(1234) override o : i32 = 1;
 
 @stage(compute) @workgroup_size(1)
 fn main() {