diff --git a/docs/origin-trial-changes.md b/docs/origin-trial-changes.md index 71be9107b3..9e19b620ec 100644 --- a/docs/origin-trial-changes.md +++ b/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 diff --git a/src/BUILD.gn b/src/BUILD.gn index 62e24d2950..ca35959f72 100644 --- a/src/BUILD.gn +++ b/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", diff --git a/src/CMakeLists.txt b/src/CMakeLists.txt index dff6c748b6..6c83d79fdd 100644 --- a/src/CMakeLists.txt +++ b/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 diff --git a/src/ast/group_attribute_test.cc b/src/ast/group_attribute_test.cc index f65066f0a2..a124170486 100644 --- a/src/ast/group_attribute_test.cc +++ b/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 { diff --git a/src/ast/id_attribute.cc b/src/ast/id_attribute.cc new file mode 100644 index 0000000000..608625fe75 --- /dev/null +++ b/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 + +#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(src, value); +} + +} // namespace ast +} // namespace tint diff --git a/src/ast/override_attribute.h b/src/ast/id_attribute.h similarity index 56% rename from src/ast/override_attribute.h rename to src/ast/id_attribute.h index f3cbcc5cfc..db12aa1645 100644 --- a/src/ast/override_attribute.h +++ b/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_ -#define SRC_AST_OVERRIDE_ATTRIBUTE_H_ +#ifndef SRC_AST_ID_ATTRIBUTE_H_ +#define SRC_AST_ID_ATTRIBUTE_H_ #include @@ -22,19 +22,15 @@ namespace tint { namespace ast { -/// An override attribute -class OverrideAttribute : public Castable { +/// An id attribute for pipeline-overridable constants +class IdAttribute : public Castable { 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); - /// Create an override attribute with a specific id value. - /// @param pid the identifier of the program that owns this node - /// @param src the source of this node - /// @param val the override value - OverrideAttribute(ProgramID pid, const Source& src, uint32_t val); - ~OverrideAttribute() override; + /// @param val the numeric id value + IdAttribute(ProgramID pid, const Source& src, uint32_t val); + ~IdAttribute() override; /// @returns the WGSL name for the attribute std::string Name() const override; @@ -43,16 +39,13 @@ class OverrideAttribute : public Castable { /// `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 - const bool has_value; - - /// The override id value + /// The id value const uint32_t value; }; } // namespace ast } // namespace tint -#endif // SRC_AST_OVERRIDE_ATTRIBUTE_H_ +#endif // SRC_AST_ID_ATTRIBUTE_H_ diff --git a/src/ast/override_attribute_test.cc b/src/ast/id_attribute_test.cc similarity index 69% rename from src/ast/override_attribute_test.cc rename to src/ast/id_attribute_test.cc index df6afba495..d40db86d39 100644 --- a/src/ast/override_attribute_test.cc +++ b/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) { - auto* d = create(12); - EXPECT_TRUE(d->has_value); +TEST_F(IdAttributeTest, Creation) { + auto* d = create(12); EXPECT_EQ(12u, d->value); } -TEST_F(OverrideAttributeTest, Creation_WithoutValue) { - auto* d = create(); - EXPECT_FALSE(d->has_value); -} - } // namespace } // namespace ast } // namespace tint diff --git a/src/ast/location_attribute_test.cc b/src/ast/location_attribute_test.cc index e2ebc23dd6..60941f7c1c 100644 --- a/src/ast/location_attribute_test.cc +++ b/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 { diff --git a/src/ast/override_attribute.cc b/src/ast/override_attribute.cc deleted file mode 100644 index 62993a95de..0000000000 --- a/src/ast/override_attribute.cc +++ /dev/null @@ -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 - -#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(src, value); - } else { - return ctx->dst->create(src); - } -} - -} // namespace ast -} // namespace tint diff --git a/src/ast/variable.cc b/src/ast/variable.cc index 3609700718..6e3f97deb4 100644 --- a/src/ast/variable.cc +++ b/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(src, sym, declared_storage_class, - declared_access, ty, is_const, ctor, attrs); + declared_access, ty, is_const, + is_overridable, ctor, attrs); } } // namespace ast diff --git a/src/ast/variable.h b/src/ast/variable.h index 80ac8e3d34..7a40171b33 100644 --- a/src/ast/variable.h +++ b/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" -/// declaration, "let" declaration, or formal parameter to a function. +/// An instance of this class represents one of four constructs in WGSL: "var" +/// 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 { /// @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 { 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 { /// 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; diff --git a/src/ast/variable_test.cc b/src/ast/variable_test.cc index ab6778dc33..f126f1b956 100644 --- a/src/ast/variable_test.cc +++ b/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(1), create(Builtin::kPosition), - create(1200), + create(1200), }); auto& attributes = var->attributes; EXPECT_TRUE(ast::HasAttribute(attributes)); EXPECT_TRUE(ast::HasAttribute(attributes)); - EXPECT_TRUE(ast::HasAttribute(attributes)); + EXPECT_TRUE(ast::HasAttribute(attributes)); auto* location = ast::GetAttribute(attributes); ASSERT_NE(nullptr, location); diff --git a/src/inspector/inspector.cc b/src/inspector/inspector.cc index c897708c7a..975e73da82 100644 --- a/src/inspector/inspector.cc +++ b/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 Inspector::GetEntryPoints() { overridable_constant.is_initialized = global->Declaration()->constructor; - auto* override_attr = ast::GetAttribute( - global->Declaration()->attributes); overridable_constant.is_numeric_id_specified = - override_attr ? override_attr->has_value : false; + ast::HasAttribute( + global->Declaration()->attributes); entry_point.overridable_constants.push_back(overridable_constant); } diff --git a/src/inspector/inspector_test.cc b/src/inspector/inspector_test.cc index 790b149106..f1a4232899 100644 --- a/src/inspector/inspector_test.cc +++ b/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" diff --git a/src/inspector/test_inspector_builder.h b/src/inspector/test_inspector_builder.h index f2ac38d995..64f5c00f05 100644 --- a/src/inspector/test_inspector_builder.h +++ b/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, - ast::AttributeList{ - Override(id), - }); + return Override(name, type, constructor, {Id(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, - ast::AttributeList{ - Override(), - }); + return Override(name, type, constructor); } /// Generates a function that references module-scoped, plain-typed constant diff --git a/src/program_builder.h b/src/program_builder.h index ebc39b173b..0db8eea5fc 100644 --- a/src/program_builder.h +++ b/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)...); return create(Sym(std::forward(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)...); - return create(source, Sym(std::forward(name)), - opts.storage, opts.access, type, false, - opts.constructor, std::move(opts.attributes)); + return create( + source, Sym(std::forward(name)), opts.storage, opts.access, type, + 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( Sym(std::forward(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( source, Sym(std::forward(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( Sym(std::forward(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( source, Sym(std::forward(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 + const ast::Variable* Override(NAME&& name, + const ast::Type* type, + const ast::Expression* constructor, + ast::AttributeList attributes = {}) { + auto* var = create( + source_, Sym(std::forward(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 + const ast::Variable* Override(const Source& source, + NAME&& name, + const ast::Type* type, + const ast::Expression* constructor, + ast::AttributeList attributes = {}) { + auto* var = create( + source, Sym(std::forward(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(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) { - return create(source, id); + const ast::IdAttribute* Id(const Source& source, uint32_t id) { + return create(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) { - 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(source); - } - - /// Creates an ast::OverrideAttribute without a constant ID - /// @returns the override attribute pointer - const ast::OverrideAttribute* Override() { return Override(source_); } + const ast::IdAttribute* Id(uint32_t id) { return Id(source_, id); } /// Creates an ast::StageAttribute /// @param source the source information diff --git a/src/reader/spirv/function.cc b/src/reader/spirv/function.cc index ed35ca4e0d..33d7da2b9a 100644 --- a/src/reader/spirv/function.cc +++ b/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, - ast::AttributeList{}); + param->result_id(), ast::StorageClass::kNone, type, true, false, + 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(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( - Source{}, - parser_impl_.MakeVariable(id, ast::StorageClass::kNone, storage_type, - false, nullptr, ast::AttributeList{}))); + Source{}, parser_impl_.MakeVariable(id, ast::StorageClass::kNone, + storage_type, false, false, nullptr, + 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, - ast::AttributeList{}); + inst.result_id(), ast::StorageClass::kNone, expr.type, true, false, + expr.expr, ast::AttributeList{}); if (!ast_const) { return false; } diff --git a/src/reader/spirv/parser_impl.cc b/src/reader/spirv/parser_impl.cc index e80fdec047..c6b192ef08 100644 --- a/src/reader/spirv/parser_impl.cc +++ b/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(Source{}, id); + auto* cid = create(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, 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(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, diff --git a/src/reader/spirv/parser_impl.h b/src/reader/spirv/parser_impl.h index 592fc0a5b4..ffa62512ca 100644 --- a/src/reader/spirv/parser_impl.h +++ b/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); diff --git a/src/reader/spirv/parser_impl_module_var_test.cc b/src/reader/spirv/parser_impl_module_var_test.cc index 5e1051819e..ccd11a7a15 100644 --- a/src/reader/spirv/parser_impl_module_var_test.cc +++ b/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, - HasSubstr("@override(12) let myconst : bool = false;")) + EXPECT_THAT(module_str, HasSubstr("@id(12) override 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) { diff --git a/src/reader/wgsl/lexer.cc b/src/reader/wgsl/lexer.cc index a905ec5eb0..56e56adc5e 100644 --- a/src/reader/wgsl/lexer.cc +++ b/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") diff --git a/src/reader/wgsl/lexer_test.cc b/src/reader/wgsl/lexer_test.cc index 687c4a1d2d..ff6c9caf3f 100644 --- a/src/reader/wgsl/lexer_test.cc +++ b/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}, diff --git a/src/reader/wgsl/parser_impl.cc b/src/reader/wgsl/parser_impl.cc index e6e45500f4..8eb761ba3e 100644 --- a/src/reader/wgsl/parser_impl.cc +++ b/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 == kLocationAttribute || t == kOverrideAttribute || + t == kGroupAttribute || t == kIdAttribute || + t == kInterpolateAttribute || t == kLocationAttribute || t == kSizeAttribute || t == kStageAttribute || t == kStrideAttribute || t == kWorkgroupSizeAttribute; } @@ -485,22 +485,29 @@ Maybe 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 -// : attribute_list* LET variable_ident_decl global_const_initializer? +// global_constant_decl : +// | LET (ident | variable_ident_decl) global_const_initializer +// | attribute* override (ident | variable_ident_decl) (equal expression)? // global_const_initializer // : EQUAL const_expr Maybe 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 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 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 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 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 ParserImpl::attribute() { }); } - if (t == kOverrideAttribute) { - const char* use = "override attribute"; + if (t == kIdAttribute) { + 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)) { - // @override(x) - return expect_paren_block(use, [&]() -> Result { - auto val = expect_positive_sint(use); - if (val.errored) - return Failure::kErrored; - - return create(t.source(), val.value); - }); - } else { - // [[override]] - return create(t.source()); - } + return create(t.source(), val.value); + }); } return Failure::kNoMatch; diff --git a/src/reader/wgsl/parser_impl_error_msg_test.cc b/src/reader/wgsl/parser_impl_error_msg_test.cc index bf3d5b894a..2a800edc8b 100644 --- a/src/reader/wgsl/parser_impl_error_msg_test.cc +++ b/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; + ^^^^^^^^^ )"); } diff --git a/src/reader/wgsl/parser_impl_global_constant_decl_test.cc b/src/reader/wgsl/parser_impl_global_constant_decl_test.cc index a5c60629ea..20bc9e9211 100644 --- a/src/reader/wgsl/parser_impl_global_constant_decl_test.cc +++ b/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()); @@ -43,8 +44,6 @@ TEST_F(ParserImplTest, GlobalConstantDecl) { ASSERT_NE(e->constructor, nullptr); EXPECT_TRUE(e->constructor->Is()); - - EXPECT_FALSE(ast::HasAttribute(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()); - - EXPECT_FALSE(ast::HasAttribute(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()); 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()); auto* override_attr = - ast::GetAttribute(e.value->attributes); + ast::GetAttribute(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()); 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()); - auto* override_attr = - ast::GetAttribute(e.value->attributes); - ASSERT_NE(override_attr, nullptr); - EXPECT_FALSE(override_attr->has_value); + auto* id_attr = ast::GetAttribute(e.value->attributes); + ASSERT_EQ(id_attr, nullptr); } 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 diff --git a/src/reader/wgsl/parser_impl_struct_attribute_decl_test.cc b/src/reader/wgsl/parser_impl_struct_attribute_decl_test.cc index 1db834396e..6f1651d3b1 100644 --- a/src/reader/wgsl/parser_impl_struct_attribute_decl_test.cc +++ b/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(); - EXPECT_TRUE(override_attr->Is()); + auto* invariant = attrs.value[0]->As(); + EXPECT_TRUE(invariant->Is()); } 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(); - EXPECT_TRUE(override_attr->Is()); + auto* invariant_attr = attrs.value[0]->As(); + EXPECT_TRUE(invariant_attr->Is()); } // 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 diff --git a/src/reader/wgsl/token.cc b/src/reader/wgsl/token.cc index 6103d0ad95..c7531ba540 100644 --- a/src/reader/wgsl/token.cc +++ b/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: diff --git a/src/reader/wgsl/token.h b/src/reader/wgsl/token.h index cc40fe0564..25f55f7c69 100644 --- a/src/reader/wgsl/token.h +++ b/src/reader/wgsl/token.h @@ -187,6 +187,8 @@ class Token { kMat4x3, /// A 'mat4x4' kMat4x4, + /// A 'override' + kOverride, /// A 'private' kPrivate, /// A 'ptr' diff --git a/src/resolver/attribute_validation_test.cc b/src/resolver/attribute_validation_test.cc index ad68170d7a..4479e53b31 100644 --- a/src/resolver/attribute_validation_test.cc +++ b/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(source, 1u)}; + case AttributeKind::kId: + return {builder.create(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(source, 0u)}; case AttributeKind::kOffset: return {builder.create(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(Source{{12, 34}}), - create(Source{{56, 78}}, 1), + create(Source{{12, 34}}, 0), + create(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)"); } diff --git a/src/resolver/dependency_graph.cc b/src/resolver/dependency_graph.cc index 7aa5c52080..d3ac5f4b6e 100644 --- a/src/resolver/dependency_graph.cc +++ b/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>()) { diff --git a/src/resolver/pipeline_overridable_constant_test.cc b/src/resolver/pipeline_overridable_constant_test.cc index 074d5610b1..64ca315866 100644 --- a/src/resolver/pipeline_overridable_constant_test.cc +++ b/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 variables; - auto* a = GlobalConst("a", ty.f32(), Expr(1.f), {Override()}); - auto* b = GlobalConst("b", ty.f32(), Expr(1.f), {Override()}); - auto* c = GlobalConst("c", ty.f32(), Expr(1.f), {Override(2u)}); - auto* d = GlobalConst("d", ty.f32(), Expr(1.f), {Override(4u)}); - auto* e = GlobalConst("e", ty.f32(), Expr(1.f), {Override()}); - auto* f = GlobalConst("f", ty.f32(), Expr(1.f), {Override(1u)}); + auto* a = Override("a", ty.f32(), Expr(1.f)); + auto* b = Override("b", ty.f32(), Expr(1.f)); + auto* c = Override("c", ty.f32(), Expr(1.f), {Id(2u)}); + auto* d = Override("d", ty.f32(), Expr(1.f), {Id(4u)}); + auto* e = Override("e", ty.f32(), Expr(1.f)); + 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)}); - GlobalConst("b", ty.f32(), Expr(1.f), {Override(Source{{56, 78}}, 7u)}); + Override("a", ty.f32(), Expr(1.f), {Id(Source{{12, 34}}, 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()); diff --git a/src/resolver/resolver.cc b/src/resolver/resolver.cc index 66a6c6c5b5..98697559ee 100644 --- a/src/resolver/resolver.cc +++ b/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 && - !ast::HasAttribute(var->attributes)) { + } else if (var->is_const && !var->is_overridable && + 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 = - ast::GetAttribute(var->attributes); - bool has_const_val = rhs && var->is_const && !override; - + bool has_const_val = rhs && var->is_const && !var->is_overridable; auto* global = builder_->create( 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) { - global->SetConstantId(static_cast(override->value)); + if (auto* id = ast::GetAttribute(var->attributes)) { + global->SetConstantId(static_cast(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(); - if (!var) { - continue; - } - auto* override_attr = - ast::GetAttribute(var->attributes); - if (!override_attr) { + if (!var || !var->is_overridable) { continue; } uint16_t constant_id; - if (override_attr->has_value) { - constant_id = static_cast(override_attr->value); + if (auto* id_attr = ast::GetAttribute(var->attributes)) { + constant_id = static_cast(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()) { + if (auto* id_attr = attr->As()) { // Track the constant IDs that are specified in the shader. - if (override_attr->has_value) { - constant_ids_.emplace(override_attr->value, sem); - } + constant_ids_.emplace(id_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. - if (ast::HasAttribute(decl->attributes)) { + // Capture the constant if it is pipeline-overridable. + 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()) { // Make sure the identifier is a non-overridable module-scope constant. - auto* var = ResolvedSymbol(ident); - if (!var || !var->Is() || - !var->Declaration()->is_const) { + auto* var = ResolvedSymbol(ident); + if (!var || !var->Declaration()->is_const) { AddError("array size identifier must be a module-scope constant", size_source); return nullptr; } - if (ast::HasAttribute( - var->Declaration()->attributes)) { + if (var->IsOverridable()) { AddError("array size expression must not be pipeline-overridable", size_source); return nullptr; diff --git a/src/resolver/resolver_test.cc b/src/resolver/resolver_test.cc index ba2add0ee2..f4fd4e9fb2 100644 --- a/src/resolver/resolver_test.cc +++ b/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; - // @override(1) let height = 8; - // @override(2) let depth = 2; + // @id(0) override width = 16; + // @id(1) override height = 8; + // @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* height = GlobalConst("height", ty.i32(), Expr(8), {Override(1)}); - auto* depth = GlobalConst("depth", ty.i32(), Expr(2), {Override(2)}); + auto* width = Override("width", ty.i32(), Expr(16), {Id(0)}); + auto* height = Override("height", ty.i32(), Expr(8), {Id(1)}); + 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; - // @override(1) let height : i32; - // @override(2) let depth : i32; + // @id(0) override width : i32; + // @id(1) override height : 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* height = GlobalConst("height", ty.i32(), nullptr, {Override(1)}); - auto* depth = GlobalConst("depth", ty.i32(), nullptr, {Override(2)}); + auto* width = Override("width", ty.i32(), nullptr, {Id(0)}); + auto* height = Override("height", ty.i32(), nullptr, {Id(1)}); + 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), diff --git a/src/resolver/resolver_validation.cc b/src/resolver/resolver_validation.cc index f9dadab8c5..a46c6fbb90 100644 --- a/src/resolver/resolver_validation.cc +++ b/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()) { - if (override_attr->has_value) { - uint32_t id = override_attr->value; - auto it = constant_ids_.find(id); - if (it != constant_ids_.end() && it->second != var) { - AddError("pipeline constant IDs must be unique", attr->source); - AddNote("a pipeline constant with an ID of " + std::to_string(id) + - " was previously declared " - "here:", - ast::GetAttribute( - it->second->Declaration()->attributes) - ->source); - return false; - } - if (id > 65535) { - AddError("pipeline constant IDs must be between 0 and 65535", - attr->source); - return false; - } + if (auto* id_attr = attr->As()) { + uint32_t id = id_attr->value; + auto it = constant_ids_.find(id); + if (it != constant_ids_.end() && it->second != var) { + AddError("pipeline constant IDs must be unique", attr->source); + AddNote("a pipeline constant with an ID of " + std::to_string(id) + + " was previously declared " + "here:", + ast::GetAttribute( + it->second->Declaration()->attributes) + ->source); + return false; + } + if (id > 65535) { + AddError("pipeline constant IDs must be between 0 and 65535", + attr->source); + return false; } } else { AddError("attribute is not valid for constants", attr->source); diff --git a/src/resolver/type_validation_test.cc b/src/resolver/type_validation_test.cc index a3d353600d..1c8f00b064 100644 --- a/src/resolver/type_validation_test.cc +++ b/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; - GlobalConst(Source{{12, 34}}, "a", ty.i32(), nullptr, - ast::AttributeList{create(0)}); + // @id(0) override a :i32; + Override(Source{{12, 34}}, "a", ty.i32(), nullptr, ast::AttributeList{Id(0)}); EXPECT_TRUE(r()->Resolve()) << r()->error(); } @@ -98,7 +97,7 @@ TEST_F(ResolverTypeValidationTest, GlobalConstantWithStorageClass_Fail) { AST().AddGlobalVariable(create( 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 a : array; - 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()); diff --git a/src/transform/binding_remapper.cc b/src/transform/binding_remapper.cc index 91c8ebb533..54fb554877 100644 --- a/src/transform/binding_remapper.cc +++ b/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( 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); } diff --git a/src/transform/calculate_array_length.cc b/src/transform/calculate_array_length.cc index b745bd23b2..87e4eb51dd 100644 --- a/src/transform/calculate_array_length.cc +++ b/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( 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(), diff --git a/src/transform/decompose_memory_access.cc b/src/transform/decompose_memory_access.cc index bc62b86695..647977d57a 100644 --- a/src/transform/decompose_memory_access.cc +++ b/src/transform/decompose_memory_access.cc @@ -457,7 +457,7 @@ struct DecomposeMemoryAccess::State { // array. b.create(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(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(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()), }; diff --git a/src/transform/external_texture_transform.cc b/src/transform/external_texture_transform.cc index 45323f0f2f..b597313c69 100644 --- a/src/transform/external_texture_transform.cc +++ b/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( 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); diff --git a/src/transform/renamer_test.cc b/src/transform/renamer_test.cc index 328605efee..0cdbdb43ec 100644 --- a/src/transform/renamer_test.cc +++ b/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", diff --git a/src/transform/robustness_test.cc b/src/transform/robustness_test.cc index ac422c0206..3b0747bdd7 100644 --- a/src/transform/robustness_test.cc +++ b/src/transform/robustness_test.cc @@ -1056,7 +1056,7 @@ var a : mat3x2; // 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 // 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 // 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 // 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 // var b : f32 = a[1][idx] // diff --git a/src/transform/single_entry_point.cc b/src/transform/single_entry_point.cc index 914585ab54..6e25c85b92 100644 --- a/src/transform/single_entry_point.cc +++ b/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()) { if (referenced_vars.count(var)) { - if (var->is_const) { - if (auto* attr = - ast::GetAttribute(var->attributes)) { - // It is an overridable constant - if (!attr->has_value) { - // If the attribute doesn't have numeric ID specified explicitly - // Make their ids explicitly assigned in the attribute so that - // they won't be affected by other stripped away constants - auto* global = sem.Get(var)->As(); - const auto* new_deco = - ctx.dst->Override(attr->source, global->ConstantId()); - ctx.Replace(attr, new_deco); - } + if (var->is_overridable) { + // It is an overridable constant + if (!ast::HasAttribute(var->attributes)) { + // If the constant doesn't already have an @id() attribute, add one + // so that its allocated ID so that it won't be affected by other + // stripped away constants + auto* global = sem.Get(var)->As(); + const auto* id = ctx.dst->Id(global->ConstantId()); + ctx.InsertFront(var->attributes, id); } } ctx.dst->AST().AddGlobalVariable(ctx.Clone(var)); diff --git a/src/transform/single_entry_point_test.cc b/src/transform/single_entry_point_test.cc index 6b950fed1c..9831ae110a 100644 --- a/src/transform/single_entry_point_test.cc +++ b/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; -[[override]] let c2 : u32 = 1u; -@override(0) let c3 : u32 = 1u; -@override(9999) let c4 : u32 = 1u; +@id(1001) override c1 : u32 = 1u; + override c2 : u32 = 1u; +@id(0) override c3 : 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() { diff --git a/src/transform/unshadow.cc b/src/transform/unshadow.cc index 4c9a795557..d71e087ee4 100644 --- a/src/transform/unshadow.cc +++ b/src/transform/unshadow.cc @@ -57,7 +57,7 @@ struct Unshadow::State { auto attributes = ctx.Clone(decl->attributes); return ctx.dst->create( 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* { diff --git a/src/transform/zero_init_workgroup_memory_test.cc b/src/transform/zero_init_workgroup_memory_test.cc index 54e1aa26ba..09baba6e92 100644 --- a/src/transform/zero_init_workgroup_memory_test.cc +++ b/src/transform/zero_init_workgroup_memory_test.cc @@ -506,7 +506,7 @@ var b : S; var c : array; -@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 b : S; var c : array; -@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 b : S; var c : array; -@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 b : S; var c : array; -@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) { diff --git a/src/writer/glsl/generator_impl.cc b/src/writer/glsl/generator_impl.cc index f31a4a8349..2331565160 100644 --- a/src/writer/glsl/generator_impl.cc +++ b/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()) { + if (!d->Is()) { diagnostics_.add_error(diag::System::Writer, "Decorated const values not valid"); return false; diff --git a/src/writer/glsl/generator_impl_function_test.cc b/src/writer/glsl/generator_impl_function_test.cc index 09713a1ff4..0b539ebbe9 100644 --- a/src/writer/glsl/generator_impl_function_test.cc +++ b/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)}); - GlobalConst("height", ty.i32(), Construct(ty.i32(), 3), {Override(8u)}); - GlobalConst("depth", ty.i32(), Construct(ty.i32(), 4), {Override(9u)}); + Override("width", ty.i32(), Construct(ty.i32(), 2), {Id(7u)}); + Override("height", ty.i32(), Construct(ty.i32(), 3), {Id(8u)}); + Override("depth", ty.i32(), Construct(ty.i32(), 4), {Id(9u)}); Func("main", ast::VariableList{}, ty.void_(), {}, { Stage(ast::PipelineStage::kCompute), diff --git a/src/writer/glsl/generator_impl_module_constant_test.cc b/src/writer/glsl/generator_impl_module_constant_test.cc index 0f1fb78224..036f1e58d6 100644 --- a/src/writer/glsl/generator_impl_module_constant_test.cc +++ b/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), - ast::AttributeList{ - create(23), - }); + auto* var = Override("pos", ty.f32(), Expr(3.0f), + ast::AttributeList{ + 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, - ast::AttributeList{ - create(23), - }); + auto* var = Override("pos", ty.f32(), nullptr, + ast::AttributeList{ + 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), - ast::AttributeList{ - create(0), - }); - auto* b = GlobalConst("b", ty.f32(), Expr(2.0f), - ast::AttributeList{ - create(), - }); + auto* a = Override("a", ty.f32(), Expr(3.0f), + ast::AttributeList{ + Id(0), + }); + auto* b = Override("b", ty.f32(), Expr(2.0f)); GeneratorImpl& gen = Build(); diff --git a/src/writer/glsl/generator_impl_workgroup_var_test.cc b/src/writer/glsl/generator_impl_workgroup_var_test.cc index fbc36fb2f5..41ee6fa82f 100644 --- a/src/writer/glsl/generator_impl_workgroup_var_test.cc +++ b/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" diff --git a/src/writer/hlsl/generator_impl.cc b/src/writer/hlsl/generator_impl.cc index e8caac3777..ee275b50f4 100644 --- a/src/writer/hlsl/generator_impl.cc +++ b/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()) { + if (!d->Is()) { diagnostics_.add_error(diag::System::Writer, "Decorated const values not valid"); return false; diff --git a/src/writer/hlsl/generator_impl_function_test.cc b/src/writer/hlsl/generator_impl_function_test.cc index d5e0b92847..d17620ce3b 100644 --- a/src/writer/hlsl/generator_impl_function_test.cc +++ b/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)}); - GlobalConst("height", ty.i32(), Construct(ty.i32(), 3), {Override(8u)}); - GlobalConst("depth", ty.i32(), Construct(ty.i32(), 4), {Override(9u)}); + Override("width", ty.i32(), Construct(ty.i32(), 2), {Id(7u)}); + Override("height", ty.i32(), Construct(ty.i32(), 3), {Id(8u)}); + Override("depth", ty.i32(), Construct(ty.i32(), 4), {Id(9u)}); Func("main", ast::VariableList{}, ty.void_(), {}, { Stage(ast::PipelineStage::kCompute), diff --git a/src/writer/hlsl/generator_impl_module_constant_test.cc b/src/writer/hlsl/generator_impl_module_constant_test.cc index a5698b0583..7502288232 100644 --- a/src/writer/hlsl/generator_impl_module_constant_test.cc +++ b/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), - ast::AttributeList{ - create(23), - }); + auto* var = Override("pos", ty.f32(), Expr(3.0f), + ast::AttributeList{ + 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, - ast::AttributeList{ - create(23), - }); + auto* var = Override("pos", ty.f32(), nullptr, + ast::AttributeList{ + 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), - ast::AttributeList{ - create(0), - }); - auto* b = GlobalConst("b", ty.f32(), Expr(2.0f), - ast::AttributeList{ - create(), - }); + auto* a = Override("a", ty.f32(), Expr(3.0f), + ast::AttributeList{ + Id(0), + }); + auto* b = Override("b", ty.f32(), Expr(2.0f)); GeneratorImpl& gen = Build(); diff --git a/src/writer/hlsl/generator_impl_workgroup_var_test.cc b/src/writer/hlsl/generator_impl_workgroup_var_test.cc index 5e6c669139..4da3a8123b 100644 --- a/src/writer/hlsl/generator_impl_workgroup_var_test.cc +++ b/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" diff --git a/src/writer/msl/generator_impl.cc b/src/writer/msl/generator_impl.cc index adace185a8..2930fcfd5d 100644 --- a/src/writer/msl/generator_impl.cc +++ b/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()) { + if (!d->Is()) { diagnostics_.add_error(diag::System::Writer, "Decorated const values not valid"); return false; diff --git a/src/writer/msl/generator_impl_module_constant_test.cc b/src/writer/msl/generator_impl_module_constant_test.cc index c2aa9c9cf2..0d2a7eb057 100644 --- a/src/writer/msl/generator_impl_module_constant_test.cc +++ b/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), - ast::AttributeList{ - create(23), - }); + auto* var = Override("pos", ty.f32(), Expr(3.f), + ast::AttributeList{ + 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, - ast::AttributeList{ - create(0), - }); - auto* var_b = GlobalConst("b", ty.f32(), nullptr, - ast::AttributeList{ - create(), - }); + auto* var_a = Override("a", ty.f32(), nullptr, + ast::AttributeList{ + Id(0), + }); + auto* var_b = Override("b", ty.f32(), nullptr); GeneratorImpl& gen = Build(); diff --git a/src/writer/spirv/builder.cc b/src/writer/spirv/builder.cc index c3012f47ce..84bd6331d9 100644 --- a/src/writer/spirv/builder.cc +++ b/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 - // attribute. - if (!ast::HasAttribute(var->attributes)) { + // Constants must have an initializer unless they are overridable. + if (!var->is_overridable) { 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*) { diff --git a/src/writer/spirv/builder_function_attribute_test.cc b/src/writer/spirv/builder_function_attribute_test.cc index 7500895f1b..a5f8a5ca24 100644 --- a/src/writer/spirv/builder_function_attribute_test.cc +++ b/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)}); - GlobalConst("height", ty.i32(), Construct(ty.i32(), 3), {Override(8u)}); - GlobalConst("depth", ty.i32(), Construct(ty.i32(), 4), {Override(9u)}); + Override("width", ty.i32(), Construct(ty.i32(), 2), {Id(7u)}); + Override("height", ty.i32(), Construct(ty.i32(), 3), {Id(8u)}); + 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{ diff --git a/src/writer/spirv/builder_global_variable_test.cc b/src/writer/spirv/builder_global_variable_test.cc index 46004353f9..9181fa74d5 100644 --- a/src/writer/spirv/builder_global_variable_test.cc +++ b/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), - ast::AttributeList{ - create(1200), - }); + auto* v = Override("var", ty.bool_(), Expr(true), + ast::AttributeList{ + 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(), - ast::AttributeList{ - create(1200), - }); + auto* v = Override("var", ty.bool_(), Construct(), + ast::AttributeList{ + 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, - ast::AttributeList{ - create(1200), - }); + auto* v = Override("var", ty.bool_(), nullptr, + ast::AttributeList{ + 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), - ast::AttributeList{ - create(0), - }); + auto* v = Override("var", ty.f32(), Expr(2.f), + ast::AttributeList{ + 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(), - ast::AttributeList{ - create(0), - }); + auto* v = Override("var", ty.f32(), Construct(), + ast::AttributeList{ + 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, - ast::AttributeList{ - create(0), - }); + auto* v = Override("var", ty.f32(), nullptr, + ast::AttributeList{ + 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, - ast::AttributeList{ - create(0), - }); + auto* v = Override("var", ty.i32(), nullptr, + ast::AttributeList{ + 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, - ast::AttributeList{ - create(0), - }); + auto* v = Override("var", ty.u32(), nullptr, + ast::AttributeList{ + 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), - ast::AttributeList{ - create(0), - }); - auto* var_b = GlobalConst("b", ty.bool_(), Expr(false), - ast::AttributeList{ - create(), - }); + auto* var_a = Override("a", ty.bool_(), Expr(true), + ast::AttributeList{ + Id(0), + }); + auto* var_b = Override("b", ty.bool_(), Expr(false)); spirv::Builder& b = Build(); diff --git a/src/writer/wgsl/generator_impl.cc b/src/writer/wgsl/generator_impl.cc index 96a459ab96..89268393bd 100644 --- a/src/writer/wgsl/generator_impl.cc +++ b/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) { - out << "override"; - if (override_deco->has_value) { - out << "(" << override_deco->value << ")"; - } + [&](const ast::IdAttribute* override_deco) { + out << "id(" << override_deco->value << ")"; return true; }, [&](const ast::StructMemberSizeAttribute* size) { diff --git a/src/writer/wgsl/generator_impl_global_decl_test.cc b/src/writer/wgsl/generator_impl_global_decl_test.cc index 1f5d048a11..87493c6301 100644 --- a/src/writer/wgsl/generator_impl_global_decl_test.cc +++ b/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()}); - GlobalConst("b", ty.f32(), nullptr, {Override(7u)}); + Override("a", ty.f32(), nullptr); + 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; )"); } diff --git a/src/writer/wgsl/generator_impl_literal_test.cc b/src/writer/wgsl/generator_impl_literal_test.cc index 9fe68b8b5d..55fc6783ef 100644 --- a/src/writer/wgsl/generator_impl_literal_test.cc +++ b/src/writer/wgsl/generator_impl_literal_test.cc @@ -14,7 +14,6 @@ #include -#include "src/ast/override_attribute.h" #include "src/writer/wgsl/test_helper.h" namespace tint { diff --git a/src/writer/wgsl/generator_impl_variable_test.cc b/src/writer/wgsl/generator_impl_variable_test.cc index 427b4aa617..fc1654c853 100644 --- a/src/writer/wgsl/generator_impl_variable_test.cc +++ b/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" diff --git a/test/BUILD.gn b/test/BUILD.gn index 76c24b7e2a..a1a99f6966 100644 --- a/test/BUILD.gn +++ b/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", diff --git a/test/var/override/named/no_init/bool.wgsl b/test/var/override/named/no_init/bool.wgsl index fe69ac3f42..04ca499584 100644 --- a/test/var/override/named/no_init/bool.wgsl +++ b/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() { diff --git a/test/var/override/named/no_init/bool.wgsl.expected.wgsl b/test/var/override/named/no_init/bool.wgsl.expected.wgsl index c4b5efd192..344ec0e312 100644 --- a/test/var/override/named/no_init/bool.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/no_init/f32.wgsl b/test/var/override/named/no_init/f32.wgsl index 5328e9dcce..50dee2f087 100644 --- a/test/var/override/named/no_init/f32.wgsl +++ b/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() { diff --git a/test/var/override/named/no_init/f32.wgsl.expected.wgsl b/test/var/override/named/no_init/f32.wgsl.expected.wgsl index 1d0945293f..bf38cdd930 100644 --- a/test/var/override/named/no_init/f32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/no_init/i32.wgsl b/test/var/override/named/no_init/i32.wgsl index 0a3ed57909..d91c2f2381 100644 --- a/test/var/override/named/no_init/i32.wgsl +++ b/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() { diff --git a/test/var/override/named/no_init/i32.wgsl.expected.wgsl b/test/var/override/named/no_init/i32.wgsl.expected.wgsl index 92a7426f43..95d395c9e9 100644 --- a/test/var/override/named/no_init/i32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/no_init/u32.wgsl b/test/var/override/named/no_init/u32.wgsl index bce0c7769c..5c11bc962a 100644 --- a/test/var/override/named/no_init/u32.wgsl +++ b/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() { diff --git a/test/var/override/named/no_init/u32.wgsl.expected.wgsl b/test/var/override/named/no_init/u32.wgsl.expected.wgsl index 31db0667b8..8d79972804 100644 --- a/test/var/override/named/no_init/u32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/val_init/bool.wgsl b/test/var/override/named/val_init/bool.wgsl index 9edeffaca1..8eaca90b85 100644 --- a/test/var/override/named/val_init/bool.wgsl +++ b/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() { diff --git a/test/var/override/named/val_init/bool.wgsl.expected.wgsl b/test/var/override/named/val_init/bool.wgsl.expected.wgsl index cbcd033c0b..02f5b7d2d9 100644 --- a/test/var/override/named/val_init/bool.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/val_init/f32.wgsl b/test/var/override/named/val_init/f32.wgsl index a58f753f65..0f718e87a3 100644 --- a/test/var/override/named/val_init/f32.wgsl +++ b/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() { diff --git a/test/var/override/named/val_init/f32.wgsl.expected.wgsl b/test/var/override/named/val_init/f32.wgsl.expected.wgsl index 18e6ceb1ce..1c53942f13 100644 --- a/test/var/override/named/val_init/f32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/val_init/i32.wgsl b/test/var/override/named/val_init/i32.wgsl index 642096dcdb..9f4b021ed1 100644 --- a/test/var/override/named/val_init/i32.wgsl +++ b/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() { diff --git a/test/var/override/named/val_init/i32.wgsl.expected.wgsl b/test/var/override/named/val_init/i32.wgsl.expected.wgsl index 46adcb78ac..f36ff37a32 100644 --- a/test/var/override/named/val_init/i32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/val_init/u32.wgsl b/test/var/override/named/val_init/u32.wgsl index 582f7abceb..77ca84c4ad 100644 --- a/test/var/override/named/val_init/u32.wgsl +++ b/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() { diff --git a/test/var/override/named/val_init/u32.wgsl.expected.wgsl b/test/var/override/named/val_init/u32.wgsl.expected.wgsl index 06b1aa0237..4a56f44644 100644 --- a/test/var/override/named/val_init/u32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/zero_init/bool.wgsl b/test/var/override/named/zero_init/bool.wgsl index 8869f16b58..1dff97f0d1 100644 --- a/test/var/override/named/zero_init/bool.wgsl +++ b/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() { diff --git a/test/var/override/named/zero_init/bool.wgsl.expected.wgsl b/test/var/override/named/zero_init/bool.wgsl.expected.wgsl index f68661234d..cfd341ee0e 100644 --- a/test/var/override/named/zero_init/bool.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/zero_init/f32.wgsl b/test/var/override/named/zero_init/f32.wgsl index 4af2466523..c4f949ab58 100644 --- a/test/var/override/named/zero_init/f32.wgsl +++ b/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() { diff --git a/test/var/override/named/zero_init/f32.wgsl.expected.wgsl b/test/var/override/named/zero_init/f32.wgsl.expected.wgsl index 9398382b0f..159b6a90c5 100644 --- a/test/var/override/named/zero_init/f32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/zero_init/i32.wgsl b/test/var/override/named/zero_init/i32.wgsl index 132987a493..2b44da76ea 100644 --- a/test/var/override/named/zero_init/i32.wgsl +++ b/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() { diff --git a/test/var/override/named/zero_init/i32.wgsl.expected.wgsl b/test/var/override/named/zero_init/i32.wgsl.expected.wgsl index afa82db937..e02284c769 100644 --- a/test/var/override/named/zero_init/i32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/named/zero_init/u32.wgsl b/test/var/override/named/zero_init/u32.wgsl index 047aeb57b9..de11d5a9c0 100644 --- a/test/var/override/named/zero_init/u32.wgsl +++ b/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() { diff --git a/test/var/override/named/zero_init/u32.wgsl.expected.wgsl b/test/var/override/named/zero_init/u32.wgsl.expected.wgsl index 9146204c69..62250f44b1 100644 --- a/test/var/override/named/zero_init/u32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/numbered/no_init/bool.wgsl b/test/var/override/numbered/no_init/bool.wgsl index 407941d1c3..98ad2ecc1b 100644 --- a/test/var/override/numbered/no_init/bool.wgsl +++ b/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() { diff --git a/test/var/override/numbered/no_init/bool.wgsl.expected.wgsl b/test/var/override/numbered/no_init/bool.wgsl.expected.wgsl index d51cd82ab3..d2077c0006 100644 --- a/test/var/override/numbered/no_init/bool.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/numbered/no_init/f32.wgsl b/test/var/override/numbered/no_init/f32.wgsl index 796c8add42..d15b4ef38e 100644 --- a/test/var/override/numbered/no_init/f32.wgsl +++ b/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() { diff --git a/test/var/override/numbered/no_init/f32.wgsl.expected.wgsl b/test/var/override/numbered/no_init/f32.wgsl.expected.wgsl index 2315774fdf..1d3c7ca79f 100644 --- a/test/var/override/numbered/no_init/f32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/numbered/no_init/i32.wgsl b/test/var/override/numbered/no_init/i32.wgsl index 5778c734e2..dca18e22a3 100644 --- a/test/var/override/numbered/no_init/i32.wgsl +++ b/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() { diff --git a/test/var/override/numbered/no_init/i32.wgsl.expected.wgsl b/test/var/override/numbered/no_init/i32.wgsl.expected.wgsl index cdd1bdda26..89e30497eb 100644 --- a/test/var/override/numbered/no_init/i32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/numbered/no_init/u32.wgsl b/test/var/override/numbered/no_init/u32.wgsl index 7992da9989..5cdb5982d9 100644 --- a/test/var/override/numbered/no_init/u32.wgsl +++ b/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() { diff --git a/test/var/override/numbered/no_init/u32.wgsl.expected.wgsl b/test/var/override/numbered/no_init/u32.wgsl.expected.wgsl index 2a2fc965f5..d330537a46 100644 --- a/test/var/override/numbered/no_init/u32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/numbered/val_init/bool.wgsl b/test/var/override/numbered/val_init/bool.wgsl index 91cbe009b1..815e5102d1 100644 --- a/test/var/override/numbered/val_init/bool.wgsl +++ b/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() { diff --git a/test/var/override/numbered/val_init/bool.wgsl.expected.wgsl b/test/var/override/numbered/val_init/bool.wgsl.expected.wgsl index 57c6d05a2e..623733603a 100644 --- a/test/var/override/numbered/val_init/bool.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/numbered/val_init/f32.wgsl b/test/var/override/numbered/val_init/f32.wgsl index 72519868ea..c565064ee9 100644 --- a/test/var/override/numbered/val_init/f32.wgsl +++ b/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() { diff --git a/test/var/override/numbered/val_init/f32.wgsl.expected.wgsl b/test/var/override/numbered/val_init/f32.wgsl.expected.wgsl index d2a4882caf..8599cbae21 100644 --- a/test/var/override/numbered/val_init/f32.wgsl.expected.wgsl +++ b/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() { diff --git a/test/var/override/numbered/val_init/i32.wgsl b/test/var/override/numbered/val_init/i32.wgsl index 1036ba5578..78cb191005 100644 --- a/test/var/override/numbered/val_init/i32.wgsl +++ b/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() { diff --git a/test/var/override/numbered/val_init/i32.wgsl.expected.wgsl b/test/var/override/numbered/val_init/i32.wgsl.expected.wgsl index 8f8e5f1b48..062ca36492 100644 --- a/test/var/override/numbered/val_init/i32.wgsl.expected.wgsl +++ b/test/var/override/numbered/val_init/i32.wgsl.expected.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : i32 = 1; +@id(1234) override o : i32 = 1; @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/val_init/u32.wgsl b/test/var/override/numbered/val_init/u32.wgsl index 3ee87d39ba..bbc740930e 100644 --- a/test/var/override/numbered/val_init/u32.wgsl +++ b/test/var/override/numbered/val_init/u32.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : u32 = 1u; +@id(1234) override o : u32 = 1u; @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/val_init/u32.wgsl.expected.wgsl b/test/var/override/numbered/val_init/u32.wgsl.expected.wgsl index 4712fda610..8a65ad64a0 100644 --- a/test/var/override/numbered/val_init/u32.wgsl.expected.wgsl +++ b/test/var/override/numbered/val_init/u32.wgsl.expected.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : u32 = 1u; +@id(1234) override o : u32 = 1u; @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/zero_init/bool.wgsl b/test/var/override/numbered/zero_init/bool.wgsl index 27a68b3620..b411e9cad7 100644 --- a/test/var/override/numbered/zero_init/bool.wgsl +++ b/test/var/override/numbered/zero_init/bool.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : bool = bool(); +@id(1234) override o : bool = bool(); @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/zero_init/bool.wgsl.expected.wgsl b/test/var/override/numbered/zero_init/bool.wgsl.expected.wgsl index 2e68a0e38a..41ae914fba 100644 --- a/test/var/override/numbered/zero_init/bool.wgsl.expected.wgsl +++ b/test/var/override/numbered/zero_init/bool.wgsl.expected.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : bool = bool(); +@id(1234) override o : bool = bool(); @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/zero_init/f32.wgsl b/test/var/override/numbered/zero_init/f32.wgsl index a022f0b592..92dffa4ce0 100644 --- a/test/var/override/numbered/zero_init/f32.wgsl +++ b/test/var/override/numbered/zero_init/f32.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : f32 = f32(); +@id(1234) override o : f32 = f32(); @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/zero_init/f32.wgsl.expected.wgsl b/test/var/override/numbered/zero_init/f32.wgsl.expected.wgsl index e5d2086a72..caaafb029a 100644 --- a/test/var/override/numbered/zero_init/f32.wgsl.expected.wgsl +++ b/test/var/override/numbered/zero_init/f32.wgsl.expected.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : f32 = f32(); +@id(1234) override o : f32 = f32(); @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/zero_init/i32.wgsl b/test/var/override/numbered/zero_init/i32.wgsl index 448f14adae..79f3f2ca26 100644 --- a/test/var/override/numbered/zero_init/i32.wgsl +++ b/test/var/override/numbered/zero_init/i32.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : i32 = i32(); +@id(1234) override o : i32 = i32(); @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/zero_init/i32.wgsl.expected.wgsl b/test/var/override/numbered/zero_init/i32.wgsl.expected.wgsl index c250241d66..f10ef51793 100644 --- a/test/var/override/numbered/zero_init/i32.wgsl.expected.wgsl +++ b/test/var/override/numbered/zero_init/i32.wgsl.expected.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : i32 = i32(); +@id(1234) override o : i32 = i32(); @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/zero_init/u32.wgsl b/test/var/override/numbered/zero_init/u32.wgsl index ff50cba980..a0e7de6b54 100644 --- a/test/var/override/numbered/zero_init/u32.wgsl +++ b/test/var/override/numbered/zero_init/u32.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : u32 = u32(); +@id(1234) override o : u32 = u32(); @stage(compute) @workgroup_size(1) fn main() { diff --git a/test/var/override/numbered/zero_init/u32.wgsl.expected.wgsl b/test/var/override/numbered/zero_init/u32.wgsl.expected.wgsl index 1dc6121073..710e3ebea7 100644 --- a/test/var/override/numbered/zero_init/u32.wgsl.expected.wgsl +++ b/test/var/override/numbered/zero_init/u32.wgsl.expected.wgsl @@ -1,4 +1,4 @@ -@override(1234) let o : u32 = u32(); +@id(1234) override o : u32 = u32(); @stage(compute) @workgroup_size(1) fn main() {