tint: Refactor Extensions / Enables.

* Extract ast::Enable::ExtensionKind to ast::Extension.
* Move the parsing out of ast::Enable and next to ast/extension.h
* Change the ast::Enable constructor to take the Extension, instead of
  a std::string. It's the WGSL parser's responsibility to parse, not the
  AST nodes.
* Add ProgramBuilder::Enable() helper.
* Keep ast::Module simple - keep track of the declared AST Enable nodes,
  don't do any deduplicating of the enabled extensions.
* Add the de-duplicated ast::Extensions to the sem::Module.
* Remove the kInternalExtensionForTesting enum value - we have kF16
  now, which can be used instead for testing.
* Rename kNoExtension to kNone.

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

src/dawn/tests/unittests/validation/ShaderModuleValidationTests.cpp

@@ -663,7 +663,7 @@ TEST_F(ShaderModuleValidationTest, MissingDecorations) {
 // Test that WGSL extension used by enable directives must be allowed by WebGPU.
 TEST_F(ShaderModuleValidationTest, ExtensionMustBeAllowed) {
     ASSERT_DEVICE_ERROR(utils::CreateShaderModule(device, R"(
-enable InternalExtensionForTesting;
+enable f16;
 
 @stage(compute) @workgroup_size(1) fn main() {})"));
 }

src/tint/BUILD.gn

@@ -226,6 +226,8 @@ libtint_source_set("libtint_core_all_src") {
     "ast/enable.h",
     "ast/expression.cc",
     "ast/expression.h",
+    "ast/extension.cc",
+    "ast/extension.h",
     "ast/external_texture.cc",
     "ast/external_texture.h",
     "ast/f16.cc",

src/tint/CMakeLists.txt

@@ -114,6 +114,8 @@ set(TINT_LIB_SRCS
   ast/enable.h
   ast/expression.cc
   ast/expression.h
+  ast/extension.cc
+  ast/extension.h
   ast/external_texture.cc
   ast/external_texture.h
   ast/f16.cc
@@ -691,6 +693,7 @@ if(TINT_BUILD_TESTS)
     ast/depth_texture_test.cc
     ast/discard_statement_test.cc
     ast/enable_test.cc
+    ast/extension_test.cc
     ast/external_texture_test.cc
     ast/f16_test.cc
     ast/f32_test.cc

src/tint/ast/enable.cc

@@ -21,47 +21,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::ast::Enable);
 
 namespace tint::ast {
 
-Enable::ExtensionKind Enable::NameToKind(const std::string& name) {
-    if (name == "chromium_experimental_dp4a") {
-        return Enable::ExtensionKind::kChromiumExperimentalDP4a;
-    }
-    if (name == "chromium_disable_uniformity_analysis") {
-        return Enable::ExtensionKind::kChromiumDisableUniformityAnalysis;
-    }
-    if (name == "f16") {
-        return Enable::ExtensionKind::kF16;
-    }
-
-    // The reserved internal extension name for testing
-    if (name == "InternalExtensionForTesting") {
-        return Enable::ExtensionKind::kInternalExtensionForTesting;
-    }
-
-    return Enable::ExtensionKind::kNoExtension;
-}
-
-std::string Enable::KindToName(ExtensionKind kind) {
-    switch (kind) {
-        case ExtensionKind::kChromiumExperimentalDP4a:
-            return "chromium_experimental_dp4a";
-        case ExtensionKind::kChromiumDisableUniformityAnalysis:
-            return "chromium_disable_uniformity_analysis";
-        case ExtensionKind::kF16:
-            return "f16";
-        // The reserved internal extension for testing
-        case ExtensionKind::kInternalExtensionForTesting:
-            return "InternalExtensionForTesting";
-        case ExtensionKind::kNoExtension:
-            // Return an empty string for kNoExtension
-            return {};
-            // No default case, as this switch must cover all ExtensionKind values.
-    }
-    // This return shall never get hit.
-    return {};
-}
-
-Enable::Enable(ProgramID pid, const Source& src, const std::string& ext_name)
-    : Base(pid, src), name(ext_name), kind(NameToKind(ext_name)) {}
+Enable::Enable(ProgramID pid, const Source& src, Extension ext) : Base(pid, src), extension(ext) {}
 
 Enable::Enable(Enable&&) = default;
 
@@ -69,6 +29,6 @@ Enable::~Enable() = default;
 
 const Enable* Enable::Clone(CloneContext* ctx) const {
     auto src = ctx->Clone(source);
-    return ctx->dst->create<Enable>(src, name);
+    return ctx->dst->create<Enable>(src, extension);
 }
 }  // namespace tint::ast

src/tint/ast/enable.h

@@ -16,63 +16,26 @@
 #define SRC_TINT_AST_ENABLE_H_
 
 #include <string>
-#include <unordered_set>
 #include <utility>
+#include <vector>
 
-#include "src/tint/ast/access.h"
-#include "src/tint/ast/expression.h"
+#include "src/tint/ast/extension.h"
+#include "src/tint/ast/node.h"
 
 namespace tint::ast {
 
-/// An instance of this class represents one extension mentioned in a
-/// "enable" derictive. Example:
+/// An "enable" directive. Example:
+/// ```
 ///   // Enable an extension named "f16"
 ///   enable f16;
-class Enable : public Castable<Enable, Node> {
+/// ```
+class Enable final : public Castable<Enable, Node> {
   public:
-    ///  The enum class identifing each supported WGSL extension
-    enum class ExtensionKind {
-        /// An internal reserved extension for test, named
-        /// "InternalExtensionForTesting".
-        kInternalExtensionForTesting,
-        /// WGSL Extension "f16"
-        kF16,
-
-        /// An extension for the experimental feature
-        /// "chromium_experimental_dp4a".
-        /// See crbug.com/tint/1497 for more details
-        kChromiumExperimentalDP4a,
-        /// A Chromium-specific extension for disabling uniformity analysis.
-        kChromiumDisableUniformityAnalysis,
-
-        /// Reserved for representing "No extension required" or "Not a valid extension".
-        kNoExtension,
-    };
-
-    /// Convert a string of extension name into one of ExtensionKind enum value,
-    /// the result will be ExtensionKind::kNoExtension if the name is not a
-    /// known extension name. A extension node of kind kNoExtension must not
-    /// exist in the AST tree, and using a unknown extension name in WGSL code
-    /// should result in a shader-creation error.
-    /// @param name string of the extension name
-    /// @return the ExtensionKind enum value for the extension of given name, or
-    /// kNoExtension if no known extension has the given name
-    static ExtensionKind NameToKind(const std::string& name);
-
-    /// Convert the ExtensionKind enum value to corresponding extension name
-    /// string. If the given enum value is kNoExtension or don't have a known
-    /// name, return an empty string instead.
-    /// @param kind the ExtensionKind enum value
-    /// @return string of the extension name corresponding to the given kind, or
-    /// an empty string if the given enum value is kNoExtension or don't have a
-    /// known corresponding name
-    static std::string KindToName(ExtensionKind kind);
-
     /// Create a extension
     /// @param pid the identifier of the program that owns this node
     /// @param src the source of this node
-    /// @param name the name of extension
-    Enable(ProgramID pid, const Source& src, const std::string& name);
+    /// @param ext the extension
+    Enable(ProgramID pid, const Source& src, Extension ext);
     /// Move constructor
     Enable(Enable&&);
 
@@ -85,14 +48,11 @@ class Enable : public Castable<Enable, Node> {
     const Enable* Clone(CloneContext* ctx) const override;
 
     /// The extension name
-    const std::string name;
-
-    /// The extension kind
-    const ExtensionKind kind;
+    const Extension extension;
 };
 
-///  A set of extension kinds
-using ExtensionSet = std::unordered_set<Enable::ExtensionKind>;
+/// A list of enables
+using EnableList = std::vector<const Enable*>;
 
 }  // namespace tint::ast
 

src/tint/ast/enable_test.cc

@@ -19,40 +19,15 @@
 namespace tint::ast {
 namespace {
 
-using AstExtensionTest = TestHelper;
+using EnableTest = TestHelper;
 
-TEST_F(AstExtensionTest, Creation) {
-    auto* ext =
-        create<Enable>(Source{Source::Range{Source::Location{20, 2}, Source::Location{20, 5}}},
-                       "InternalExtensionForTesting");
+TEST_F(EnableTest, Creation) {
+    auto* ext = create<ast::Enable>(Source{{{20, 2}, {20, 5}}}, Extension::kF16);
     EXPECT_EQ(ext->source.range.begin.line, 20u);
     EXPECT_EQ(ext->source.range.begin.column, 2u);
     EXPECT_EQ(ext->source.range.end.line, 20u);
     EXPECT_EQ(ext->source.range.end.column, 5u);
-    EXPECT_EQ(ext->kind, ast::Enable::ExtensionKind::kInternalExtensionForTesting);
-}
-
-TEST_F(AstExtensionTest, Creation_InvalidName) {
-    auto* ext = create<Enable>(
-        Source{Source::Range{Source::Location{20, 2}, Source::Location{20, 5}}}, std::string());
-    EXPECT_EQ(ext->source.range.begin.line, 20u);
-    EXPECT_EQ(ext->source.range.begin.column, 2u);
-    EXPECT_EQ(ext->source.range.end.line, 20u);
-    EXPECT_EQ(ext->source.range.end.column, 5u);
-    EXPECT_EQ(ext->kind, ast::Enable::ExtensionKind::kNoExtension);
-}
-
-TEST_F(AstExtensionTest, NameToKind_InvalidName) {
-    EXPECT_EQ(ast::Enable::NameToKind(std::string()), ast::Enable::ExtensionKind::kNoExtension);
-    EXPECT_EQ(ast::Enable::NameToKind("__ImpossibleExtensionName"),
-              ast::Enable::ExtensionKind::kNoExtension);
-    EXPECT_EQ(ast::Enable::NameToKind("123"), ast::Enable::ExtensionKind::kNoExtension);
-}
-
-TEST_F(AstExtensionTest, KindToName) {
-    EXPECT_EQ(ast::Enable::KindToName(ast::Enable::ExtensionKind::kInternalExtensionForTesting),
-              "InternalExtensionForTesting");
-    EXPECT_EQ(ast::Enable::KindToName(ast::Enable::ExtensionKind::kNoExtension), std::string());
+    EXPECT_EQ(ext->extension, Extension::kF16);
 }
 
 }  // namespace

src/tint/ast/extension.cc

@@ -0,0 +1,51 @@
+// Copyright 2022 The Tint Authors.
+//
+// Licensed under the Apache License, Version 2.0 (the "License");
+// you may not use this file except in compliance with the License.
+// You may obtain a copy of the License at
+//
+//     http://www.apache.org/licenses/LICENSE-2.0
+//
+// Unless required by applicable law or agreed to in writing, software
+// distributed under the License is distributed on an "AS IS" BASIS,
+// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+// See the License for the specific language governing permissions and
+// limitations under the License.
+
+#include "src/tint/ast/extension.h"
+
+namespace tint::ast {
+
+Extension ParseExtension(const std::string& name) {
+    if (name == "chromium_experimental_dp4a") {
+        return Extension::kChromiumExperimentalDP4a;
+    }
+    if (name == "chromium_disable_uniformity_analysis") {
+        return Extension::kChromiumDisableUniformityAnalysis;
+    }
+    if (name == "f16") {
+        return Extension::kF16;
+    }
+    return Extension::kNone;
+}
+
+const char* str(Extension ext) {
+    switch (ext) {
+        case Extension::kChromiumExperimentalDP4a:
+            return "chromium_experimental_dp4a";
+        case Extension::kChromiumDisableUniformityAnalysis:
+            return "chromium_disable_uniformity_analysis";
+        case Extension::kF16:
+            return "f16";
+        case Extension::kNone:
+            return "<none>";
+    }
+    return "<unknown>";
+}
+
+std::ostream& operator<<(std::ostream& out, Extension i) {
+    out << str(i);
+    return out;
+}
+
+}  // namespace tint::ast

src/tint/ast/extension.h

@@ -0,0 +1,68 @@
+// 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.
+
+#ifndef SRC_TINT_AST_EXTENSION_H_
+#define SRC_TINT_AST_EXTENSION_H_
+
+#include <sstream>
+#include <string>
+
+#include "src/tint/utils/unique_vector.h"
+
+namespace tint::ast {
+
+/// An enumerator of WGSL extensions
+enum class Extension {
+    /// WGSL Extension "f16"
+    kF16,
+
+    /// An extension for the experimental feature
+    /// "chromium_experimental_dp4a".
+    /// See crbug.com/tint/1497 for more details
+    kChromiumExperimentalDP4a,
+    /// A Chromium-specific extension for disabling uniformity analysis.
+    kChromiumDisableUniformityAnalysis,
+
+    /// Reserved for representing "No extension required" or "Not a valid extension".
+    kNone,
+};
+
+/// Convert a string of extension name into one of Extension enum value, the result will be
+/// Extension::kNone if the name is not a known extension name. A extension node of kind
+/// kNone must not exist in the AST tree, and using a unknown extension name in WGSL code
+/// should result in a shader-creation error.
+/// @param name string of the extension name
+/// @return the Extension enum value for the extension of given name, or kNone if no known extension
+/// has the given name
+Extension ParseExtension(const std::string& name);
+
+/// Convert the Extension enum value to corresponding extension name string.
+/// @param ext the Extension enum value
+/// @return string of the extension name corresponding to the given kind, or
+/// an empty string if the given enum value is kNone or don't have a
+/// known corresponding name
+const char* ExtensionName(Extension ext);
+
+/// @returns the name of the extension.
+const char* str(Extension i);
+
+/// Emits the name of the extension type.
+std::ostream& operator<<(std::ostream& out, Extension i);
+
+// A unique vector of extensions
+using Extensions = utils::UniqueVector<Extension>;
+
+}  // namespace tint::ast
+
+#endif  // SRC_TINT_AST_EXTENSION_H_

src/tint/ast/extension_test.cc

@@ -0,0 +1,36 @@
+
+// Copyright 2021 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/extension.h"
+
+#include "gtest/gtest.h"
+
+namespace tint::ast {
+namespace {
+
+TEST(ExtensionTest, NameToKind_InvalidName) {
+    EXPECT_EQ(ParseExtension("f16"), Extension::kF16);
+    EXPECT_EQ(ParseExtension(""), Extension::kNone);
+    EXPECT_EQ(ParseExtension("__ImpossibleExtensionName"), Extension::kNone);
+    EXPECT_EQ(ParseExtension("123"), Extension::kNone);
+}
+
+TEST(ExtensionTest, KindToName) {
+    EXPECT_EQ(std::string(str(Extension::kF16)), "f16");
+    EXPECT_EQ(std::string(str(Extension::kNone)), "<none>");
+}
+
+}  // namespace
+}  // namespace tint::ast

src/tint/ast/module.cc

@@ -68,18 +68,18 @@ void Module::BinGlobalDeclaration(const tint::ast::Node* decl, diag::List& diags
             TINT_ASSERT_PROGRAM_IDS_EQUAL_IF_VALID(AST, var, program_id);
             global_variables_.push_back(var);
         },
-        [&](const Enable* ext) {
-            TINT_ASSERT_PROGRAM_IDS_EQUAL_IF_VALID(AST, ext, program_id);
-            extensions_.insert(ext->kind);
+        [&](const Enable* enable) {
+            TINT_ASSERT_PROGRAM_IDS_EQUAL_IF_VALID(AST, enable, program_id);
+            enables_.push_back(enable);
         },
         [&](Default) { TINT_ICE(AST, diags) << "Unknown global declaration type"; });
 }
 
-void Module::AddEnable(const ast::Enable* ext) {
-    TINT_ASSERT(AST, ext);
-    TINT_ASSERT_PROGRAM_IDS_EQUAL_IF_VALID(AST, ext, program_id);
-    global_declarations_.push_back(ext);
-    extensions_.insert(ext->kind);
+void Module::AddEnable(const ast::Enable* enable) {
+    TINT_ASSERT(AST, enable);
+    TINT_ASSERT_PROGRAM_IDS_EQUAL_IF_VALID(AST, enable, program_id);
+    global_declarations_.push_back(enable);
+    enables_.push_back(enable);
 }
 
 void Module::AddGlobalVariable(const ast::Variable* var) {
@@ -117,7 +117,7 @@ void Module::Copy(CloneContext* ctx, const Module* src) {
     type_decls_.clear();
     functions_.clear();
     global_variables_.clear();
-    extensions_.clear();
+    enables_.clear();
 
     for (auto* decl : global_declarations_) {
         if (!decl) {

src/tint/ast/module.h

@@ -78,7 +78,7 @@ class Module final : public Castable<Module, Node> {
     VariableList& GlobalVariables() { return global_variables_; }
 
     /// @returns the extension set for the module
-    const ExtensionSet& Extensions() const { return extensions_; }
+    const EnableList& Enables() const { return enables_; }
 
     /// Adds a type declaration to the Builder.
     /// @param decl the type declaration to add
@@ -120,7 +120,7 @@ class Module final : public Castable<Module, Node> {
     std::vector<const TypeDecl*> type_decls_;
     FunctionList functions_;
     VariableList global_variables_;
-    ExtensionSet extensions_;
+    EnableList enables_;
 };
 
 }  // namespace tint::ast

src/tint/inspector/inspector.cc

@@ -19,6 +19,7 @@
 
 #include "src/tint/ast/bool_literal_expression.h"
 #include "src/tint/ast/call_expression.h"
+#include "src/tint/ast/extension.h"
 #include "src/tint/ast/float_literal_expression.h"
 #include "src/tint/ast/id_attribute.h"
 #include "src/tint/ast/interpolate_attribute.h"
@@ -32,6 +33,7 @@
 #include "src/tint/sem/function.h"
 #include "src/tint/sem/i32.h"
 #include "src/tint/sem/matrix.h"
+#include "src/tint/sem/module.h"
 #include "src/tint/sem/multisampled_texture.h"
 #include "src/tint/sem/sampled_texture.h"
 #include "src/tint/sem/statement.h"
@@ -544,16 +546,13 @@ uint32_t Inspector::GetWorkgroupStorageSize(const std::string& entry_point) {
 }
 
 std::vector<std::string> Inspector::GetUsedExtensionNames() {
-    std::vector<std::string> result;
-
-    ast::ExtensionSet set = program_->AST().Extensions();
-    result.reserve(set.size());
-    for (auto kind : set) {
-        std::string name = ast::Enable::KindToName(kind);
-        result.push_back(name);
+    auto& extensions = program_->Sem().Module()->Extensions();
+    std::vector<std::string> out;
+    out.reserve(extensions.size());
+    for (auto ext : extensions) {
+        out.push_back(ast::str(ext));
     }
-
-    return result;
+    return out;
 }
 
 std::vector<std::pair<std::string, Source>> Inspector::GetEnableDirectives() {
@@ -563,7 +562,7 @@ std::vector<std::pair<std::string, Source>> Inspector::GetEnableDirectives() {
     auto global_decls = program_->AST().GlobalDeclarations();
     for (auto* node : global_decls) {
         if (auto* ext = node->As<ast::Enable>()) {
-            result.push_back({ext->name, ext->source});
+            result.push_back({ast::str(ext->extension), ext->source});
         }
     }
 

src/tint/inspector/inspector_test.cc

@@ -2849,7 +2849,7 @@ fn main() {
 // Test calling GetUsedExtensionNames on a shader with valid extension.
 TEST_F(InspectorGetUsedExtensionNamesTest, Simple) {
     std::string shader = R"(
-enable InternalExtensionForTesting;
+enable f16;
 
 @stage(fragment)
 fn main() {
@@ -2859,15 +2859,15 @@ fn main() {
 
     auto result = inspector.GetUsedExtensionNames();
     EXPECT_EQ(result.size(), 1u);
-    EXPECT_EQ(result[0], "InternalExtensionForTesting");
+    EXPECT_EQ(result[0], "f16");
 }
 
 // Test calling GetUsedExtensionNames on a shader with a extension enabled for
 // multiple times.
 TEST_F(InspectorGetUsedExtensionNamesTest, Duplicated) {
     std::string shader = R"(
-enable InternalExtensionForTesting;
-enable InternalExtensionForTesting;
+enable f16;
+enable f16;
 
 @stage(fragment)
 fn main() {
@@ -2877,7 +2877,7 @@ fn main() {
 
     auto result = inspector.GetUsedExtensionNames();
     EXPECT_EQ(result.size(), 1u);
-    EXPECT_EQ(result[0], "InternalExtensionForTesting");
+    EXPECT_EQ(result[0], "f16");
 }
 
 // Test calling GetEnableDirectives on a empty shader.
@@ -2906,7 +2906,7 @@ fn main() {
 // Test calling GetEnableDirectives on a shader with valid extension.
 TEST_F(InspectorGetEnableDirectivesTest, Simple) {
     std::string shader = R"(
-enable InternalExtensionForTesting;
+enable f16;
 
 @stage(fragment)
 fn main() {
@@ -2916,17 +2916,17 @@ fn main() {
 
     auto result = inspector.GetEnableDirectives();
     EXPECT_EQ(result.size(), 1u);
-    EXPECT_EQ(result[0].first, "InternalExtensionForTesting");
-    EXPECT_EQ(result[0].second.range, (Source::Range{{2, 8}, {2, 35}}));
+    EXPECT_EQ(result[0].first, "f16");
+    EXPECT_EQ(result[0].second.range, (Source::Range{{2, 8}, {2, 11}}));
 }
 
 // Test calling GetEnableDirectives on a shader with a extension enabled for
 // multiple times.
 TEST_F(InspectorGetEnableDirectivesTest, Duplicated) {
     std::string shader = R"(
-enable InternalExtensionForTesting;
+enable f16;
 
-enable InternalExtensionForTesting;
+enable f16;
 @stage(fragment)
 fn main() {
 })";
@@ -2935,10 +2935,10 @@ fn main() {
 
     auto result = inspector.GetEnableDirectives();
     EXPECT_EQ(result.size(), 2u);
-    EXPECT_EQ(result[0].first, "InternalExtensionForTesting");
-    EXPECT_EQ(result[0].second.range, (Source::Range{{2, 8}, {2, 35}}));
-    EXPECT_EQ(result[1].first, "InternalExtensionForTesting");
-    EXPECT_EQ(result[1].second.range, (Source::Range{{4, 8}, {4, 35}}));
+    EXPECT_EQ(result[0].first, "f16");
+    EXPECT_EQ(result[0].second.range, (Source::Range{{2, 8}, {2, 11}}));
+    EXPECT_EQ(result[1].first, "f16");
+    EXPECT_EQ(result[1].second.range, (Source::Range{{4, 8}, {4, 11}}));
 }
 
 // Crash was occuring in ::GenerateSamplerTargets, when

src/tint/program_builder.h

@@ -39,6 +39,7 @@
 #include "src/tint/ast/disable_validation_attribute.h"
 #include "src/tint/ast/discard_statement.h"
 #include "src/tint/ast/enable.h"
+#include "src/tint/ast/extension.h"
 #include "src/tint/ast/external_texture.h"
 #include "src/tint/ast/f16.h"
 #include "src/tint/ast/f32.h"
@@ -1307,6 +1308,15 @@ class ProgramBuilder {
         return Construct(ty.array(subtype, std::forward<EXPR>(n)), std::forward<ARGS>(args)...);
     }
 
+    /// Adds the extension to the list of enable directives at the top of the module.
+    /// @param ext the extension to enable
+    /// @return an `ast::Enable` enabling the given extension.
+    const ast::Enable* Enable(ast::Extension ext) {
+        auto* enable = create<ast::Enable>(ext);
+        AST().AddEnable(enable);
+        return enable;
+    }
+
     /// @param name the variable name
     /// @param type the variable type
     /// @param optional the optional variable settings.

src/tint/reader/wgsl/parser_impl.cc

@@ -366,13 +366,11 @@ Maybe<bool> ParserImpl::enable_directive() {
             return Failure::kErrored;
         }
 
-        if (ast::Enable::NameToKind(name.value) != ast::Enable::ExtensionKind::kNoExtension) {
-            const ast::Enable* extension = create<ast::Enable>(name.source, name.value);
-            builder_.AST().AddEnable(extension);
-        } else {
-            // Error if an unknown extension is used
+        auto extension = ast::ParseExtension(name.value);
+        if (extension == ast::Extension::kNone) {
             return add_error(name.source, "unsupported extension: '" + name.value + "'");
         }
+        builder_.AST().AddEnable(create<ast::Enable>(name.source, extension));
 
         return true;
     });

src/tint/reader/wgsl/parser_impl_enable_directive_test.cc

@@ -23,41 +23,36 @@ using EnableDirectiveTest = ParserImplTest;
 
 // Test a valid enable directive.
 TEST_F(EnableDirectiveTest, Valid) {
-    auto p = parser("enable InternalExtensionForTesting;");
+    auto p = parser("enable f16;");
     p->enable_directive();
     EXPECT_FALSE(p->has_error()) << p->error();
     auto program = p->program();
     auto& ast = program.AST();
-    EXPECT_EQ(ast.Extensions(),
-              ast::ExtensionSet{ast::Enable::ExtensionKind::kInternalExtensionForTesting});
-    EXPECT_EQ(ast.GlobalDeclarations().size(), 1u);
-    auto* node = ast.GlobalDeclarations()[0]->As<ast::Enable>();
-    EXPECT_TRUE(node != nullptr);
-    EXPECT_EQ(node->name, "InternalExtensionForTesting");
-    EXPECT_EQ(node->kind, ast::Enable::ExtensionKind::kInternalExtensionForTesting);
+    ASSERT_EQ(ast.Enables().size(), 1u);
+    auto* enable = ast.Enables()[0];
+    EXPECT_EQ(enable->extension, ast::Extension::kF16);
+    ASSERT_EQ(ast.GlobalDeclarations().size(), 1u);
+    EXPECT_EQ(ast.GlobalDeclarations()[0], enable);
 }
 
 // Test multiple enable directives for a same extension.
 TEST_F(EnableDirectiveTest, EnableMultipleTime) {
     auto p = parser(R"(
-enable InternalExtensionForTesting;
-enable InternalExtensionForTesting;
+enable f16;
+enable f16;
 )");
     p->translation_unit();
     EXPECT_FALSE(p->has_error()) << p->error();
     auto program = p->program();
     auto& ast = program.AST();
-    EXPECT_EQ(ast.Extensions(),
-              ast::ExtensionSet{ast::Enable::ExtensionKind::kInternalExtensionForTesting});
-    EXPECT_EQ(ast.GlobalDeclarations().size(), 2u);
-    auto* node1 = ast.GlobalDeclarations()[0]->As<ast::Enable>();
-    EXPECT_TRUE(node1 != nullptr);
-    EXPECT_EQ(node1->name, "InternalExtensionForTesting");
-    EXPECT_EQ(node1->kind, ast::Enable::ExtensionKind::kInternalExtensionForTesting);
-    auto* node2 = ast.GlobalDeclarations()[1]->As<ast::Enable>();
-    EXPECT_TRUE(node2 != nullptr);
-    EXPECT_EQ(node2->name, "InternalExtensionForTesting");
-    EXPECT_EQ(node2->kind, ast::Enable::ExtensionKind::kInternalExtensionForTesting);
+    ASSERT_EQ(ast.Enables().size(), 2u);
+    auto* enable_a = ast.Enables()[0];
+    auto* enable_b = ast.Enables()[1];
+    EXPECT_EQ(enable_a->extension, ast::Extension::kF16);
+    EXPECT_EQ(enable_b->extension, ast::Extension::kF16);
+    ASSERT_EQ(ast.GlobalDeclarations().size(), 2u);
+    EXPECT_EQ(ast.GlobalDeclarations()[0], enable_a);
+    EXPECT_EQ(ast.GlobalDeclarations()[1], enable_b);
 }
 
 // Test an unknown extension identifier.
@@ -69,42 +64,42 @@ TEST_F(EnableDirectiveTest, InvalidIdentifier) {
     EXPECT_EQ(p->error(), "1:8: unsupported extension: 'NotAValidExtensionName'");
     auto program = p->program();
     auto& ast = program.AST();
-    EXPECT_EQ(ast.Extensions().size(), 0u);
+    EXPECT_EQ(ast.Enables().size(), 0u);
     EXPECT_EQ(ast.GlobalDeclarations().size(), 0u);
 }
 
-// Test an enable directive missing ending semiclon.
-TEST_F(EnableDirectiveTest, MissingEndingSemiclon) {
-    auto p = parser("enable InternalExtensionForTesting");
+// Test an enable directive missing ending semicolon.
+TEST_F(EnableDirectiveTest, MissingEndingSemicolon) {
+    auto p = parser("enable f16");
     p->translation_unit();
     EXPECT_TRUE(p->has_error());
-    EXPECT_EQ(p->error(), "1:35: expected ';' for enable directive");
+    EXPECT_EQ(p->error(), "1:11: expected ';' for enable directive");
     auto program = p->program();
     auto& ast = program.AST();
-    EXPECT_EQ(ast.Extensions().size(), 0u);
+    EXPECT_EQ(ast.Enables().size(), 0u);
     EXPECT_EQ(ast.GlobalDeclarations().size(), 0u);
 }
 
 // Test using invalid tokens in an enable directive.
 TEST_F(EnableDirectiveTest, InvalidTokens) {
     {
-        auto p = parser("enable InternalExtensionForTesting<;");
+        auto p = parser("enable f16<;");
         p->translation_unit();
         EXPECT_TRUE(p->has_error());
-        EXPECT_EQ(p->error(), "1:35: expected ';' for enable directive");
+        EXPECT_EQ(p->error(), "1:11: expected ';' for enable directive");
         auto program = p->program();
         auto& ast = program.AST();
-        EXPECT_EQ(ast.Extensions().size(), 0u);
+        EXPECT_EQ(ast.Enables().size(), 0u);
         EXPECT_EQ(ast.GlobalDeclarations().size(), 0u);
     }
     {
-        auto p = parser("enable <InternalExtensionForTesting;");
+        auto p = parser("enable <f16;");
         p->translation_unit();
         EXPECT_TRUE(p->has_error());
         EXPECT_EQ(p->error(), "1:8: invalid extension name");
         auto program = p->program();
         auto& ast = program.AST();
-        EXPECT_EQ(ast.Extensions().size(), 0u);
+        EXPECT_EQ(ast.Enables().size(), 0u);
         EXPECT_EQ(ast.GlobalDeclarations().size(), 0u);
     }
     {
@@ -114,7 +109,7 @@ TEST_F(EnableDirectiveTest, InvalidTokens) {
         EXPECT_EQ(p->error(), "1:8: invalid extension name");
         auto program = p->program();
         auto& ast = program.AST();
-        EXPECT_EQ(ast.Extensions().size(), 0u);
+        EXPECT_EQ(ast.Enables().size(), 0u);
         EXPECT_EQ(ast.GlobalDeclarations().size(), 0u);
     }
     {
@@ -124,7 +119,7 @@ TEST_F(EnableDirectiveTest, InvalidTokens) {
         EXPECT_EQ(p->error(), "1:8: invalid extension name");
         auto program = p->program();
         auto& ast = program.AST();
-        EXPECT_EQ(ast.Extensions().size(), 0u);
+        EXPECT_EQ(ast.Enables().size(), 0u);
         EXPECT_EQ(ast.GlobalDeclarations().size(), 0u);
     }
 }
@@ -133,35 +128,39 @@ TEST_F(EnableDirectiveTest, InvalidTokens) {
 TEST_F(EnableDirectiveTest, FollowingOtherGlobalDecl) {
     auto p = parser(R"(
 var<private> t: f32 = 0f;
-enable InternalExtensionForTesting;
+enable f16;
 )");
     p->translation_unit();
     EXPECT_TRUE(p->has_error());
     EXPECT_EQ(p->error(), "3:1: enable directives must come before all global declarations");
     auto program = p->program();
     auto& ast = program.AST();
-    // Accept the enable directive although it cause an error
-    EXPECT_EQ(ast.Extensions(),
-              ast::ExtensionSet{ast::Enable::ExtensionKind::kInternalExtensionForTesting});
-    EXPECT_EQ(ast.GlobalDeclarations().size(), 2u);
+    // Accept the enable directive although it caused an error
+    ASSERT_EQ(ast.Enables().size(), 1u);
+    auto* enable = ast.Enables()[0];
+    EXPECT_EQ(enable->extension, ast::Extension::kF16);
+    ASSERT_EQ(ast.GlobalDeclarations().size(), 2u);
+    EXPECT_EQ(ast.GlobalDeclarations()[1], enable);
 }
 
-// Test an enable directive go after an empty semiclon.
-TEST_F(EnableDirectiveTest, FollowingEmptySemiclon) {
+// Test an enable directive go after an empty semicolon.
+TEST_F(EnableDirectiveTest, FollowingEmptySemicolon) {
     auto p = parser(R"(
 ;
-enable InternalExtensionForTesting;
+enable f16;
 )");
     p->translation_unit();
-    // An empty semiclon is treated as a global declaration
+    // An empty semicolon is treated as a global declaration
     EXPECT_TRUE(p->has_error());
     EXPECT_EQ(p->error(), "3:1: enable directives must come before all global declarations");
     auto program = p->program();
     auto& ast = program.AST();
     // Accept the enable directive although it cause an error
-    EXPECT_EQ(ast.Extensions(),
-              ast::ExtensionSet{ast::Enable::ExtensionKind::kInternalExtensionForTesting});
-    EXPECT_EQ(ast.GlobalDeclarations().size(), 1u);
+    ASSERT_EQ(ast.Enables().size(), 1u);
+    auto* enable = ast.Enables()[0];
+    EXPECT_EQ(enable->extension, ast::Extension::kF16);
+    ASSERT_EQ(ast.GlobalDeclarations().size(), 1u);
+    EXPECT_EQ(ast.GlobalDeclarations()[0], enable);
 }
 
 }  // namespace

src/tint/resolver/builtin_validation_test.cc

@@ -378,10 +378,7 @@ using ResolverDP4aExtensionValidationTest = ResolverTest;
 TEST_F(ResolverDP4aExtensionValidationTest, Dot4I8PackedWithExtension) {
     // enable chromium_experimental_dp4a;
     // fn func { return dot4I8Packed(1u, 2u); }
-    auto* ext =
-        create<ast::Enable>(Source{Source::Range{Source::Location{10, 2}, Source::Location{10, 5}}},
-                            "chromium_experimental_dp4a");
-    AST().AddEnable(ext);
+    Enable(ast::Extension::kChromiumExperimentalDP4a);
 
     Func("func", {}, ty.i32(),
          {
@@ -409,10 +406,7 @@ TEST_F(ResolverDP4aExtensionValidationTest, Dot4I8PackedWithoutExtension) {
 TEST_F(ResolverDP4aExtensionValidationTest, Dot4U8PackedWithExtension) {
     // enable chromium_experimental_dp4a;
     // fn func { return dot4U8Packed(1u, 2u); }
-    auto* ext =
-        create<ast::Enable>(Source{Source::Range{Source::Location{10, 2}, Source::Location{10, 5}}},
-                            "chromium_experimental_dp4a");
-    AST().AddEnable(ext);
+    Enable(ast::Extension::kChromiumExperimentalDP4a);
 
     Func("func", {}, ty.u32(),
          {

src/tint/resolver/resolver.cc

@@ -101,9 +101,6 @@ bool Resolver::Resolve() {
         return false;
     }
 
-    // Create the semantic module
-    builder_->Sem().SetModule(builder_->create<sem::Module>(dependencies_.ordered_globals));
-
     bool result = ResolveInternal();
 
     if (!result && !diagnostics_.contains_errors()) {
@@ -111,6 +108,10 @@ bool Resolver::Resolve() {
         return false;
     }
 
+    // Create the semantic module
+    builder_->Sem().SetModule(builder_->create<sem::Module>(
+        std::move(dependencies_.ordered_globals), std::move(enabled_extensions_)));
+
     return result;
 }
 
@@ -120,19 +121,16 @@ bool Resolver::ResolveInternal() {
     // Process all module-scope declarations in dependency order.
     for (auto* decl : dependencies_.ordered_globals) {
         Mark(decl);
-        // Enable directives don't have sem node.
-        if (decl->Is<ast::Enable>()) {
-            continue;
-        }
-        if (!Switch(
+        if (!Switch<bool>(
                 decl,  //
+                [&](const ast::Enable* e) { return Enable(e); },
                 [&](const ast::TypeDecl* td) { return TypeDecl(td); },
                 [&](const ast::Function* func) { return Function(func); },
                 [&](const ast::Variable* var) { return GlobalVariable(var); },
                 [&](Default) {
                     TINT_UNREACHABLE(Resolver, diagnostics_)
                         << "unhandled global declaration: " << decl->TypeInfo().name;
-                    return nullptr;
+                    return false;
                 })) {
             return false;
         }
@@ -146,9 +144,11 @@ bool Resolver::ResolveInternal() {
         return false;
     }
 
+    if (!enabled_extensions_.contains(ast::Extension::kChromiumDisableUniformityAnalysis)) {
         if (!AnalyzeUniformity(builder_, dependencies_)) {
             // TODO(jrprice): Reject programs that fail uniformity analysis.
         }
+    }
 
     bool result = true;
     for (auto* node : builder_->ASTNodes().Objects()) {
@@ -174,7 +174,7 @@ sem::Type* Resolver::Type(const ast::Type* ty) {
         [&](const ast::U32*) { return builder_->create<sem::U32>(); },
         [&](const ast::F16* t) -> sem::F16* {
             // Validate if f16 type is allowed.
-            if (builder_->AST().Extensions().count(ast::Enable::ExtensionKind::kF16) == 0) {
+            if (!enabled_extensions_.contains(ast::Extension::kF16)) {
                 AddError("f16 used without 'f16' extension enabled", t->source);
                 return nullptr;
             }
@@ -1358,7 +1358,7 @@ sem::Call* Resolver::BuiltinCall(const ast::CallExpression* expr,
 
     current_function_->AddDirectlyCalledBuiltin(builtin);
 
-    if (!validator_.RequiredExtensionForBuiltinFunction(call, builder_->AST().Extensions())) {
+    if (!validator_.RequiredExtensionForBuiltinFunction(call, enabled_extensions_)) {
         return nullptr;
     }
 
@@ -1750,6 +1750,11 @@ sem::Expression* Resolver::UnaryOp(const ast::UnaryOpExpression* unary) {
     return sem;
 }
 
+bool Resolver::Enable(const ast::Enable* enable) {
+    enabled_extensions_.add(enable->extension);
+    return true;
+}
+
 sem::Type* Resolver::TypeDecl(const ast::TypeDecl* named_type) {
     sem::Type* result = nullptr;
     if (auto* alias = named_type->As<ast::Alias>()) {

src/tint/resolver/resolver.h

@@ -228,6 +228,10 @@ class Resolver {
     /// @param ty the ast::Type
     sem::Type* Type(const ast::Type* ty);
 
+    /// @param enable the enable declaration
+    /// @returns the resolved extension
+    bool Enable(const ast::Enable* enable);
+
     /// @param named_type the named type to resolve
     /// @returns the resolved semantic type
     sem::Type* TypeDecl(const ast::TypeDecl* named_type);
@@ -351,6 +355,7 @@ class Resolver {
     DependencyGraph dependencies_;
     SemHelper sem_;
     Validator validator_;
+    ast::Extensions enabled_extensions_;
     std::vector<sem::Function*> entry_points_;
     std::unordered_map<const sem::Type*, const Source&> atomic_composite_info_;
     std::unordered_set<const ast::Node*> marked_;

src/tint/resolver/type_validation_test.cc

@@ -665,8 +665,8 @@ TEST_F(ResolverTypeValidationTest, BuiltinAsType) {
 TEST_F(ResolverTypeValidationTest, F16TypeUsedWithExtension) {
     // enable f16;
     // var<private> v : f16;
-    auto* ext = create<ast::Enable>("f16");
-    AST().AddEnable(ext);
+    Enable(ast::Extension::kF16);
+
     Global("v", ty.f16(), ast::StorageClass::kPrivate);
 
     EXPECT_TRUE(r()->Resolve()) << r()->error();

src/tint/resolver/uniformity.cc

@@ -1548,11 +1548,6 @@ class UniformityGraph {
 }  // namespace
 
 bool AnalyzeUniformity(ProgramBuilder* builder, const DependencyGraph& dependency_graph) {
-    if (builder->AST().Extensions().count(
-            ast::Enable::ExtensionKind::kChromiumDisableUniformityAnalysis)) {
-        return true;
-    }
-
     UniformityGraph graph(builder);
     return graph.Build(dependency_graph);
 }

src/tint/resolver/validator.cc

@@ -1553,21 +1553,22 @@ bool Validator::TextureBuiltinFunction(const sem::Call* call) const {
            check_arg_is_constexpr(sem::ParameterUsage::kComponent, 0, 3);
 }
 
-bool Validator::RequiredExtensionForBuiltinFunction(const sem::Call* call,
-                                                    const ast::ExtensionSet& extensionSet) const {
+bool Validator::RequiredExtensionForBuiltinFunction(
+    const sem::Call* call,
+    const ast::Extensions& enabled_extensions) const {
     const auto* builtin = call->Target()->As<sem::Builtin>();
     if (!builtin) {
         return true;
     }
 
     const auto extension = builtin->RequiredExtension();
-    if (extension == ast::Enable::ExtensionKind::kNoExtension) {
+    if (extension == ast::Extension::kNone) {
         return true;
     }
 
-    if (extensionSet.find(extension) == extensionSet.cend()) {
+    if (!enabled_extensions.contains(extension)) {
         AddError("cannot call built-in function '" + std::string(builtin->str()) +
-                     "' without extension " + ast::Enable::KindToName(extension),
+                     "' without extension " + ast::str(extension),
                  call->Declaration()->source);
         return false;
     }

src/tint/resolver/validator.h

@@ -361,10 +361,10 @@ class Validator {
 
     /// Validates an optional builtin function and its required extension.
     /// @param call the builtin call to validate
-    /// @param extensionSet all the extensions declared in current module
+    /// @param enabled_extensions all the extensions declared in current module
     /// @returns true on success, false otherwise
     bool RequiredExtensionForBuiltinFunction(const sem::Call* call,
-                                             const ast::ExtensionSet& extensionSet) const;
+                                             const ast::Extensions& enabled_extensions) const;
 
     /// Validates there are no duplicate attributes
     /// @param attributes the list of attributes to validate

src/tint/sem/builtin.cc

@@ -153,11 +153,11 @@ bool Builtin::HasSideEffects() const {
     return false;
 }
 
-ast::Enable::ExtensionKind Builtin::RequiredExtension() const {
+ast::Extension Builtin::RequiredExtension() const {
     if (IsDP4a()) {
-        return ast::Enable::ExtensionKind::kChromiumExperimentalDP4a;
+        return ast::Extension::kChromiumExperimentalDP4a;
     }
-    return ast::Enable::ExtensionKind::kNoExtension;
+    return ast::Extension::kNone;
 }
 
 }  // namespace tint::sem

src/tint/sem/builtin.h

@@ -18,6 +18,7 @@
 #include <string>
 #include <vector>
 
+#include "src/tint/ast/extension.h"
 #include "src/tint/sem/builtin_type.h"
 #include "src/tint/sem/call_target.h"
 #include "src/tint/sem/pipeline_stage_set.h"
@@ -144,8 +145,8 @@ class Builtin final : public Castable<Builtin, CallTarget> {
     bool HasSideEffects() const;
 
     /// @returns the required extension of this builtin function. Returns
-    /// ast::Enable::ExtensionKind::kNoExtension if no extension is required.
-    ast::Enable::ExtensionKind RequiredExtension() const;
+    /// ast::Extension::kNone if no extension is required.
+    ast::Extension RequiredExtension() const;
 
   private:
     const BuiltinType type_;

src/tint/sem/module.cc

@@ -21,8 +21,8 @@ TINT_INSTANTIATE_TYPEINFO(tint::sem::Module);
 
 namespace tint::sem {
 
-Module::Module(std::vector<const ast::Node*> dep_ordered_decls)
-    : dep_ordered_decls_(std::move(dep_ordered_decls)) {}
+Module::Module(std::vector<const ast::Node*> dep_ordered_decls, ast::Extensions extensions)
+    : dep_ordered_decls_(std::move(dep_ordered_decls)), extensions_(std::move(extensions)) {}
 
 Module::~Module() = default;
 

src/tint/sem/module.h

@@ -17,6 +17,7 @@
 
 #include <vector>
 
+#include "src/tint/ast/extension.h"
 #include "src/tint/sem/node.h"
 
 // Forward declarations
@@ -33,7 +34,8 @@ class Module final : public Castable<Module, Node> {
   public:
     /// Constructor
     /// @param dep_ordered_decls the dependency-ordered module-scope declarations
-    explicit Module(std::vector<const ast::Node*> dep_ordered_decls);
+    /// @param extensions the list of enabled extensions in the module
+    Module(std::vector<const ast::Node*> dep_ordered_decls, ast::Extensions extensions);
 
     /// Destructor
     ~Module() override;
@@ -43,8 +45,12 @@ class Module final : public Castable<Module, Node> {
         return dep_ordered_decls_;
     }
 
+    /// @returns the list of enabled extensions in the module
+    const ast::Extensions& Extensions() const { return extensions_; }
+
   private:
     const std::vector<const ast::Node*> dep_ordered_decls_;
+    ast::Extensions extensions_;
 };
 
 }  // namespace tint::sem

src/tint/transform/disable_uniformity_analysis.cc

@@ -17,6 +17,7 @@
 #include <utility>
 
 #include "src/tint/program_builder.h"
+#include "src/tint/sem/module.h"
 
 TINT_INSTANTIATE_TYPEINFO(tint::transform::DisableUniformityAnalysis);
 
@@ -27,13 +28,12 @@ DisableUniformityAnalysis::DisableUniformityAnalysis() = default;
 DisableUniformityAnalysis::~DisableUniformityAnalysis() = default;
 
 bool DisableUniformityAnalysis::ShouldRun(const Program* program, const DataMap&) const {
-    return !program->AST().Extensions().count(
-        ast::Enable::ExtensionKind::kChromiumDisableUniformityAnalysis);
+    return !program->Sem().Module()->Extensions().contains(
+        ast::Extension::kChromiumDisableUniformityAnalysis);
 }
 
 void DisableUniformityAnalysis::Run(CloneContext& ctx, const DataMap&, DataMap&) const {
-    ctx.dst->AST().AddEnable(ctx.dst->create<ast::Enable>(
-        ast::Enable::KindToName(ast::Enable::ExtensionKind::kChromiumDisableUniformityAnalysis)));
+    ctx.dst->Enable(ast::Extension::kChromiumDisableUniformityAnalysis);
     ctx.Clone();
 }
 

src/tint/writer/hlsl/generator_impl_builtin_test.cc

@@ -727,10 +727,7 @@ void main() {
 }
 
 TEST_F(HlslGeneratorImplTest_Builtin, Dot4I8Packed) {
-    auto* ext =
-        create<ast::Enable>(Source{Source::Range{Source::Location{10, 2}, Source::Location{10, 5}}},
-                            "chromium_experimental_dp4a");
-    AST().AddEnable(ext);
+    Enable(ast::Extension::kChromiumExperimentalDP4a);
 
     auto* val1 = Var("val1", ty.u32());
     auto* val2 = Var("val2", ty.u32());
@@ -756,10 +753,7 @@ void test_function() {
 }
 
 TEST_F(HlslGeneratorImplTest_Builtin, Dot4U8Packed) {
-    auto* ext =
-        create<ast::Enable>(Source{Source::Range{Source::Location{10, 2}, Source::Location{10, 5}}},
-                            "chromium_experimental_dp4a");
-    AST().AddEnable(ext);
+    Enable(ast::Extension::kChromiumExperimentalDP4a);
 
     auto* val1 = Var("val1", ty.u32());
     auto* val2 = Var("val2", ty.u32());

src/tint/writer/spirv/builder.cc

@@ -256,7 +256,7 @@ bool Builder::Build() {
     push_memory_model(spv::Op::OpMemoryModel,
                       {U32Operand(SpvAddressingModelLogical), U32Operand(SpvMemoryModelGLSL450)});
 
-    for (auto ext : builder_.AST().Extensions()) {
+    for (auto ext : builder_.Sem().Module()->Extensions()) {
         GenerateExtension(ext);
     }
 
@@ -366,7 +366,7 @@ void Builder::push_capability(uint32_t cap) {
     }
 }
 
-bool Builder::GenerateExtension(ast::Enable::ExtensionKind) {
+bool Builder::GenerateExtension(ast::Extension) {
     /*
     For each supported extension, push corresponding capability into the builder.
     For example:

src/tint/writer/spirv/builder.h

@@ -224,11 +224,11 @@ class Builder {
                                      ast::InterpolationType type,
                                      ast::InterpolationSampling sampling);
 
-    /// Generates a extension for the given extension kind. Emits an error and
-    /// returns false if the extension kind is not supported.
-    /// @param kind ExtensionKind of the extension to generate
+    /// Generates the enabling of an extension. Emits an error and returns false if the extension is
+    /// not supported.
+    /// @param ext the extension to generate
     /// @returns true on success.
-    bool GenerateExtension(ast::Enable::ExtensionKind kind);
+    bool GenerateExtension(ast::Extension ext);
     /// Generates a label for the given id. Emits an error and returns false if
     /// we're currently outside a function.
     /// @param id the id to use for the label

src/tint/writer/wgsl/generator_impl.cc

@@ -62,12 +62,12 @@ GeneratorImpl::~GeneratorImpl() = default;
 
 bool GeneratorImpl::Generate() {
     // Generate enable directives before any other global declarations.
-    for (auto ext : program_->AST().Extensions()) {
-        if (!EmitEnableDirective(ext)) {
+    for (auto enable : program_->AST().Enables()) {
+        if (!EmitEnable(enable)) {
             return false;
         }
     }
-    if (!program_->AST().Extensions().empty()) {
+    if (!program_->AST().Enables().empty()) {
         line();
     }
     // Generate global declarations in the order they appear in the module.
@@ -94,13 +94,9 @@ bool GeneratorImpl::Generate() {
     return true;
 }
 
-bool GeneratorImpl::EmitEnableDirective(const ast::Enable::ExtensionKind ext) {
+bool GeneratorImpl::EmitEnable(const ast::Enable* enable) {
     auto out = line();
-    auto extension = ast::Enable::KindToName(ext);
-    if (extension == "") {
-        return false;
-    }
-    out << "enable " << extension << ";";
+    out << "enable " << enable->extension << ";";
     return true;
 }
 

src/tint/writer/wgsl/generator_impl.h

@@ -53,9 +53,9 @@ class GeneratorImpl : public TextGenerator {
     bool Generate();
 
     /// Handles generating a enable directive
-    /// @param ext the extension kind in the enable directive to generate
+    /// @param enable the enable node
     /// @returns true if the enable directive was emitted
-    bool EmitEnableDirective(const ast::Enable::ExtensionKind ext);
+    bool EmitEnable(const ast::Enable* enable);
     /// Handles generating a declared type
     /// @param ty the declared type to generate
     /// @returns true if the declared type was emitted

src/tint/writer/wgsl/generator_impl_enable_test.cc

@@ -20,10 +20,12 @@ namespace {
 using WgslGeneratorImplTest = TestHelper;
 
 TEST_F(WgslGeneratorImplTest, Emit_Enable) {
+    auto* enable = Enable(ast::Extension::kF16);
+
     GeneratorImpl& gen = Build();
 
-    ASSERT_TRUE(gen.EmitEnableDirective(ast::Enable::ExtensionKind::kInternalExtensionForTesting));
-    EXPECT_EQ(gen.result(), R"(enable InternalExtensionForTesting;
+    ASSERT_TRUE(gen.EmitEnable(enable));
+    EXPECT_EQ(gen.result(), R"(enable f16;
 )");
 }
 

test/tint/BUILD.gn

@@ -164,6 +164,7 @@ tint_unittests_source_set("tint_unittests_ast_src") {
     "../../src/tint/ast/depth_texture_test.cc",
     "../../src/tint/ast/discard_statement_test.cc",
     "../../src/tint/ast/enable_test.cc",
+    "../../src/tint/ast/extension_test.cc",
     "../../src/tint/ast/external_texture_test.cc",
     "../../src/tint/ast/f16_test.cc",
     "../../src/tint/ast/f32_test.cc",

test/tint/extensions/InternalExtensionForTesting/simple_with_InternalExtensionForTesting.wgsl

@@ -1,26 +0,0 @@
-// Copyright 2022 The Tint Authors.
-//
-// Licensed under the Apache License, Version 2.0 (the "License");
-// you may not use this file except in compliance with the License.
-// You may obtain a copy of the License at
-//
-//     http://www.apache.org/licenses/LICENSE-2.0
-//
-// Unless required by applicable law or agreed to in writing, software
-// distributed under the License is distributed on an "AS IS" BASIS,
-// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
-// See the License for the specific language governing permissions and
-// limitations under the License.
-
-// Enable a void internal extension
-enable InternalExtensionForTesting;
-
-fn bar() {
-}
-
-@stage(fragment)
-fn main() -> @location(0) vec4<f32> {
-    var a : vec2<f32> = vec2<f32>();
-    bar();
-    return vec4<f32>(0.4, 0.4, 0.8, 1.0);
-}

test/tint/extensions/InternalExtensionForTesting/simple_with_InternalExtensionForTesting.wgsl.expected.spvasm

@@ -1,47 +0,0 @@
-; SPIR-V
-; Version: 1.3
-; Generator: Google Tint Compiler; 0
-; Bound: 25
-; Schema: 0
-               OpCapability Shader
-               OpMemoryModel Logical GLSL450
-               OpEntryPoint Fragment %main "main" %value
-               OpExecutionMode %main OriginUpperLeft
-               OpName %value "value"
-               OpName %bar "bar"
-               OpName %main_inner "main_inner"
-               OpName %a "a"
-               OpName %main "main"
-               OpDecorate %value Location 0
-      %float = OpTypeFloat 32
-    %v4float = OpTypeVector %float 4
-%_ptr_Output_v4float = OpTypePointer Output %v4float
-          %5 = OpConstantNull %v4float
-      %value = OpVariable %_ptr_Output_v4float Output %5
-       %void = OpTypeVoid
-          %6 = OpTypeFunction %void
-         %10 = OpTypeFunction %v4float
-    %v2float = OpTypeVector %float 2
-         %14 = OpConstantNull %v2float
-%_ptr_Function_v2float = OpTypePointer Function %v2float
-%float_0_400000006 = OpConstant %float 0.400000006
-%float_0_800000012 = OpConstant %float 0.800000012
-    %float_1 = OpConstant %float 1
-         %21 = OpConstantComposite %v4float %float_0_400000006 %float_0_400000006 %float_0_800000012 %float_1
-        %bar = OpFunction %void None %6
-          %9 = OpLabel
-               OpReturn
-               OpFunctionEnd
- %main_inner = OpFunction %v4float None %10
-         %12 = OpLabel
-          %a = OpVariable %_ptr_Function_v2float Function %14
-               OpStore %a %14
-         %17 = OpFunctionCall %void %bar
-               OpReturnValue %21
-               OpFunctionEnd
-       %main = OpFunction %void None %6
-         %23 = OpLabel
-         %24 = OpFunctionCall %v4float %main_inner
-               OpStore %value %24
-               OpReturn
-               OpFunctionEnd

test/tint/extensions/InternalExtensionForTesting/simple_with_InternalExtensionForTesting.wgsl.expected.wgsl

@@ -1,11 +0,0 @@
-enable InternalExtensionForTesting;
-
-fn bar() {
-}
-
-@stage(fragment)
-fn main() -> @location(0) vec4<f32> {
-  var a : vec2<f32> = vec2<f32>();
-  bar();
-  return vec4<f32>(0.400000006, 0.400000006, 0.800000012, 1.0);
-}

test/tint/extensions/InternalExtensionForTesting/simple_with_duplicated_InternalExtensionForTesting.wgsl

@@ -1,28 +0,0 @@
-// Copyright 2022 The Tint Authors.
-//
-// Licensed under the Apache License, Version 2.0 (the "License");
-// you may not use this file except in compliance with the License.
-// You may obtain a copy of the License at
-//
-//     http://www.apache.org/licenses/LICENSE-2.0
-//
-// Unless required by applicable law or agreed to in writing, software
-// distributed under the License is distributed on an "AS IS" BASIS,
-// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
-// See the License for the specific language governing permissions and
-// limitations under the License.
-
-// Enable a void internal extension for multiple times
-enable InternalExtensionForTesting;
-enable InternalExtensionForTesting;
-enable InternalExtensionForTesting;
-
-fn bar() {
-}
-
-@stage(fragment)
-fn main() -> @location(0) vec4<f32> {
-    var a : vec2<f32> = vec2<f32>();
-    bar();
-    return vec4<f32>(0.4, 0.4, 0.8, 1.0);
-}

test/tint/extensions/InternalExtensionForTesting/simple_with_duplicated_InternalExtensionForTesting.wgsl.expected.spvasm

@@ -1,47 +0,0 @@
-; SPIR-V
-; Version: 1.3
-; Generator: Google Tint Compiler; 0
-; Bound: 25
-; Schema: 0
-               OpCapability Shader
-               OpMemoryModel Logical GLSL450
-               OpEntryPoint Fragment %main "main" %value
-               OpExecutionMode %main OriginUpperLeft
-               OpName %value "value"
-               OpName %bar "bar"
-               OpName %main_inner "main_inner"
-               OpName %a "a"
-               OpName %main "main"
-               OpDecorate %value Location 0
-      %float = OpTypeFloat 32
-    %v4float = OpTypeVector %float 4
-%_ptr_Output_v4float = OpTypePointer Output %v4float
-          %5 = OpConstantNull %v4float
-      %value = OpVariable %_ptr_Output_v4float Output %5
-       %void = OpTypeVoid
-          %6 = OpTypeFunction %void
-         %10 = OpTypeFunction %v4float
-    %v2float = OpTypeVector %float 2
-         %14 = OpConstantNull %v2float
-%_ptr_Function_v2float = OpTypePointer Function %v2float
-%float_0_400000006 = OpConstant %float 0.400000006
-%float_0_800000012 = OpConstant %float 0.800000012
-    %float_1 = OpConstant %float 1
-         %21 = OpConstantComposite %v4float %float_0_400000006 %float_0_400000006 %float_0_800000012 %float_1
-        %bar = OpFunction %void None %6
-          %9 = OpLabel
-               OpReturn
-               OpFunctionEnd
- %main_inner = OpFunction %v4float None %10
-         %12 = OpLabel
-          %a = OpVariable %_ptr_Function_v2float Function %14
-               OpStore %a %14
-         %17 = OpFunctionCall %void %bar
-               OpReturnValue %21
-               OpFunctionEnd
-       %main = OpFunction %void None %6
-         %23 = OpLabel
-         %24 = OpFunctionCall %v4float %main_inner
-               OpStore %value %24
-               OpReturn
-               OpFunctionEnd

test/tint/extensions/InternalExtensionForTesting/simple_with_duplicated_InternalExtensionForTesting.wgsl.expected.wgsl

@@ -1,11 +0,0 @@
-enable InternalExtensionForTesting;
-
-fn bar() {
-}
-
-@stage(fragment)
-fn main() -> @location(0) vec4<f32> {
-  var a : vec2<f32> = vec2<f32>();
-  bar();
-  return vec4<f32>(0.400000006, 0.400000006, 0.800000012, 1.0);
-}

test/tint/extensions/parsing/basic.wgsl

@@ -0,0 +1,7 @@
+// Enable a void internal extension
+enable f16;
+
+@stage(fragment)
+fn main() -> @location(0) vec4<f32> {
+    return vec4<f32>(0.1, 0.2, 0.3, 0.4);
+}

test/tint/extensions/parsing/basic.wgsl.expected.glsl

@@ -2,13 +2,8 @@
 precision mediump float;
 
 layout(location = 0) out vec4 value;
-void bar() {
-}
-
 vec4 tint_symbol() {
-  vec2 a = vec2(0.0f, 0.0f);
-  bar();
-  return vec4(0.400000006f, 0.400000006f, 0.800000012f, 1.0f);
+  return vec4(0.100000001f, 0.200000003f, 0.300000012f, 0.400000006f);
 }
 
 void main() {

test/tint/extensions/parsing/basic.wgsl.expected.hlsl

@@ -1,14 +1,9 @@
-void bar() {
-}
-
 struct tint_symbol {
   float4 value : SV_Target0;
 };
 
 float4 main_inner() {
-  float2 a = float2(0.0f, 0.0f);
-  bar();
-  return float4(0.400000006f, 0.400000006f, 0.800000012f, 1.0f);
+  return float4(0.100000001f, 0.200000003f, 0.300000012f, 0.400000006f);
 }
 
 tint_symbol main() {

test/tint/extensions/parsing/basic.wgsl.expected.msl

@@ -1,17 +1,12 @@
 #include <metal_stdlib>
 
 using namespace metal;
-void bar() {
-}
-
 struct tint_symbol_1 {
   float4 value [[color(0)]];
 };
 
 float4 tint_symbol_inner() {
-  float2 a = float2();
-  bar();
-  return float4(0.400000006f, 0.400000006f, 0.800000012f, 1.0f);
+  return float4(0.100000001f, 0.200000003f, 0.300000012f, 0.400000006f);
 }
 
 fragment tint_symbol_1 tint_symbol() {

test/tint/extensions/parsing/basic.wgsl.expected.spvasm

@@ -0,0 +1,36 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 19
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Fragment %main "main" %value
+               OpExecutionMode %main OriginUpperLeft
+               OpName %value "value"
+               OpName %main_inner "main_inner"
+               OpName %main "main"
+               OpDecorate %value Location 0
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+          %6 = OpTypeFunction %v4float
+%float_0_100000001 = OpConstant %float 0.100000001
+%float_0_200000003 = OpConstant %float 0.200000003
+%float_0_300000012 = OpConstant %float 0.300000012
+%float_0_400000006 = OpConstant %float 0.400000006
+         %13 = OpConstantComposite %v4float %float_0_100000001 %float_0_200000003 %float_0_300000012 %float_0_400000006
+       %void = OpTypeVoid
+         %14 = OpTypeFunction %void
+ %main_inner = OpFunction %v4float None %6
+          %8 = OpLabel
+               OpReturnValue %13
+               OpFunctionEnd
+       %main = OpFunction %void None %14
+         %17 = OpLabel
+         %18 = OpFunctionCall %v4float %main_inner
+               OpStore %value %18
+               OpReturn
+               OpFunctionEnd

test/tint/extensions/parsing/basic.wgsl.expected.wgsl

@@ -0,0 +1,6 @@
+enable f16;
+
+@stage(fragment)
+fn main() -> @location(0) vec4<f32> {
+  return vec4<f32>(0.100000001, 0.200000003, 0.300000012, 0.400000006);
+}

test/tint/extensions/parsing/duplicated_extensions.wgsl

@@ -0,0 +1,9 @@
+// Enable a void internal extension for multiple times
+enable f16;
+enable f16;
+enable f16;
+
+@stage(fragment)
+fn main() -> @location(0) vec4<f32> {
+    return vec4<f32>(0.1, 0.2, 0.3, 0.4);
+}

test/tint/extensions/parsing/duplicated_extensions.wgsl.expected.glsl

@@ -2,13 +2,8 @@
 precision mediump float;
 
 layout(location = 0) out vec4 value;
-void bar() {
-}
-
 vec4 tint_symbol() {
-  vec2 a = vec2(0.0f, 0.0f);
-  bar();
-  return vec4(0.400000006f, 0.400000006f, 0.800000012f, 1.0f);
+  return vec4(0.100000001f, 0.200000003f, 0.300000012f, 0.400000006f);
 }
 
 void main() {

test/tint/extensions/parsing/duplicated_extensions.wgsl.expected.hlsl

@@ -1,14 +1,9 @@
-void bar() {
-}
-
 struct tint_symbol {
   float4 value : SV_Target0;
 };
 
 float4 main_inner() {
-  float2 a = float2(0.0f, 0.0f);
-  bar();
-  return float4(0.400000006f, 0.400000006f, 0.800000012f, 1.0f);
+  return float4(0.100000001f, 0.200000003f, 0.300000012f, 0.400000006f);
 }
 
 tint_symbol main() {

test/tint/extensions/parsing/duplicated_extensions.wgsl.expected.msl

@@ -1,17 +1,12 @@
 #include <metal_stdlib>
 
 using namespace metal;
-void bar() {
-}
-
 struct tint_symbol_1 {
   float4 value [[color(0)]];
 };
 
 float4 tint_symbol_inner() {
-  float2 a = float2();
-  bar();
-  return float4(0.400000006f, 0.400000006f, 0.800000012f, 1.0f);
+  return float4(0.100000001f, 0.200000003f, 0.300000012f, 0.400000006f);
 }
 
 fragment tint_symbol_1 tint_symbol() {

test/tint/extensions/parsing/duplicated_extensions.wgsl.expected.spvasm

@@ -0,0 +1,36 @@
+; SPIR-V
+; Version: 1.3
+; Generator: Google Tint Compiler; 0
+; Bound: 19
+; Schema: 0
+               OpCapability Shader
+               OpMemoryModel Logical GLSL450
+               OpEntryPoint Fragment %main "main" %value
+               OpExecutionMode %main OriginUpperLeft
+               OpName %value "value"
+               OpName %main_inner "main_inner"
+               OpName %main "main"
+               OpDecorate %value Location 0
+      %float = OpTypeFloat 32
+    %v4float = OpTypeVector %float 4
+%_ptr_Output_v4float = OpTypePointer Output %v4float
+          %5 = OpConstantNull %v4float
+      %value = OpVariable %_ptr_Output_v4float Output %5
+          %6 = OpTypeFunction %v4float
+%float_0_100000001 = OpConstant %float 0.100000001
+%float_0_200000003 = OpConstant %float 0.200000003
+%float_0_300000012 = OpConstant %float 0.300000012
+%float_0_400000006 = OpConstant %float 0.400000006
+         %13 = OpConstantComposite %v4float %float_0_100000001 %float_0_200000003 %float_0_300000012 %float_0_400000006
+       %void = OpTypeVoid
+         %14 = OpTypeFunction %void
+ %main_inner = OpFunction %v4float None %6
+          %8 = OpLabel
+               OpReturnValue %13
+               OpFunctionEnd
+       %main = OpFunction %void None %14
+         %17 = OpLabel
+         %18 = OpFunctionCall %v4float %main_inner
+               OpStore %value %18
+               OpReturn
+               OpFunctionEnd

test/tint/extensions/parsing/duplicated_extensions.wgsl.expected.wgsl

@@ -0,0 +1,8 @@
+enable f16;
+enable f16;
+enable f16;
+
+@stage(fragment)
+fn main() -> @location(0) vec4<f32> {
+  return vec4<f32>(0.100000001, 0.200000003, 0.300000012, 0.400000006);
+}