Add Partial Tint Dual Source Blending Extension Adds dual source blending extension string to Tint. Adds @index attribute to ast and resolver. Includes basic ast and resolver tests. Bug: dawn:1709 Change-Id: I6bf05b42978c3338e1f5f45d21bd2a3fb5fed08d Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/137281 Reviewed-by: Ben Clayton <bclayton@google.com> Kokoro: Kokoro <noreply+kokoro@google.com> Commit-Queue: Ben Clayton <bclayton@google.com>
diff --git a/src/tint/BUILD.gn b/src/tint/BUILD.gn index efa4753..1d71177 100644 --- a/src/tint/BUILD.gn +++ b/src/tint/BUILD.gn
@@ -577,6 +577,7 @@ "ast/if_statement.h", "ast/increment_decrement_statement.h", "ast/index_accessor_expression.h", + "ast/index_attribute.h", "ast/int_literal_expression.h", "ast/internal_attribute.h", "ast/interpolate_attribute.h", @@ -666,6 +667,7 @@ "ast/if_statement.cc", "ast/increment_decrement_statement.cc", "ast/index_accessor_expression.cc", + "ast/index_attribute.cc", "ast/int_literal_expression.cc", "ast/internal_attribute.cc", "ast/interpolate_attribute.cc", @@ -1564,6 +1566,7 @@ "ast/if_statement_test.cc", "ast/increment_decrement_statement_test.cc", "ast/index_accessor_expression_test.cc", + "ast/index_attribute_test.cc", "ast/int_literal_expression_test.cc", "ast/interpolate_attribute_test.cc", "ast/location_attribute_test.cc", @@ -1673,6 +1676,7 @@ "resolver/control_block_validation_test.cc", "resolver/dependency_graph_test.cc", "resolver/diagnostic_control_test.cc", + "resolver/dual_source_blending_extension_test.cc", "resolver/entry_point_validation_test.cc", "resolver/evaluation_stage_test.cc", "resolver/expression_kind_test.cc",
diff --git a/src/tint/CMakeLists.txt b/src/tint/CMakeLists.txt index 3ac8ed4..5c0bf6f 100644 --- a/src/tint/CMakeLists.txt +++ b/src/tint/CMakeLists.txt
@@ -158,6 +158,8 @@ ast/increment_decrement_statement.h ast/index_accessor_expression.cc ast/index_accessor_expression.h + ast/index_attribute.cc + ast/index_attribute.h ast/int_literal_expression.cc ast/int_literal_expression.h ast/internal_attribute.cc @@ -939,6 +941,7 @@ ast/if_statement_test.cc ast/increment_decrement_statement_test.cc ast/index_accessor_expression_test.cc + ast/index_attribute_test.cc ast/int_literal_expression_test.cc ast/interpolate_attribute_test.cc ast/location_attribute_test.cc @@ -1008,6 +1011,7 @@ resolver/control_block_validation_test.cc resolver/dependency_graph_test.cc resolver/diagnostic_control_test.cc + resolver/dual_source_blending_extension_test.cc resolver/entry_point_validation_test.cc resolver/evaluation_stage_test.cc resolver/expression_kind_test.cc
diff --git a/src/tint/ast/index_attribute.cc b/src/tint/ast/index_attribute.cc new file mode 100644 index 0000000..7916226 --- /dev/null +++ b/src/tint/ast/index_attribute.cc
@@ -0,0 +1,41 @@ +// Copyright 2023 The Tint Authors. +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. + +#include "src/tint/ast/index_attribute.h" + +#include <string> + +#include "src/tint/program_builder.h" + +TINT_INSTANTIATE_TYPEINFO(tint::ast::IndexAttribute); + +namespace tint::ast { + +IndexAttribute::IndexAttribute(ProgramID pid, NodeID nid, const Source& src, const Expression* exp) + : Base(pid, nid, src), expr(exp) {} + +IndexAttribute::~IndexAttribute() = default; + +std::string IndexAttribute::Name() const { + return "index"; +} + +const IndexAttribute* IndexAttribute::Clone(CloneContext* ctx) const { + // Clone arguments outside of create() call to have deterministic ordering + auto src = ctx->Clone(source); + auto* expr_ = ctx->Clone(expr); + return ctx->dst->create<IndexAttribute>(src, expr_); +} + +} // namespace tint::ast
diff --git a/src/tint/ast/index_attribute.h b/src/tint/ast/index_attribute.h new file mode 100644 index 0000000..e241ba6 --- /dev/null +++ b/src/tint/ast/index_attribute.h
@@ -0,0 +1,51 @@ +// Copyright 2023 The Tint Authors. +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. + +#ifndef SRC_TINT_AST_INDEX_ATTRIBUTE_H_ +#define SRC_TINT_AST_INDEX_ATTRIBUTE_H_ + +#include <string> + +#include "src/tint/ast/attribute.h" +#include "src/tint/ast/expression.h" + +namespace tint::ast { + +/// An id attribute for pipeline-overridable constants +class IndexAttribute final : public utils::Castable<IndexAttribute, Attribute> { + public: + /// Create an index attribute. + /// @param pid the identifier of the program that owns this node + /// @param nid the unique node identifier + /// @param src the source of this node + /// @param expr the numeric id expression + IndexAttribute(ProgramID pid, NodeID nid, const Source& src, const Expression* expr); + ~IndexAttribute() override; + + /// @returns the WGSL name for the attribute + std::string Name() const override; + + /// Clones this node and all transitive child nodes using the `CloneContext` + /// `ctx`. + /// @param ctx the clone context + /// @return the newly cloned node + const IndexAttribute* Clone(CloneContext* ctx) const override; + + /// The id expression + const Expression* const expr; +}; + +} // namespace tint::ast + +#endif // SRC_TINT_AST_INDEX_ATTRIBUTE_H_
diff --git a/src/tint/ast/index_attribute_test.cc b/src/tint/ast/index_attribute_test.cc new file mode 100644 index 0000000..bad37c6 --- /dev/null +++ b/src/tint/ast/index_attribute_test.cc
@@ -0,0 +1,31 @@ +// Copyright 2023 The Tint Authors. +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. + +#include "src/tint/ast/id_attribute.h" + +#include "src/tint/ast/test_helper.h" + +namespace tint::ast { +namespace { + +using namespace tint::number_suffixes; // NOLINT +using IndexAttributeTest = TestHelper; + +TEST_F(IndexAttributeTest, Creation) { + auto* d = Index(1_a); + EXPECT_TRUE(d->expr->Is<IntLiteralExpression>()); +} + +} // namespace +} // namespace tint::ast
diff --git a/src/tint/builtin/attribute.cc b/src/tint/builtin/attribute.cc index 33a5171..0647825 100644 --- a/src/tint/builtin/attribute.cc +++ b/src/tint/builtin/attribute.cc
@@ -52,6 +52,9 @@ if (str == "id") { return Attribute::kId; } + if (str == "index") { + return Attribute::kIndex; + } if (str == "interpolate") { return Attribute::kInterpolate; } @@ -96,6 +99,8 @@ return out << "group"; case Attribute::kId: return out << "id"; + case Attribute::kIndex: + return out << "index"; case Attribute::kInterpolate: return out << "interpolate"; case Attribute::kInvariant:
diff --git a/src/tint/builtin/attribute.h b/src/tint/builtin/attribute.h index c975267..f9a1868 100644 --- a/src/tint/builtin/attribute.h +++ b/src/tint/builtin/attribute.h
@@ -41,6 +41,7 @@ kFragment, kGroup, kId, + kIndex, kInterpolate, kInvariant, kLocation, @@ -61,9 +62,9 @@ Attribute ParseAttribute(std::string_view str); constexpr const char* kAttributeStrings[] = { - "align", "binding", "builtin", "compute", "diagnostic", - "fragment", "group", "id", "interpolate", "invariant", - "location", "must_use", "size", "vertex", "workgroup_size", + "align", "binding", "builtin", "compute", "diagnostic", "fragment", + "group", "id", "index", "interpolate", "invariant", "location", + "must_use", "size", "vertex", "workgroup_size", }; } // namespace tint::builtin
diff --git a/src/tint/builtin/attribute_bench.cc b/src/tint/builtin/attribute_bench.cc index a25fbf1..267a905 100644 --- a/src/tint/builtin/attribute_bench.cc +++ b/src/tint/builtin/attribute_bench.cc
@@ -87,55 +87,62 @@ "i", "rrd", "iG", - "FFnterpolate", - "iEtrplat", - "inerporrate", + "inFFex", + "iE", + "inrrx", + "index", + "inx", + "inJJD", + "ie", + "inerpklae", + "intrpolate", + "inJerpolae", "interpolate", - "inteplate", - "XterJJoDate", - "inepol8t", - "nvark1n", - "invriant", - "Jnvarant", + "interpocate", + "interpolaOe", + "__nttevvpoKKate", + "xnvari5n8", + "inFq__ant", + "iqqariant", "invariant", - "invaricnt", - "invariaOt", - "invttKK_ianvv", - "lxxcati8", - "Focqq__o", - "locaiqqn", + "invar6a33O", + "i96arQttanoo", + "inari66nt", + "lOxati6zz", + "locyytion", + "lHHtion", "location", - "loc33tio6", - "ltto6at9QQn", - "loc66tio", - "mOxt_u6zz", - "musyy_use", - "mHH_use", + "qWW4caton", + "locOOton", + "ocatiYn", + "m_use", + "mutFuse", + "wust_us", "must_use", - "qWW4st_se", - "musOO_se", - "ust_uYe", - "i", - "Fie", - "siw", + "Kst_sff", + "qusKK_use", + "mFsmm_3se", + "ize", + "sze", + "sbbb", "size", - "zff", - "sizqK", - "s3zmm", - "ertex", - "vereq", - "vbtbbx", + "iie", + "siqe", + "svvTTe", + "vertFFx", + "vrQ00P", + "vePtex", "vertex", - "irtex", - "vOOteq", - "vertTvvx", - "woFFkgroup_size", - "wfr00grPupsiQe", - "workgrouP_size", + "vsste77", + "veCtRRbb", + "verteXX", + "workgqou_siCCOOO", + "worsgroupsuzL", + "wXrkgroup_size", "workgroup_size", - "workgroup77sise", - "RRobbkgroupCsize", - "wXXrkgroup_size", + "workgroup_sze", + "wqqrOgoupize", + "workg22oup_size", }; for (auto _ : state) { for (auto* str : kStrings) {
diff --git a/src/tint/builtin/attribute_test.cc b/src/tint/builtin/attribute_test.cc index 9919467..33f557e 100644 --- a/src/tint/builtin/attribute_test.cc +++ b/src/tint/builtin/attribute_test.cc
@@ -43,21 +43,14 @@ } static constexpr Case kValidCases[] = { - {"align", Attribute::kAlign}, - {"binding", Attribute::kBinding}, - {"builtin", Attribute::kBuiltin}, - {"compute", Attribute::kCompute}, - {"diagnostic", Attribute::kDiagnostic}, - {"fragment", Attribute::kFragment}, - {"group", Attribute::kGroup}, - {"id", Attribute::kId}, - {"interpolate", Attribute::kInterpolate}, - {"invariant", Attribute::kInvariant}, - {"location", Attribute::kLocation}, - {"must_use", Attribute::kMustUse}, - {"size", Attribute::kSize}, - {"vertex", Attribute::kVertex}, - {"workgroup_size", Attribute::kWorkgroupSize}, + {"align", Attribute::kAlign}, {"binding", Attribute::kBinding}, + {"builtin", Attribute::kBuiltin}, {"compute", Attribute::kCompute}, + {"diagnostic", Attribute::kDiagnostic}, {"fragment", Attribute::kFragment}, + {"group", Attribute::kGroup}, {"id", Attribute::kId}, + {"index", Attribute::kIndex}, {"interpolate", Attribute::kInterpolate}, + {"invariant", Attribute::kInvariant}, {"location", Attribute::kLocation}, + {"must_use", Attribute::kMustUse}, {"size", Attribute::kSize}, + {"vertex", Attribute::kVertex}, {"workgroup_size", Attribute::kWorkgroupSize}, }; static constexpr Case kInvalidCases[] = { @@ -85,27 +78,30 @@ {"d", Attribute::kUndefined}, {"i", Attribute::kUndefined}, {"OVd", Attribute::kUndefined}, - {"inyerpolae", Attribute::kUndefined}, - {"rrnterpolll77Ge", Attribute::kUndefined}, - {"inte4pol00te", Attribute::kUndefined}, - {"inoornt", Attribute::kUndefined}, - {"inzzriat", Attribute::kUndefined}, - {"n11pariiin", Attribute::kUndefined}, - {"XXocation", Attribute::kUndefined}, - {"lIIc9955nnon", Attribute::kUndefined}, - {"aaoHHatioYSS", Attribute::kUndefined}, - {"mkksue", Attribute::kUndefined}, - {"gjs_RRs", Attribute::kUndefined}, - {"msb_se", Attribute::kUndefined}, - {"jize", Attribute::kUndefined}, - {"sze", Attribute::kUndefined}, - {"qz", Attribute::kUndefined}, - {"vNNtex", Attribute::kUndefined}, - {"vevvx", Attribute::kUndefined}, - {"veQQex", Attribute::kUndefined}, - {"workgrrupffie", Attribute::kUndefined}, - {"workgroup_sije", Attribute::kUndefined}, - {"workgoupNNwsiz8", Attribute::kUndefined}, + {"ndyx", Attribute::kUndefined}, + {"n77rrldGx", Attribute::kUndefined}, + {"inde40", Attribute::kUndefined}, + {"itooolate", Attribute::kUndefined}, + {"intezplate", Attribute::kUndefined}, + {"ppnerii1olat", Attribute::kUndefined}, + {"invarianXX", Attribute::kUndefined}, + {"inv55ria99nII", Attribute::kUndefined}, + {"irrvariaSSaHH", Attribute::kUndefined}, + {"lkkcin", Attribute::kUndefined}, + {"gjctRRo", Attribute::kUndefined}, + {"lcbton", Attribute::kUndefined}, + {"mustjuse", Attribute::kUndefined}, + {"must_se", Attribute::kUndefined}, + {"muquse", Attribute::kUndefined}, + {"szNN", Attribute::kUndefined}, + {"zvv", Attribute::kUndefined}, + {"QQze", Attribute::kUndefined}, + {"eterf", Attribute::kUndefined}, + {"vertjx", Attribute::kUndefined}, + {"v82wNNx", Attribute::kUndefined}, + {"worgroup_size", Attribute::kUndefined}, + {"workgrourr_size", Attribute::kUndefined}, + {"workgroGp_size", Attribute::kUndefined}, }; using AttributeParseTest = testing::TestWithParam<Case>;
diff --git a/src/tint/builtin/extension.cc b/src/tint/builtin/extension.cc index aca6220..a8d4567 100644 --- a/src/tint/builtin/extension.cc +++ b/src/tint/builtin/extension.cc
@@ -40,6 +40,9 @@ if (str == "chromium_experimental_push_constant") { return Extension::kChromiumExperimentalPushConstant; } + if (str == "chromium_internal_dual_source_blending") { + return Extension::kChromiumInternalDualSourceBlending; + } if (str == "chromium_internal_relaxed_uniform_layout") { return Extension::kChromiumInternalRelaxedUniformLayout; } @@ -61,6 +64,8 @@ return out << "chromium_experimental_full_ptr_parameters"; case Extension::kChromiumExperimentalPushConstant: return out << "chromium_experimental_push_constant"; + case Extension::kChromiumInternalDualSourceBlending: + return out << "chromium_internal_dual_source_blending"; case Extension::kChromiumInternalRelaxedUniformLayout: return out << "chromium_internal_relaxed_uniform_layout"; case Extension::kF16:
diff --git a/src/tint/builtin/extension.h b/src/tint/builtin/extension.h index 6beeda3..6883f4b 100644 --- a/src/tint/builtin/extension.h +++ b/src/tint/builtin/extension.h
@@ -36,6 +36,7 @@ kChromiumExperimentalDp4A, kChromiumExperimentalFullPtrParameters, kChromiumExperimentalPushConstant, + kChromiumInternalDualSourceBlending, kChromiumInternalRelaxedUniformLayout, kF16, }; @@ -51,9 +52,13 @@ Extension ParseExtension(std::string_view str); constexpr const char* kExtensionStrings[] = { - "chromium_disable_uniformity_analysis", "chromium_experimental_dp4a", - "chromium_experimental_full_ptr_parameters", "chromium_experimental_push_constant", - "chromium_internal_relaxed_uniform_layout", "f16", + "chromium_disable_uniformity_analysis", + "chromium_experimental_dp4a", + "chromium_experimental_full_ptr_parameters", + "chromium_experimental_push_constant", + "chromium_internal_dual_source_blending", + "chromium_internal_relaxed_uniform_layout", + "f16", }; // A unique vector of extensions
diff --git a/src/tint/builtin/extension_bench.cc b/src/tint/builtin/extension_bench.cc index 7a50281..447bf55 100644 --- a/src/tint/builtin/extension_bench.cc +++ b/src/tint/builtin/extension_bench.cc
@@ -59,20 +59,27 @@ "chromium_exp9rimFntal_ush_constant", "chrmium_experimental_push_constant", "cOOromium_experiVeHtal_puh_conRRtant", - "chromium_internl_relaxyd_uniform_layout", - "chromnnum_internrr77_Gelaxell_uniform_layout", - "chromium_intern4l_relaxe00_uniform_layout", + "chromium_internay_dual_sorce_blending", + "chrnnmium_internal_duGrr_source_bllend77ng", + "chromiu4_inter00al_dual_source_blending", + "chromium_internal_dual_source_blending", + "chromium_intoornal_dua_sorce_bleding", + "cromium_nternal_dual_sourcezzblending", + "chiiomiu_internal_dppal_source11blendig", + "XXhromium_internal_relaxed_uniform_layout", + "chromium_99nternal_reIInaxed_un55form_layout", + "chromYuaaSSinternrrl_rHHlaxed_uniform_layout", "chromium_internal_relaxed_uniform_layout", - "chrmoom_internal_relaxed_uniform_lyout", - "chroium_internal_rlaxed_uniform_layzzut", - "chromium_internaii_r11axed_uppifor_layout", - "f1XX", - "55199II", - "frSSHHa", + "chroiHm_internal_rkklaxd_uniform_layou", + "chromium_internl_relaxedguniRRorm_lajou", + "chromium_inernal_relaxedbuniorm_layout", + "fj6", + "f6", + "q", "f16", - "U", - "jV3", - "", + "fNN", + "fv", + "Q16", }; for (auto _ : state) { for (auto* str : kStrings) {
diff --git a/src/tint/builtin/extension_test.cc b/src/tint/builtin/extension_test.cc index 9ef8683..cc13b13 100644 --- a/src/tint/builtin/extension_test.cc +++ b/src/tint/builtin/extension_test.cc
@@ -48,6 +48,7 @@ {"chromium_experimental_full_ptr_parameters", Extension::kChromiumExperimentalFullPtrParameters}, {"chromium_experimental_push_constant", Extension::kChromiumExperimentalPushConstant}, + {"chromium_internal_dual_source_blending", Extension::kChromiumInternalDualSourceBlending}, {"chromium_internal_relaxed_uniform_layout", Extension::kChromiumInternalRelaxedUniformLayout}, {"f16", Extension::kF16}, }; @@ -65,12 +66,15 @@ {"chvomium_experimental_push_constiint", Extension::kUndefined}, {"chromiu8WWexperimental_push_constant", Extension::kUndefined}, {"chromium_experiMental_push_costanxx", Extension::kUndefined}, - {"chromium_internal_relaxed_unXform_layugg", Extension::kUndefined}, - {"chromiuu_iVterna_relxed_unifXrm_layout", Extension::kUndefined}, - {"chromium_internal_relaxed_uni3orm_layout", Extension::kUndefined}, - {"fE6", Extension::kUndefined}, - {"fPTT", Extension::kUndefined}, - {"dxx6", Extension::kUndefined}, + {"Xhromium_ggnternal_dual_sourceblending", Extension::kUndefined}, + {"chromium_internludual_sorce_bVenXing", Extension::kUndefined}, + {"chromium_internal_dual_source_b3ending", Extension::kUndefined}, + {"chromium_internal_rElaxed_uniform_layout", Extension::kUndefined}, + {"chromium_internalPPrTTlaed_uniform_layout", Extension::kUndefined}, + {"chroddium_internxxl_relaxed_unform_layout", Extension::kUndefined}, + {"4416", Extension::kUndefined}, + {"fSVV6", Extension::kUndefined}, + {"RR2", Extension::kUndefined}, }; using ExtensionParseTest = testing::TestWithParam<Case>;
diff --git a/src/tint/intrinsics.def b/src/tint/intrinsics.def index 0d667ab..4943c1c 100644 --- a/src/tint/intrinsics.def +++ b/src/tint/intrinsics.def
@@ -76,6 +76,8 @@ chromium_experimental_full_ptr_parameters // A Chromium-specific extension that relaxes memory layout requirements for uniform storage. chromium_internal_relaxed_uniform_layout + // A Chromium-specific extension that enables dual source blending. + chromium_internal_dual_source_blending } // https://gpuweb.github.io/gpuweb/wgsl/#storage-class @@ -260,6 +262,7 @@ fragment group id + index interpolate invariant location
diff --git a/src/tint/program_builder.h b/src/tint/program_builder.h index 4f4788f..f8866ac 100644 --- a/src/tint/program_builder.h +++ b/src/tint/program_builder.h
@@ -50,6 +50,7 @@ #include "src/tint/ast/if_statement.h" #include "src/tint/ast/increment_decrement_statement.h" #include "src/tint/ast/index_accessor_expression.h" +#include "src/tint/ast/index_attribute.h" #include "src/tint/ast/int_literal_expression.h" #include "src/tint/ast/interpolate_attribute.h" #include "src/tint/ast/invariant_attribute.h" @@ -3264,6 +3265,23 @@ return create<ast::LocationAttribute>(source_, Expr(std::forward<EXPR>(location))); } + /// Creates an ast::IndexAttribute + /// @param source the source information + /// @param index the index value expression + /// @returns the index attribute pointer + template <typename EXPR> + const ast::IndexAttribute* Index(const Source& source, EXPR&& index) { + return create<ast::IndexAttribute>(source, Expr(std::forward<EXPR>(index))); + } + + /// Creates an ast::IndexAttribute + /// @param index the index value expression + /// @returns the index attribute pointer + template <typename EXPR> + const ast::IndexAttribute* Index(EXPR&& index) { + return create<ast::IndexAttribute>(source_, Expr(std::forward<EXPR>(index))); + } + /// Creates an ast::IdAttribute /// @param source the source information /// @param id the id value
diff --git a/src/tint/reader/wgsl/parser_impl_enable_directive_test.cc b/src/tint/reader/wgsl/parser_impl_enable_directive_test.cc index 3c63a02..621206f 100644 --- a/src/tint/reader/wgsl/parser_impl_enable_directive_test.cc +++ b/src/tint/reader/wgsl/parser_impl_enable_directive_test.cc
@@ -164,7 +164,7 @@ // Error when unknown extension found EXPECT_TRUE(p->has_error()); EXPECT_EQ(p->error(), R"(1:8: expected extension -Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_relaxed_uniform_layout', 'f16')"); +Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_dual_source_blending', 'chromium_internal_relaxed_uniform_layout', 'f16')"); auto program = p->program(); auto& ast = program.AST(); EXPECT_EQ(ast.Enables().Length(), 0u); @@ -178,7 +178,7 @@ EXPECT_TRUE(p->has_error()); EXPECT_EQ(p->error(), R"(1:8: expected extension Did you mean 'f16'? -Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_relaxed_uniform_layout', 'f16')"); +Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_dual_source_blending', 'chromium_internal_relaxed_uniform_layout', 'f16')"); auto program = p->program(); auto& ast = program.AST(); EXPECT_EQ(ast.Enables().Length(), 0u); @@ -226,7 +226,7 @@ p->translation_unit(); EXPECT_TRUE(p->has_error()); EXPECT_EQ(p->error(), R"(1:8: expected extension -Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_relaxed_uniform_layout', 'f16')"); +Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_dual_source_blending', 'chromium_internal_relaxed_uniform_layout', 'f16')"); auto program = p->program(); auto& ast = program.AST(); EXPECT_EQ(ast.Enables().Length(), 0u); @@ -237,7 +237,7 @@ p->translation_unit(); EXPECT_TRUE(p->has_error()); EXPECT_EQ(p->error(), R"(1:8: expected extension -Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_relaxed_uniform_layout', 'f16')"); +Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_dual_source_blending', 'chromium_internal_relaxed_uniform_layout', 'f16')"); auto program = p->program(); auto& ast = program.AST(); EXPECT_EQ(ast.Enables().Length(), 0u); @@ -249,7 +249,7 @@ EXPECT_TRUE(p->has_error()); EXPECT_EQ(p->error(), R"(1:8: expected extension Did you mean 'f16'? -Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_relaxed_uniform_layout', 'f16')"); +Possible values: 'chromium_disable_uniformity_analysis', 'chromium_experimental_dp4a', 'chromium_experimental_full_ptr_parameters', 'chromium_experimental_push_constant', 'chromium_internal_dual_source_blending', 'chromium_internal_relaxed_uniform_layout', 'f16')"); auto program = p->program(); auto& ast = program.AST(); EXPECT_EQ(ast.Enables().Length(), 0u);
diff --git a/src/tint/reader/wgsl/parser_impl_error_resync_test.cc b/src/tint/reader/wgsl/parser_impl_error_resync_test.cc index f3e306a..9817bec 100644 --- a/src/tint/reader/wgsl/parser_impl_error_resync_test.cc +++ b/src/tint/reader/wgsl/parser_impl_error_resync_test.cc
@@ -54,7 +54,7 @@ ^ test.wgsl:4:2 error: expected attribute -Possible values: 'align', 'binding', 'builtin', 'compute', 'diagnostic', 'fragment', 'group', 'id', 'interpolate', 'invariant', 'location', 'must_use', 'size', 'vertex', 'workgroup_size' +Possible values: 'align', 'binding', 'builtin', 'compute', 'diagnostic', 'fragment', 'group', 'id', 'index', 'interpolate', 'invariant', 'location', 'must_use', 'size', 'vertex', 'workgroup_size' @_ fn -> {} ^ )"); @@ -122,7 +122,7 @@ ^^^^ test.wgsl:7:6 error: expected attribute -Possible values: 'align', 'binding', 'builtin', 'compute', 'diagnostic', 'fragment', 'group', 'id', 'interpolate', 'invariant', 'location', 'must_use', 'size', 'vertex', 'workgroup_size' +Possible values: 'align', 'binding', 'builtin', 'compute', 'diagnostic', 'fragment', 'group', 'id', 'index', 'interpolate', 'invariant', 'location', 'must_use', 'size', 'vertex', 'workgroup_size' @- x : i32, ^ )");
diff --git a/src/tint/reader/wgsl/parser_impl_function_attribute_list_test.cc b/src/tint/reader/wgsl/parser_impl_function_attribute_list_test.cc index ae43a0f..52d7da1 100644 --- a/src/tint/reader/wgsl/parser_impl_function_attribute_list_test.cc +++ b/src/tint/reader/wgsl/parser_impl_function_attribute_list_test.cc
@@ -53,7 +53,7 @@ EXPECT_TRUE(attrs.value.IsEmpty()); EXPECT_EQ(p->error(), R"(1:2: expected attribute Did you mean 'invariant'? -Possible values: 'align', 'binding', 'builtin', 'compute', 'diagnostic', 'fragment', 'group', 'id', 'interpolate', 'invariant', 'location', 'must_use', 'size', 'vertex', 'workgroup_size')"); +Possible values: 'align', 'binding', 'builtin', 'compute', 'diagnostic', 'fragment', 'group', 'id', 'index', 'interpolate', 'invariant', 'location', 'must_use', 'size', 'vertex', 'workgroup_size')"); } } // namespace
diff --git a/src/tint/reader/wgsl/parser_impl_variable_attribute_list_test.cc b/src/tint/reader/wgsl/parser_impl_variable_attribute_list_test.cc index 1751e98..76c27e7 100644 --- a/src/tint/reader/wgsl/parser_impl_variable_attribute_list_test.cc +++ b/src/tint/reader/wgsl/parser_impl_variable_attribute_list_test.cc
@@ -51,7 +51,7 @@ EXPECT_TRUE(attrs.value.IsEmpty()); EXPECT_EQ(p->error(), R"(1:2: expected attribute Did you mean 'invariant'? -Possible values: 'align', 'binding', 'builtin', 'compute', 'diagnostic', 'fragment', 'group', 'id', 'interpolate', 'invariant', 'location', 'must_use', 'size', 'vertex', 'workgroup_size')"); +Possible values: 'align', 'binding', 'builtin', 'compute', 'diagnostic', 'fragment', 'group', 'id', 'index', 'interpolate', 'invariant', 'location', 'must_use', 'size', 'vertex', 'workgroup_size')"); } } // namespace
diff --git a/src/tint/resolver/dependency_graph.cc b/src/tint/resolver/dependency_graph.cc index 504e8ac..84932ed 100644 --- a/src/tint/resolver/dependency_graph.cc +++ b/src/tint/resolver/dependency_graph.cc
@@ -34,6 +34,7 @@ #include "src/tint/ast/identifier.h" #include "src/tint/ast/if_statement.h" #include "src/tint/ast/increment_decrement_statement.h" +#include "src/tint/ast/index_attribute.h" #include "src/tint/ast/internal_attribute.h" #include "src/tint/ast/interpolate_attribute.h" #include "src/tint/ast/invariant_attribute.h" @@ -390,6 +391,10 @@ TraverseExpression(id->expr); return true; }, + [&](const ast::IndexAttribute* index) { + TraverseExpression(index->expr); + return true; + }, [&](const ast::InterpolateAttribute* interpolate) { TraverseExpression(interpolate->type); TraverseExpression(interpolate->sampling);
diff --git a/src/tint/resolver/dual_source_blending_extension_test.cc b/src/tint/resolver/dual_source_blending_extension_test.cc new file mode 100644 index 0000000..ebaf61c --- /dev/null +++ b/src/tint/resolver/dual_source_blending_extension_test.cc
@@ -0,0 +1,88 @@ +// Copyright 2022 The Tint Authors. +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. + +#include "src/tint/resolver/resolver.h" +#include "src/tint/resolver/resolver_test_helper.h" + +#include "gmock/gmock.h" + +using namespace tint::number_suffixes; // NOLINT + +namespace tint::resolver { +namespace { + +using DualSourceBlendingExtensionTest = ResolverTest; + +// Using the @index attribute without chromium_internal_dual_source_blending enabled should fail. +TEST_F(DualSourceBlendingExtensionTest, UseIndexAttribWithoutExtensionError) { + Structure("Output", utils::Vector{ + Member("a", ty.vec4<f32>(), + utils::Vector{Location(0_a), Index(Source{{12, 34}}, 0_a)}), + }); + + EXPECT_FALSE(r()->Resolve()); + EXPECT_EQ(r()->error(), + "12:34 error: use of '@index' attribute requires enabling extension " + "'chromium_internal_dual_source_blending'"); +} + +TEST_F(DualSourceBlendingExtensionTest, IndexF32Error) { + Enable(builtin::Extension::kChromiumInternalDualSourceBlending); + + Structure("Output", utils::Vector{ + Member(Source{{12, 34}}, "a", ty.vec4<f32>(), + utils::Vector{Location(0_a), Index(Source{{12, 34}}, 0_f)}), + }); + + EXPECT_FALSE(r()->Resolve()); + EXPECT_EQ(r()->error(), "12:34 error: @location must be an i32 or u32 value"); +} + +TEST_F(DualSourceBlendingExtensionTest, IndexFloatValueError) { + Enable(builtin::Extension::kChromiumInternalDualSourceBlending); + + Structure("Output", utils::Vector{ + Member(Source{{12, 34}}, "a", ty.vec4<f32>(), + utils::Vector{Location(0_a), Index(Source{{12, 34}}, 1.0_a)}), + }); + EXPECT_FALSE(r()->Resolve()); + EXPECT_EQ(r()->error(), "12:34 error: @location must be an i32 or u32 value"); +} + +TEST_F(DualSourceBlendingExtensionTest, IndexNegativeValue) { + Enable(builtin::Extension::kChromiumInternalDualSourceBlending); + + Structure("Output", utils::Vector{ + Member(Source{{12, 34}}, "a", ty.vec4<f32>(), + utils::Vector{Location(0_a), Index(Source{{12, 34}}, -1_a)}), + }); + + EXPECT_FALSE(r()->Resolve()); + EXPECT_EQ(r()->error(), "12:34 error: @index value must be zero or one"); +} + +TEST_F(DualSourceBlendingExtensionTest, IndexValueAboveOne) { + Enable(builtin::Extension::kChromiumInternalDualSourceBlending); + + Structure("Output", utils::Vector{ + Member(Source{{12, 34}}, "a", ty.vec4<f32>(), + utils::Vector{Location(0_a), Index(Source{{12, 34}}, 2_a)}), + }); + + EXPECT_FALSE(r()->Resolve()); + EXPECT_EQ(r()->error(), "12:34 error: @index value must be zero or one"); +} + +} // namespace +} // namespace tint::resolver
diff --git a/src/tint/resolver/resolver.cc b/src/tint/resolver/resolver.cc index 74a462a..51bf27c 100644 --- a/src/tint/resolver/resolver.cc +++ b/src/tint/resolver/resolver.cc
@@ -3570,6 +3570,30 @@ return static_cast<uint32_t>(value); } +utils::Result<uint32_t> Resolver::IndexAttribute(const ast::IndexAttribute* attr) { + ExprEvalStageConstraint constraint{sem::EvaluationStage::kConstant, "@index value"}; + TINT_SCOPED_ASSIGNMENT(expr_eval_stage_constraint_, constraint); + + auto* materialized = Materialize(ValueExpression(attr->expr)); + if (!materialized) { + return utils::Failure; + } + + if (!materialized->Type()->IsAnyOf<type::I32, type::U32>()) { + AddError("@location must be an i32 or u32 value", attr->source); + return utils::Failure; + } + + auto const_value = materialized->ConstantValue(); + auto value = const_value->ValueAs<AInt>(); + if (value != 0 && value != 1) { + AddError("@index value must be zero or one", attr->source); + return utils::Failure; + } + + return static_cast<uint32_t>(value); +} + utils::Result<uint32_t> Resolver::BindingAttribute(const ast::BindingAttribute* attr) { ExprEvalStageConstraint constraint{sem::EvaluationStage::kConstant, "@binding"}; TINT_SCOPED_ASSIGNMENT(expr_eval_stage_constraint_, constraint); @@ -4142,6 +4166,14 @@ attributes.location = value.Get(); return true; }, + [&](const ast::IndexAttribute* attr) { + auto value = IndexAttribute(attr); + if (!value) { + return false; + } + attributes.index = value.Get(); + return true; + }, [&](const ast::BuiltinAttribute* attr) { auto value = BuiltinAttribute(attr); if (!value) {
diff --git a/src/tint/resolver/resolver.h b/src/tint/resolver/resolver.h index 5d13d25..46ac83f 100644 --- a/src/tint/resolver/resolver.h +++ b/src/tint/resolver/resolver.h
@@ -320,6 +320,10 @@ /// @returns the location value on success. utils::Result<uint32_t> LocationAttribute(const ast::LocationAttribute* attr); + /// Resolves the `@index` attribute @p attr + /// @returns the index value on success. + utils::Result<uint32_t> IndexAttribute(const ast::IndexAttribute* attr); + /// Resolves the `@binding` attribute @p attr /// @returns the binding value on success. utils::Result<uint32_t> BindingAttribute(const ast::BindingAttribute* attr);
diff --git a/src/tint/resolver/validator.cc b/src/tint/resolver/validator.cc index eb009c2..17b75e1 100644 --- a/src/tint/resolver/validator.cc +++ b/src/tint/resolver/validator.cc
@@ -29,6 +29,7 @@ #include "src/tint/ast/for_loop_statement.h" #include "src/tint/ast/id_attribute.h" #include "src/tint/ast/if_statement.h" +#include "src/tint/ast/index_attribute.h" #include "src/tint/ast/internal_attribute.h" #include "src/tint/ast/interpolate_attribute.h" #include "src/tint/ast/loop_statement.h" @@ -1115,6 +1116,8 @@ is_input)) { return false; } + } else if (auto* index_attr = attr->As<ast::IndexAttribute>()) { + return IndexAttribute(index_attr); } else if (auto* interpolate = attr->As<ast::InterpolateAttribute>()) { if (decl->PipelineStage() == ast::PipelineStage::kCompute) { is_invalid_compute_shader_attribute = true; @@ -2113,6 +2116,7 @@ } return true; }, + [&](const ast::IndexAttribute* index) { return IndexAttribute(index); }, [&](const ast::BuiltinAttribute* builtin_attr) { if (!BuiltinAttribute(builtin_attr, member->Type(), stage, /* is_input */ false)) { @@ -2197,6 +2201,18 @@ return true; } +bool Validator::IndexAttribute(const ast::IndexAttribute* index_attr) const { + if (!enabled_extensions_.Contains(builtin::Extension::kChromiumInternalDualSourceBlending)) { + AddError( + "use of '@index' attribute requires enabling extension " + "'chromium_internal_dual_source_blending'", + index_attr->source); + return false; + } + + return false; +} + bool Validator::Return(const ast::ReturnStatement* ret, const type::Type* func_type, const type::Type* ret_type,
diff --git a/src/tint/resolver/validator.h b/src/tint/resolver/validator.h index 999bc49..19ceb5e 100644 --- a/src/tint/resolver/validator.h +++ b/src/tint/resolver/validator.h
@@ -329,6 +329,11 @@ const Source& source, const bool is_input = false) const; + /// Validates a index attribute + /// @param index_attr the index attribute to validate + /// @returns true on success, false otherwise. + bool IndexAttribute(const ast::IndexAttribute* index_attr) const; + /// Validates a loop statement /// @param stmt the loop statement /// @returns true on success, false otherwise.
diff --git a/src/tint/type/struct.h b/src/tint/type/struct.h index 720beca..907485c 100644 --- a/src/tint/type/struct.h +++ b/src/tint/type/struct.h
@@ -185,6 +185,8 @@ struct StructMemberAttributes { /// The value of a `@location` attribute std::optional<uint32_t> location; + /// The value of a `@index` attribute + std::optional<uint32_t> index; /// The value of a `@builtin` attribute std::optional<builtin::BuiltinValue> builtin; /// The values of a `@interpolate` attribute