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
