diff --git a/src/dawn/native/ShaderModule.cpp b/src/dawn/native/ShaderModule.cpp
index 4c81bdde64..c9871e8c77 100644
--- a/src/dawn/native/ShaderModule.cpp
+++ b/src/dawn/native/ShaderModule.cpp
@@ -806,9 +806,8 @@ ResultOrError<std::unique_ptr<EntryPointMetadata>> ReflectEntryPointUsingTint(
                         resource.binding, resource.bind_group);
     }
 
-    std::vector<tint::inspector::SamplerTexturePair> samplerTextureUses =
-        inspector->GetSamplerTextureUses(entryPoint.name);
-    metadata->samplerTexturePairs.reserve(samplerTextureUses.size());
+    auto samplerTextureUses = inspector->GetSamplerTextureUses(entryPoint.name);
+    metadata->samplerTexturePairs.reserve(samplerTextureUses.Length());
     std::transform(samplerTextureUses.begin(), samplerTextureUses.end(),
                    std::back_inserter(metadata->samplerTexturePairs),
                    [](const tint::inspector::SamplerTexturePair& pair) {
diff --git a/src/tint/ast/extension.h b/src/tint/ast/extension.h
index ce3d48add8..eea9becb72 100644
--- a/src/tint/ast/extension.h
+++ b/src/tint/ast/extension.h
@@ -50,7 +50,7 @@ std::ostream& operator<<(std::ostream& out, Extension value);
 Extension ParseExtension(std::string_view str);
 
 // A unique vector of extensions
-using Extensions = utils::UniqueVector<Extension>;
+using Extensions = utils::UniqueVector<Extension, 4>;
 
 }  // namespace tint::ast
 
diff --git a/src/tint/ast/extension.h.tmpl b/src/tint/ast/extension.h.tmpl
index 395aeecd95..29feff962a 100644
--- a/src/tint/ast/extension.h.tmpl
+++ b/src/tint/ast/extension.h.tmpl
@@ -25,7 +25,7 @@ namespace tint::ast {
 {{ Eval "DeclareEnum" $enum}}
 
 // A unique vector of extensions
-using Extensions = utils::UniqueVector<Extension>;
+using Extensions = utils::UniqueVector<Extension, 4>;
 
 }  // namespace tint::ast
 
diff --git a/src/tint/inspector/inspector.cc b/src/tint/inspector/inspector.cc
index e9c4d887e6..60734709d8 100644
--- a/src/tint/inspector/inspector.cc
+++ b/src/tint/inspector/inspector.cc
@@ -505,7 +505,7 @@ std::vector<ResourceBinding> Inspector::GetExternalTextureResourceBindings(
                                       ResourceBinding::ResourceType::kExternalTexture);
 }
 
-std::vector<sem::SamplerTexturePair> Inspector::GetSamplerTextureUses(
+utils::Vector<sem::SamplerTexturePair, 4> Inspector::GetSamplerTextureUses(
     const std::string& entry_point) {
     auto* func = FindEntryPointByName(entry_point);
     if (!func) {
@@ -570,7 +570,7 @@ uint32_t Inspector::GetWorkgroupStorageSize(const std::string& entry_point) {
 std::vector<std::string> Inspector::GetUsedExtensionNames() {
     auto& extensions = program_->Sem().Module()->Extensions();
     std::vector<std::string> out;
-    out.reserve(extensions.size());
+    out.reserve(extensions.Length());
     for (auto ext : extensions) {
         out.push_back(utils::ToString(ext));
     }
@@ -789,7 +789,7 @@ void Inspector::GenerateSamplerTargets() {
     }
 
     sampler_targets_ = std::make_unique<
-        std::unordered_map<std::string, utils::UniqueVector<sem::SamplerTexturePair>>>();
+        std::unordered_map<std::string, utils::UniqueVector<sem::SamplerTexturePair, 4>>>();
 
     auto& sem = program_->Sem();
 
@@ -849,7 +849,7 @@ void Inspector::GenerateSamplerTargets() {
                 for (auto* entry_point : entry_points) {
                     const auto& ep_name =
                         program_->Symbols().NameFor(entry_point->Declaration()->symbol);
-                    (*sampler_targets_)[ep_name].add(
+                    (*sampler_targets_)[ep_name].Add(
                         {sampler_binding_point, texture_binding_point});
                 }
             });
@@ -868,7 +868,7 @@ void Inspector::GetOriginatingResources(std::array<const ast::Expression*, N> ex
 
     std::array<const sem::GlobalVariable*, N> globals{};
     std::array<const sem::Parameter*, N> parameters{};
-    utils::UniqueVector<const ast::CallExpression*> callsites;
+    utils::UniqueVector<const ast::CallExpression*, 8> callsites;
 
     for (size_t i = 0; i < N; i++) {
         const sem::Variable* source_var = sem.Get(exprs[i])->SourceVariable();
@@ -882,7 +882,7 @@ void Inspector::GetOriginatingResources(std::array<const ast::Expression*, N> ex
                 return;
             }
             for (auto* call : func->CallSites()) {
-                callsites.add(call->Declaration());
+                callsites.Add(call->Declaration());
             }
             parameters[i] = param;
         } else {
@@ -893,7 +893,7 @@ void Inspector::GetOriginatingResources(std::array<const ast::Expression*, N> ex
         }
     }
 
-    if (callsites.size()) {
+    if (callsites.Length()) {
         for (auto* call_expr : callsites) {
             // Make a copy of the expressions for this callsite
             std::array<const ast::Expression*, N> call_exprs = exprs;
diff --git a/src/tint/inspector/inspector.h b/src/tint/inspector/inspector.h
index 97707db3e9..f3fe27008a 100644
--- a/src/tint/inspector/inspector.h
+++ b/src/tint/inspector/inspector.h
@@ -122,7 +122,7 @@ class Inspector {
     /// @param entry_point name of the entry point to get information about.
     /// @returns vector of all of the sampler/texture sampling pairs that are used
     /// by that entry point.
-    std::vector<sem::SamplerTexturePair> GetSamplerTextureUses(const std::string& entry_point);
+    utils::Vector<sem::SamplerTexturePair, 4> GetSamplerTextureUses(const std::string& entry_point);
 
     /// @param entry_point name of the entry point to get information about.
     /// @param placeholder the sampler binding point to use for texture-only
@@ -153,7 +153,8 @@ class Inspector {
   private:
     const Program* program_;
     diag::List diagnostics_;
-    std::unique_ptr<std::unordered_map<std::string, utils::UniqueVector<sem::SamplerTexturePair>>>
+    std::unique_ptr<
+        std::unordered_map<std::string, utils::UniqueVector<sem::SamplerTexturePair, 4>>>
         sampler_targets_;
 
     /// @param name name of the entry point to find
diff --git a/src/tint/inspector/inspector_test.cc b/src/tint/inspector/inspector_test.cc
index 543239c5b1..cb46833b94 100644
--- a/src/tint/inspector/inspector_test.cc
+++ b/src/tint/inspector/inspector_test.cc
@@ -2971,7 +2971,7 @@ fn main() {
     auto result = inspector.GetSamplerTextureUses("main");
     ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-    ASSERT_EQ(0u, result.size());
+    ASSERT_EQ(0u, result.Length());
 }
 
 TEST_F(InspectorGetSamplerTextureUsesTest, Simple) {
@@ -2989,7 +2989,7 @@ fn main(@location(0) fragUV: vec2<f32>,
     auto result = inspector.GetSamplerTextureUses("main");
     ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-    ASSERT_EQ(1u, result.size());
+    ASSERT_EQ(1u, result.Length());
 
     EXPECT_EQ(0u, result[0].sampler_binding_point.group);
     EXPECT_EQ(1u, result[0].sampler_binding_point.binding);
@@ -3053,7 +3053,7 @@ fn main(@location(0) fragUV: vec2<f32>,
     auto result = inspector.GetSamplerTextureUses("main");
     ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-    ASSERT_EQ(1u, result.size());
+    ASSERT_EQ(1u, result.Length());
 
     EXPECT_EQ(0u, result[0].sampler_binding_point.group);
     EXPECT_EQ(1u, result[0].sampler_binding_point.binding);
@@ -3080,7 +3080,7 @@ fn main(@location(0) fragUV: vec2<f32>,
     auto result = inspector.GetSamplerTextureUses("main");
     ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-    ASSERT_EQ(1u, result.size());
+    ASSERT_EQ(1u, result.Length());
 
     EXPECT_EQ(0u, result[0].sampler_binding_point.group);
     EXPECT_EQ(1u, result[0].sampler_binding_point.binding);
@@ -3107,7 +3107,7 @@ fn main(@location(0) fragUV: vec2<f32>,
     auto result = inspector.GetSamplerTextureUses("main");
     ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-    ASSERT_EQ(1u, result.size());
+    ASSERT_EQ(1u, result.Length());
 
     EXPECT_EQ(0u, result[0].sampler_binding_point.group);
     EXPECT_EQ(1u, result[0].sampler_binding_point.binding);
@@ -3134,7 +3134,7 @@ fn main(@location(0) fragUV: vec2<f32>,
     auto result = inspector.GetSamplerTextureUses("main");
     ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-    ASSERT_EQ(1u, result.size());
+    ASSERT_EQ(1u, result.Length());
 
     EXPECT_EQ(0u, result[0].sampler_binding_point.group);
     EXPECT_EQ(1u, result[0].sampler_binding_point.binding);
@@ -3188,7 +3188,7 @@ fn direct(@location(0) fragUV: vec2<f32>,
         auto result = inspector.GetSamplerTextureUses("via_call");
         ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-        ASSERT_EQ(1u, result.size());
+        ASSERT_EQ(1u, result.Length());
 
         EXPECT_EQ(0u, result[0].sampler_binding_point.group);
         EXPECT_EQ(1u, result[0].sampler_binding_point.binding);
@@ -3200,7 +3200,7 @@ fn direct(@location(0) fragUV: vec2<f32>,
         auto result = inspector.GetSamplerTextureUses("via_ptr");
         ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-        ASSERT_EQ(1u, result.size());
+        ASSERT_EQ(1u, result.Length());
 
         EXPECT_EQ(0u, result[0].sampler_binding_point.group);
         EXPECT_EQ(1u, result[0].sampler_binding_point.binding);
@@ -3212,7 +3212,7 @@ fn direct(@location(0) fragUV: vec2<f32>,
         auto result = inspector.GetSamplerTextureUses("direct");
         ASSERT_FALSE(inspector.has_error()) << inspector.error();
 
-        ASSERT_EQ(1u, result.size());
+        ASSERT_EQ(1u, result.Length());
 
         EXPECT_EQ(0u, result[0].sampler_binding_point.group);
         EXPECT_EQ(1u, result[0].sampler_binding_point.binding);
diff --git a/src/tint/reader/spirv/entry_point_info.cc b/src/tint/reader/spirv/entry_point_info.cc
index 2677494a12..9e3604afe5 100644
--- a/src/tint/reader/spirv/entry_point_info.cc
+++ b/src/tint/reader/spirv/entry_point_info.cc
@@ -22,8 +22,8 @@ EntryPointInfo::EntryPointInfo(std::string the_name,
                                ast::PipelineStage the_stage,
                                bool the_owns_inner_implementation,
                                std::string the_inner_name,
-                               std::vector<uint32_t>&& the_inputs,
-                               std::vector<uint32_t>&& the_outputs,
+                               utils::VectorRef<uint32_t> the_inputs,
+                               utils::VectorRef<uint32_t> the_outputs,
                                GridSize the_wg_size)
     : name(the_name),
       stage(the_stage),
diff --git a/src/tint/reader/spirv/entry_point_info.h b/src/tint/reader/spirv/entry_point_info.h
index bc13759f00..90077421d1 100644
--- a/src/tint/reader/spirv/entry_point_info.h
+++ b/src/tint/reader/spirv/entry_point_info.h
@@ -16,9 +16,9 @@
 #define SRC_TINT_READER_SPIRV_ENTRY_POINT_INFO_H_
 
 #include <string>
-#include <vector>
 
 #include "src/tint/ast/pipeline_stage.h"
+#include "src/tint/utils/vector.h"
 
 namespace tint::reader::spirv {
 
@@ -48,8 +48,8 @@ struct EntryPointInfo {
                    ast::PipelineStage the_stage,
                    bool the_owns_inner_implementation,
                    std::string the_inner_name,
-                   std::vector<uint32_t>&& the_inputs,
-                   std::vector<uint32_t>&& the_outputs,
+                   utils::VectorRef<uint32_t> the_inputs,
+                   utils::VectorRef<uint32_t> the_outputs,
                    GridSize the_wg_size);
     /// Copy constructor
     /// @param other the other entry point info to be built from
@@ -75,9 +75,9 @@ struct EntryPointInfo {
     /// The name of the inner implementation function of the entry point.
     std::string inner_name;
     /// IDs of pipeline input variables, sorted and without duplicates.
-    std::vector<uint32_t> inputs;
+    utils::Vector<uint32_t, 8> inputs;
     /// IDs of pipeline output variables, sorted and without duplicates.
-    std::vector<uint32_t> outputs;
+    utils::Vector<uint32_t, 8> outputs;
 
     /// If this is a compute shader, this is the workgroup size in the x, y,
     /// and z dimensions set via LocalSize, or via the composite value
diff --git a/src/tint/reader/spirv/function.cc b/src/tint/reader/spirv/function.cc
index 49753870f8..a57984b820 100644
--- a/src/tint/reader/spirv/function.cc
+++ b/src/tint/reader/spirv/function.cc
@@ -1303,7 +1303,7 @@ bool FunctionEmitter::EmitEntryPointAsWrapper() {
                                     utils::Empty)));
 
     // Pipeline outputs are mapped to the return value.
-    if (ep_info_->outputs.empty()) {
+    if (ep_info_->outputs.IsEmpty()) {
         // There is nothing to return.
         return_type = ty_.Void()->Build(builder_);
     } else {
diff --git a/src/tint/reader/spirv/parser_impl.cc b/src/tint/reader/spirv/parser_impl.cc
index e1bff314c6..f9117a5e49 100644
--- a/src/tint/reader/spirv/parser_impl.cc
+++ b/src/tint/reader/spirv/parser_impl.cc
@@ -878,17 +878,17 @@ bool ParserImpl::RegisterEntryPoints() {
         TINT_ASSERT(Reader, !inner_implementation_name.empty());
         TINT_ASSERT(Reader, ep_name != inner_implementation_name);
 
-        utils::UniqueVector<uint32_t> inputs;
-        utils::UniqueVector<uint32_t> outputs;
+        utils::UniqueVector<uint32_t, 8> inputs;
+        utils::UniqueVector<uint32_t, 8> outputs;
         for (unsigned iarg = 3; iarg < entry_point.NumInOperands(); iarg++) {
             const uint32_t var_id = entry_point.GetSingleWordInOperand(iarg);
             if (const auto* var_inst = def_use_mgr_->GetDef(var_id)) {
                 switch (SpvStorageClass(var_inst->GetSingleWordInOperand(0))) {
                     case SpvStorageClassInput:
-                        inputs.add(var_id);
+                        inputs.Add(var_id);
                         break;
                     case SpvStorageClassOutput:
-                        outputs.add(var_id);
+                        outputs.Add(var_id);
                         break;
                     default:
                         break;
@@ -896,9 +896,9 @@ bool ParserImpl::RegisterEntryPoints() {
             }
         }
         // Save the lists, in ID-sorted order.
-        std::vector<uint32_t> sorted_inputs(inputs);
+        utils::Vector<uint32_t, 8> sorted_inputs(inputs);
         std::sort(sorted_inputs.begin(), sorted_inputs.end());
-        std::vector<uint32_t> sorted_outputs(outputs);
+        utils::Vector<uint32_t, 8> sorted_outputs(outputs);
         std::sort(sorted_outputs.begin(), sorted_outputs.end());
 
         const auto ast_stage = enum_converter_.ToPipelineStage(stage);
diff --git a/src/tint/reader/spirv/parser_impl_module_var_test.cc b/src/tint/reader/spirv/parser_impl_module_var_test.cc
index c88f0d600e..8eb506b8bd 100644
--- a/src/tint/reader/spirv/parser_impl_module_var_test.cc
+++ b/src/tint/reader/spirv/parser_impl_module_var_test.cc
@@ -3428,13 +3428,13 @@ TEST_F(SpvModuleScopeVarParserTest, RegisterInputOutputVars) {
 
     const auto& info_1000 = p->GetEntryPointInfo(1000);
     EXPECT_EQ(1u, info_1000.size());
-    EXPECT_TRUE(info_1000[0].inputs.empty());
-    EXPECT_TRUE(info_1000[0].outputs.empty());
+    EXPECT_TRUE(info_1000[0].inputs.IsEmpty());
+    EXPECT_TRUE(info_1000[0].outputs.IsEmpty());
 
     const auto& info_1100 = p->GetEntryPointInfo(1100);
     EXPECT_EQ(1u, info_1100.size());
     EXPECT_THAT(info_1100[0].inputs, ElementsAre(1));
-    EXPECT_TRUE(info_1100[0].outputs.empty());
+    EXPECT_TRUE(info_1100[0].outputs.IsEmpty());
 
     const auto& info_1200 = p->GetEntryPointInfo(1200);
     EXPECT_EQ(1u, info_1200.size());
diff --git a/src/tint/resolver/dependency_graph.cc b/src/tint/resolver/dependency_graph.cc
index 6dd35a50ec..823dcada11 100644
--- a/src/tint/resolver/dependency_graph.cc
+++ b/src/tint/resolver/dependency_graph.cc
@@ -491,7 +491,7 @@ struct DependencyAnalysis {
     bool Run(const ast::Module& module) {
         // Reserve container memory
         graph_.resolved_symbols.reserve(module.GlobalDeclarations().Length());
-        sorted_.reserve(module.GlobalDeclarations().Length());
+        sorted_.Reserve(module.GlobalDeclarations().Length());
 
         // Collect all the named globals from the AST module
         GatherGlobals(module);
@@ -505,7 +505,7 @@ struct DependencyAnalysis {
         // Dump the dependency graph if TINT_DUMP_DEPENDENCY_GRAPH is non-zero
         DumpDependencyGraph();
 
-        graph_.ordered_globals = std::move(sorted_);
+        graph_.ordered_globals = sorted_.Release();
 
         return !diagnostics_.contains_errors();
     }
@@ -632,7 +632,7 @@ struct DependencyAnalysis {
         // Make sure all 'enable' directives go before any other global declarations.
         for (auto* global : declaration_order_) {
             if (auto* enable = global->node->As<ast::Enable>()) {
-                sorted_.add(enable);
+                sorted_.Add(enable);
             }
         }
 
@@ -641,31 +641,31 @@ struct DependencyAnalysis {
                 // Skip 'enable' directives here, as they are already added.
                 continue;
             }
-            utils::UniqueVector<const Global*> stack;
+            utils::UniqueVector<const Global*, 8> stack;
             TraverseDependencies(
                 global,
                 [&](const Global* g) {  // Enter
-                    if (!stack.add(g)) {
-                        CyclicDependencyFound(g, stack);
+                    if (!stack.Add(g)) {
+                        CyclicDependencyFound(g, stack.Release());
                         return false;
                     }
-                    if (sorted_.contains(g->node)) {
+                    if (sorted_.Contains(g->node)) {
                         // Visited this global already.
                         // stack was pushed, but exit() will not be called when we return
                         // false, so pop here.
-                        stack.pop_back();
+                        stack.Pop();
                         return false;
                     }
                     return true;
                 },
                 [&](const Global* g) {  // Exit. Only called if Enter returned true.
-                    sorted_.add(g->node);
-                    stack.pop_back();
+                    sorted_.Add(g->node);
+                    stack.Pop();
                 });
 
-            sorted_.add(global->node);
+            sorted_.Add(global->node);
 
-            if (!stack.empty()) {
+            if (!stack.IsEmpty()) {
                 // Each stack.push() must have a corresponding stack.pop_back().
                 TINT_ICE(Resolver, diagnostics_)
                     << "stack not empty after returning from TraverseDependencies()";
@@ -691,12 +691,12 @@ struct DependencyAnalysis {
     /// @param root is the global that starts the cyclic dependency, which must be
     /// found in `stack`.
     /// @param stack is the global dependency stack that contains a loop.
-    void CyclicDependencyFound(const Global* root, const std::vector<const Global*>& stack) {
+    void CyclicDependencyFound(const Global* root, utils::VectorRef<const Global*> stack) {
         std::stringstream msg;
         msg << "cyclic dependency found: ";
         constexpr size_t kLoopNotStarted = ~0u;
         size_t loop_start = kLoopNotStarted;
-        for (size_t i = 0; i < stack.size(); i++) {
+        for (size_t i = 0; i < stack.Length(); i++) {
             auto* e = stack[i];
             if (loop_start == kLoopNotStarted && e == root) {
                 loop_start = i;
@@ -707,9 +707,9 @@ struct DependencyAnalysis {
         }
         msg << "'" << NameOf(root->node) << "'";
         AddError(diagnostics_, msg.str(), root->node->source);
-        for (size_t i = loop_start; i < stack.size(); i++) {
+        for (size_t i = loop_start; i < stack.Length(); i++) {
             auto* from = stack[i];
-            auto* to = (i + 1 < stack.size()) ? stack[i + 1] : stack[loop_start];
+            auto* to = (i + 1 < stack.Length()) ? stack[i + 1] : stack[loop_start];
             auto info = DepInfoFor(from, to);
             AddNote(diagnostics_,
                     KindOf(from->node) + " '" + NameOf(from->node) + "' " + info.action + " " +
@@ -764,7 +764,7 @@ struct DependencyAnalysis {
     std::vector<Global*> declaration_order_;
 
     /// Globals in sorted dependency order. Populated by SortGlobals().
-    utils::UniqueVector<const ast::Node*> sorted_;
+    utils::UniqueVector<const ast::Node*, 64> sorted_;
 };
 
 }  // namespace
diff --git a/src/tint/resolver/dependency_graph.h b/src/tint/resolver/dependency_graph.h
index 0554817ebc..9f5ddc5646 100644
--- a/src/tint/resolver/dependency_graph.h
+++ b/src/tint/resolver/dependency_graph.h
@@ -46,7 +46,7 @@ struct DependencyGraph {
                       DependencyGraph& output);
 
     /// All globals in dependency-sorted order.
-    std::vector<const ast::Node*> ordered_globals;
+    utils::Vector<const ast::Node*, 32> ordered_globals;
 
     /// Map of ast::IdentifierExpression or ast::TypeName to a type, function, or
     /// variable that declares the symbol.
diff --git a/src/tint/resolver/function_validation_test.cc b/src/tint/resolver/function_validation_test.cc
index 372d91ad8e..3c4a6ac496 100644
--- a/src/tint/resolver/function_validation_test.cc
+++ b/src/tint/resolver/function_validation_test.cc
@@ -489,8 +489,8 @@ TEST_F(ResolverFunctionValidationTest, WorkgroupSize_GoodType_ConstU32) {
     ASSERT_NE(sem_x, nullptr);
     ASSERT_NE(sem_y, nullptr);
 
-    EXPECT_TRUE(sem_func->DirectlyReferencedGlobals().contains(sem_x));
-    EXPECT_TRUE(sem_func->DirectlyReferencedGlobals().contains(sem_y));
+    EXPECT_TRUE(sem_func->DirectlyReferencedGlobals().Contains(sem_x));
+    EXPECT_TRUE(sem_func->DirectlyReferencedGlobals().Contains(sem_y));
 }
 
 TEST_F(ResolverFunctionValidationTest, WorkgroupSize_GoodType_I32) {
diff --git a/src/tint/resolver/resolver.cc b/src/tint/resolver/resolver.cc
index 2edab1f650..cd8bdc8175 100644
--- a/src/tint/resolver/resolver.cc
+++ b/src/tint/resolver/resolver.cc
@@ -164,7 +164,7 @@ bool Resolver::ResolveInternal() {
         return false;
     }
 
-    if (!enabled_extensions_.contains(ast::Extension::kChromiumDisableUniformityAnalysis)) {
+    if (!enabled_extensions_.Contains(ast::Extension::kChromiumDisableUniformityAnalysis)) {
         if (!AnalyzeUniformity(builder_, dependencies_)) {
             // TODO(jrprice): Reject programs that fail uniformity analysis.
         }
@@ -194,7 +194,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 (!enabled_extensions_.contains(ast::Extension::kF16)) {
+            if (!enabled_extensions_.Contains(ast::Extension::kF16)) {
                 AddError("f16 used without 'f16' extension enabled", t->source);
                 return nullptr;
             }
@@ -2082,7 +2082,7 @@ sem::Expression* Resolver::Literal(const ast::LiteralExpression* literal) {
         return nullptr;
     }
 
-    if ((ty->Is<sem::F16>()) && (!enabled_extensions_.contains(tint::ast::Extension::kF16))) {
+    if ((ty->Is<sem::F16>()) && (!enabled_extensions_.Contains(tint::ast::Extension::kF16))) {
         AddError("f16 literal used without 'f16' extension enabled", literal->source);
         return nullptr;
     }
@@ -2442,7 +2442,7 @@ sem::Expression* Resolver::UnaryOp(const ast::UnaryOpExpression* unary) {
 }
 
 bool Resolver::Enable(const ast::Enable* enable) {
-    enabled_extensions_.add(enable->extension);
+    enabled_extensions_.Add(enable->extension);
     return true;
 }
 
diff --git a/src/tint/resolver/resolver_test.cc b/src/tint/resolver/resolver_test.cc
index ecc251a785..35ebcf5bae 100644
--- a/src/tint/resolver/resolver_test.cc
+++ b/src/tint/resolver/resolver_test.cc
@@ -800,7 +800,7 @@ TEST_F(ResolverTest, Function_RegisterInputOutputVariables) {
     EXPECT_TRUE(func_sem->ReturnType()->Is<sem::Void>());
 
     const auto& vars = func_sem->TransitivelyReferencedGlobals();
-    ASSERT_EQ(vars.size(), 3u);
+    ASSERT_EQ(vars.Length(), 3u);
     EXPECT_EQ(vars[0]->Declaration(), wg_var);
     EXPECT_EQ(vars[1]->Declaration(), sb_var);
     EXPECT_EQ(vars[2]->Declaration(), priv_var);
@@ -835,7 +835,7 @@ TEST_F(ResolverTest, Function_RegisterInputOutputVariables_SubFunction) {
     EXPECT_EQ(func2_sem->Parameters().Length(), 0u);
 
     const auto& vars = func2_sem->TransitivelyReferencedGlobals();
-    ASSERT_EQ(vars.size(), 3u);
+    ASSERT_EQ(vars.Length(), 3u);
     EXPECT_EQ(vars[0]->Declaration(), wg_var);
     EXPECT_EQ(vars[1]->Declaration(), sb_var);
     EXPECT_EQ(vars[2]->Declaration(), priv_var);
@@ -853,7 +853,7 @@ TEST_F(ResolverTest, Function_NotRegisterFunctionVariable) {
     auto* func_sem = Sem().Get(func);
     ASSERT_NE(func_sem, nullptr);
 
-    EXPECT_EQ(func_sem->TransitivelyReferencedGlobals().size(), 0u);
+    EXPECT_EQ(func_sem->TransitivelyReferencedGlobals().Length(), 0u);
     EXPECT_TRUE(func_sem->ReturnType()->Is<sem::Void>());
 }
 
@@ -868,7 +868,7 @@ TEST_F(ResolverTest, Function_NotRegisterFunctionConstant) {
     auto* func_sem = Sem().Get(func);
     ASSERT_NE(func_sem, nullptr);
 
-    EXPECT_EQ(func_sem->TransitivelyReferencedGlobals().size(), 0u);
+    EXPECT_EQ(func_sem->TransitivelyReferencedGlobals().Length(), 0u);
     EXPECT_TRUE(func_sem->ReturnType()->Is<sem::Void>());
 }
 
@@ -879,7 +879,7 @@ TEST_F(ResolverTest, Function_NotRegisterFunctionParams) {
     auto* func_sem = Sem().Get(func);
     ASSERT_NE(func_sem, nullptr);
 
-    EXPECT_EQ(func_sem->TransitivelyReferencedGlobals().size(), 0u);
+    EXPECT_EQ(func_sem->TransitivelyReferencedGlobals().Length(), 0u);
     EXPECT_TRUE(func_sem->ReturnType()->Is<sem::Void>());
 }
 
@@ -2006,7 +2006,7 @@ TEST_F(ResolverTest, TextureSampler_TextureSample) {
 
     const sem::Function* sf = Sem().Get(f);
     auto pairs = sf->TextureSamplerPairs();
-    ASSERT_EQ(pairs.size(), 1u);
+    ASSERT_EQ(pairs.Length(), 1u);
     EXPECT_TRUE(pairs[0].first != nullptr);
     EXPECT_TRUE(pairs[0].second != nullptr);
 }
@@ -2026,12 +2026,12 @@ TEST_F(ResolverTest, TextureSampler_TextureSampleInFunction) {
     EXPECT_TRUE(r()->Resolve()) << r()->error();
 
     auto inner_pairs = Sem().Get(inner_func)->TextureSamplerPairs();
-    ASSERT_EQ(inner_pairs.size(), 1u);
+    ASSERT_EQ(inner_pairs.Length(), 1u);
     EXPECT_TRUE(inner_pairs[0].first != nullptr);
     EXPECT_TRUE(inner_pairs[0].second != nullptr);
 
     auto outer_pairs = Sem().Get(outer_func)->TextureSamplerPairs();
-    ASSERT_EQ(outer_pairs.size(), 1u);
+    ASSERT_EQ(outer_pairs.Length(), 1u);
     EXPECT_TRUE(outer_pairs[0].first != nullptr);
     EXPECT_TRUE(outer_pairs[0].second != nullptr);
 }
@@ -2055,17 +2055,17 @@ TEST_F(ResolverTest, TextureSampler_TextureSampleFunctionDiamondSameVariables) {
     EXPECT_TRUE(r()->Resolve()) << r()->error();
 
     auto inner_pairs_1 = Sem().Get(inner_func_1)->TextureSamplerPairs();
-    ASSERT_EQ(inner_pairs_1.size(), 1u);
+    ASSERT_EQ(inner_pairs_1.Length(), 1u);
     EXPECT_TRUE(inner_pairs_1[0].first != nullptr);
     EXPECT_TRUE(inner_pairs_1[0].second != nullptr);
 
     auto inner_pairs_2 = Sem().Get(inner_func_2)->TextureSamplerPairs();
-    ASSERT_EQ(inner_pairs_1.size(), 1u);
+    ASSERT_EQ(inner_pairs_1.Length(), 1u);
     EXPECT_TRUE(inner_pairs_2[0].first != nullptr);
     EXPECT_TRUE(inner_pairs_2[0].second != nullptr);
 
     auto outer_pairs = Sem().Get(outer_func)->TextureSamplerPairs();
-    ASSERT_EQ(outer_pairs.size(), 1u);
+    ASSERT_EQ(outer_pairs.Length(), 1u);
     EXPECT_TRUE(outer_pairs[0].first != nullptr);
     EXPECT_TRUE(outer_pairs[0].second != nullptr);
 }
@@ -2092,17 +2092,17 @@ TEST_F(ResolverTest, TextureSampler_TextureSampleFunctionDiamondDifferentVariabl
     EXPECT_TRUE(r()->Resolve()) << r()->error();
 
     auto inner_pairs_1 = Sem().Get(inner_func_1)->TextureSamplerPairs();
-    ASSERT_EQ(inner_pairs_1.size(), 1u);
+    ASSERT_EQ(inner_pairs_1.Length(), 1u);
     EXPECT_TRUE(inner_pairs_1[0].first != nullptr);
     EXPECT_TRUE(inner_pairs_1[0].second != nullptr);
 
     auto inner_pairs_2 = Sem().Get(inner_func_2)->TextureSamplerPairs();
-    ASSERT_EQ(inner_pairs_2.size(), 1u);
+    ASSERT_EQ(inner_pairs_2.Length(), 1u);
     EXPECT_TRUE(inner_pairs_2[0].first != nullptr);
     EXPECT_TRUE(inner_pairs_2[0].second != nullptr);
 
     auto outer_pairs = Sem().Get(outer_func)->TextureSamplerPairs();
-    ASSERT_EQ(outer_pairs.size(), 2u);
+    ASSERT_EQ(outer_pairs.Length(), 2u);
     EXPECT_TRUE(outer_pairs[0].first == inner_pairs_1[0].first);
     EXPECT_TRUE(outer_pairs[0].second == inner_pairs_1[0].second);
     EXPECT_TRUE(outer_pairs[1].first == inner_pairs_2[0].first);
@@ -2119,7 +2119,7 @@ TEST_F(ResolverTest, TextureSampler_TextureDimensions) {
 
     const sem::Function* sf = Sem().Get(f);
     auto pairs = sf->TextureSamplerPairs();
-    ASSERT_EQ(pairs.size(), 1u);
+    ASSERT_EQ(pairs.Length(), 1u);
     EXPECT_TRUE(pairs[0].first != nullptr);
     EXPECT_TRUE(pairs[0].second == nullptr);
 }
diff --git a/src/tint/resolver/uniformity.cc b/src/tint/resolver/uniformity.cc
index 9ccb7ef856..e4c31c9394 100644
--- a/src/tint/resolver/uniformity.cc
+++ b/src/tint/resolver/uniformity.cc
@@ -102,14 +102,14 @@ struct Node {
     uint32_t arg_index;
 
     /// The set of edges from this node to other nodes in the graph.
-    utils::UniqueVector<Node*> edges;
+    utils::UniqueVector<Node*, 4> edges;
 
     /// The node that this node was visited from, or nullptr if not visited.
     Node* visited_from = nullptr;
 
     /// Add an edge to the `to` node.
     /// @param to the destination node
-    void AddEdge(Node* to) { edges.add(to); }
+    void AddEdge(Node* to) { edges.Add(to); }
 };
 
 /// ParameterInfo holds information about the uniformity requirements and effects for a particular
@@ -337,13 +337,13 @@ class UniformityGraph {
 
         // Look at which nodes are reachable from "RequiredToBeUniform".
         {
-            utils::UniqueVector<Node*> reachable;
+            utils::UniqueVector<Node*, 4> reachable;
             Traverse(current_function_->required_to_be_uniform, &reachable);
-            if (reachable.contains(current_function_->may_be_non_uniform)) {
+            if (reachable.Contains(current_function_->may_be_non_uniform)) {
                 MakeError(*current_function_, current_function_->may_be_non_uniform);
                 return false;
             }
-            if (reachable.contains(current_function_->cf_start)) {
+            if (reachable.Contains(current_function_->cf_start)) {
                 current_function_->callsite_tag = CallSiteRequiredToBeUniform;
             }
 
@@ -351,7 +351,7 @@ class UniformityGraph {
             // was reachable.
             for (size_t i = 0; i < func->params.Length(); i++) {
                 auto* param = func->params[i];
-                if (reachable.contains(current_function_->variables.Get(sem_.Get(param)))) {
+                if (reachable.Contains(current_function_->variables.Get(sem_.Get(param)))) {
                     current_function_->parameters[i].tag = ParameterRequiredToBeUniform;
                 }
             }
@@ -359,9 +359,9 @@ class UniformityGraph {
 
         // Look at which nodes are reachable from "CF_return"
         {
-            utils::UniqueVector<Node*> reachable;
+            utils::UniqueVector<Node*, 4> reachable;
             Traverse(current_function_->cf_return, &reachable);
-            if (reachable.contains(current_function_->may_be_non_uniform)) {
+            if (reachable.Contains(current_function_->may_be_non_uniform)) {
                 current_function_->function_tag = SubsequentControlFlowMayBeNonUniform;
             }
 
@@ -369,7 +369,7 @@ class UniformityGraph {
             // each parameter node that was reachable.
             for (size_t i = 0; i < func->params.Length(); i++) {
                 auto* param = func->params[i];
-                if (reachable.contains(current_function_->variables.Get(sem_.Get(param)))) {
+                if (reachable.Contains(current_function_->variables.Get(sem_.Get(param)))) {
                     current_function_->parameters[i].tag =
                         ParameterRequiredToBeUniformForSubsequentControlFlow;
                 }
@@ -378,9 +378,9 @@ class UniformityGraph {
 
         // If "Value_return" exists, look at which nodes are reachable from it
         if (current_function_->value_return) {
-            utils::UniqueVector<Node*> reachable;
+            utils::UniqueVector<Node*, 4> reachable;
             Traverse(current_function_->value_return, &reachable);
-            if (reachable.contains(current_function_->may_be_non_uniform)) {
+            if (reachable.Contains(current_function_->may_be_non_uniform)) {
                 current_function_->function_tag = ReturnValueMayBeNonUniform;
             }
 
@@ -388,7 +388,7 @@ class UniformityGraph {
             // parameter node that was reachable.
             for (size_t i = 0; i < func->params.Length(); i++) {
                 auto* param = func->params[i];
-                if (reachable.contains(current_function_->variables.Get(sem_.Get(param)))) {
+                if (reachable.Contains(current_function_->variables.Get(sem_.Get(param)))) {
                     current_function_->parameters[i].tag =
                         ParameterRequiredToBeUniformForReturnValue;
                 }
@@ -404,16 +404,16 @@ class UniformityGraph {
             // Reset "visited" state for all nodes.
             current_function_->ResetVisited();
 
-            utils::UniqueVector<Node*> reachable;
+            utils::UniqueVector<Node*, 4> reachable;
             Traverse(current_function_->parameters[i].pointer_return_value, &reachable);
-            if (reachable.contains(current_function_->may_be_non_uniform)) {
+            if (reachable.Contains(current_function_->may_be_non_uniform)) {
                 current_function_->parameters[i].pointer_may_become_non_uniform = true;
             }
 
             // Check every other parameter to see if they feed into this parameter's final value.
             for (size_t j = 0; j < func->params.Length(); j++) {
                 auto* param_source = sem_.Get<sem::Parameter>(func->params[j]);
-                if (reachable.contains(current_function_->parameters[j].init_value)) {
+                if (reachable.Contains(current_function_->parameters[j].init_value)) {
                     current_function_->parameters[i].pointer_param_output_sources.push_back(
                         param_source);
                 }
@@ -1356,7 +1356,7 @@ class UniformityGraph {
     /// recording which node they were reached from.
     /// @param source the starting node
     /// @param reachable the set of reachable nodes to populate, if required
-    void Traverse(Node* source, utils::UniqueVector<Node*>* reachable = nullptr) {
+    void Traverse(Node* source, utils::UniqueVector<Node*, 4>* reachable = nullptr) {
         std::vector<Node*> to_visit{source};
 
         while (!to_visit.empty()) {
@@ -1364,7 +1364,7 @@ class UniformityGraph {
             to_visit.pop_back();
 
             if (reachable) {
-                reachable->add(node);
+                reachable->Add(node);
             }
             for (auto* to : node->edges) {
                 if (to->visited_from == nullptr) {
diff --git a/src/tint/resolver/validator.cc b/src/tint/resolver/validator.cc
index f3a02ededb..6382d34d93 100644
--- a/src/tint/resolver/validator.cc
+++ b/src/tint/resolver/validator.cc
@@ -141,7 +141,7 @@ void TraverseCallChain(diag::List& diagnostics,
             callback(f);
             return;
         }
-        if (f->TransitivelyCalledFunctions().contains(to)) {
+        if (f->TransitivelyCalledFunctions().Contains(to)) {
             TraverseCallChain(diagnostics, f, to, callback);
             callback(f);
             return;
@@ -519,7 +519,7 @@ bool Validator::StorageClassLayout(const sem::Variable* var,
                                    const ast::Extensions& enabled_extensions,
                                    ValidTypeStorageLayouts& layouts) const {
     if (var->StorageClass() == ast::StorageClass::kPushConstant &&
-        !enabled_extensions.contains(ast::Extension::kChromiumExperimentalPushConstant) &&
+        !enabled_extensions.Contains(ast::Extension::kChromiumExperimentalPushConstant) &&
         IsValidationEnabled(var->Declaration()->attributes,
                             ast::DisabledValidation::kIgnoreStorageClass)) {
         AddError(
@@ -1723,7 +1723,7 @@ bool Validator::RequiredExtensionForBuiltinFunction(
         return true;
     }
 
-    if (!enabled_extensions.contains(extension)) {
+    if (!enabled_extensions.Contains(extension)) {
         AddError("cannot call built-in function '" + std::string(builtin->str()) +
                      "' without extension " + utils::ToString(extension),
                  call->Declaration()->source);
diff --git a/src/tint/sem/function.h b/src/tint/sem/function.h
index c90d749b6a..a09f81c756 100644
--- a/src/tint/sem/function.h
+++ b/src/tint/sem/function.h
@@ -81,7 +81,7 @@ class Function final : public Castable<Function, CallTarget> {
     }
 
     /// @returns all directly referenced global variables
-    const utils::UniqueVector<const GlobalVariable*>& DirectlyReferencedGlobals() const {
+    const utils::UniqueVector<const GlobalVariable*, 4>& DirectlyReferencedGlobals() const {
         return directly_referenced_globals_;
     }
 
@@ -89,12 +89,12 @@ class Function final : public Castable<Function, CallTarget> {
     /// Note: Implicitly adds this global to the transtively-called globals.
     /// @param global the module-scope variable
     void AddDirectlyReferencedGlobal(const sem::GlobalVariable* global) {
-        directly_referenced_globals_.add(global);
-        transitively_referenced_globals_.add(global);
+        directly_referenced_globals_.Add(global);
+        transitively_referenced_globals_.Add(global);
     }
 
     /// @returns all transitively referenced global variables
-    const utils::UniqueVector<const GlobalVariable*>& TransitivelyReferencedGlobals() const {
+    const utils::UniqueVector<const GlobalVariable*, 8>& TransitivelyReferencedGlobals() const {
         return transitively_referenced_globals_;
     }
 
@@ -102,29 +102,29 @@ class Function final : public Castable<Function, CallTarget> {
     /// variable.
     /// @param global the module-scoped variable
     void AddTransitivelyReferencedGlobal(const sem::GlobalVariable* global) {
-        transitively_referenced_globals_.add(global);
+        transitively_referenced_globals_.Add(global);
     }
 
     /// @returns the list of functions that this function transitively calls.
-    const utils::UniqueVector<const Function*>& TransitivelyCalledFunctions() const {
+    const utils::UniqueVector<const Function*, 8>& TransitivelyCalledFunctions() const {
         return transitively_called_functions_;
     }
 
     /// Records that this function transitively calls `function`.
     /// @param function the function this function transitively calls
     void AddTransitivelyCalledFunction(const Function* function) {
-        transitively_called_functions_.add(function);
+        transitively_called_functions_.Add(function);
     }
 
     /// @returns the list of builtins that this function directly calls.
-    const utils::UniqueVector<const Builtin*>& DirectlyCalledBuiltins() const {
+    const utils::UniqueVector<const Builtin*, 4>& DirectlyCalledBuiltins() const {
         return directly_called_builtins_;
     }
 
     /// Records that this function transitively calls `builtin`.
     /// @param builtin the builtin this function directly calls
     void AddDirectlyCalledBuiltin(const Builtin* builtin) {
-        directly_called_builtins_.add(builtin);
+        directly_called_builtins_.Add(builtin);
     }
 
     /// Adds the given texture/sampler pair to the list of unique pairs
@@ -134,12 +134,14 @@ class Function final : public Castable<Function, CallTarget> {
     /// @param texture the texture (must be non-null)
     /// @param sampler the sampler (null indicates a texture-only reference)
     void AddTextureSamplerPair(const sem::Variable* texture, const sem::Variable* sampler) {
-        texture_sampler_pairs_.add(VariablePair(texture, sampler));
+        texture_sampler_pairs_.Add(VariablePair(texture, sampler));
     }
 
     /// @returns the list of texture/sampler pairs that this function uses
     /// (directly or indirectly).
-    const std::vector<VariablePair>& TextureSamplerPairs() const { return texture_sampler_pairs_; }
+    const utils::Vector<VariablePair, 8>& TextureSamplerPairs() const {
+        return texture_sampler_pairs_;
+    }
 
     /// @returns the list of direct calls to functions / builtins made by this
     /// function
@@ -253,17 +255,20 @@ class Function final : public Castable<Function, CallTarget> {
     sem::Behaviors& Behaviors() { return behaviors_; }
 
   private:
+    Function(const Function&) = delete;
+    Function(Function&&) = delete;
+
     VariableBindings TransitivelyReferencedSamplerVariablesImpl(ast::SamplerKind kind) const;
     VariableBindings TransitivelyReferencedSampledTextureVariablesImpl(bool multisampled) const;
 
     const ast::Function* const declaration_;
 
     sem::WorkgroupSize workgroup_size_;
-    utils::UniqueVector<const GlobalVariable*> directly_referenced_globals_;
-    utils::UniqueVector<const GlobalVariable*> transitively_referenced_globals_;
-    utils::UniqueVector<const Function*> transitively_called_functions_;
-    utils::UniqueVector<const Builtin*> directly_called_builtins_;
-    utils::UniqueVector<VariablePair> texture_sampler_pairs_;
+    utils::UniqueVector<const GlobalVariable*, 4> directly_referenced_globals_;
+    utils::UniqueVector<const GlobalVariable*, 8> transitively_referenced_globals_;
+    utils::UniqueVector<const Function*, 8> transitively_called_functions_;
+    utils::UniqueVector<const Builtin*, 4> directly_called_builtins_;
+    utils::UniqueVector<VariablePair, 8> texture_sampler_pairs_;
     std::vector<const Call*> direct_calls_;
     std::vector<const Call*> callsites_;
     std::vector<const Function*> ancestor_entry_points_;
diff --git a/src/tint/sem/module.cc b/src/tint/sem/module.cc
index 7c60650572..38f0ec63ce 100644
--- a/src/tint/sem/module.cc
+++ b/src/tint/sem/module.cc
@@ -21,7 +21,7 @@ TINT_INSTANTIATE_TYPEINFO(tint::sem::Module);
 
 namespace tint::sem {
 
-Module::Module(std::vector<const ast::Node*> dep_ordered_decls, ast::Extensions extensions)
+Module::Module(utils::VectorRef<const ast::Node*> dep_ordered_decls, ast::Extensions extensions)
     : dep_ordered_decls_(std::move(dep_ordered_decls)), extensions_(std::move(extensions)) {}
 
 Module::~Module() = default;
diff --git a/src/tint/sem/module.h b/src/tint/sem/module.h
index dffe003590..b451c5bd52 100644
--- a/src/tint/sem/module.h
+++ b/src/tint/sem/module.h
@@ -15,10 +15,9 @@
 #ifndef SRC_TINT_SEM_MODULE_H_
 #define SRC_TINT_SEM_MODULE_H_
 
-#include <vector>
-
 #include "src/tint/ast/extension.h"
 #include "src/tint/sem/node.h"
+#include "src/tint/utils/vector.h"
 
 // Forward declarations
 namespace tint::ast {
@@ -34,13 +33,13 @@ class Module final : public Castable<Module, Node> {
     /// Constructor
     /// @param dep_ordered_decls the dependency-ordered module-scope declarations
     /// @param extensions the list of enabled extensions in the module
-    Module(std::vector<const ast::Node*> dep_ordered_decls, ast::Extensions extensions);
+    Module(utils::VectorRef<const ast::Node*> dep_ordered_decls, ast::Extensions extensions);
 
     /// Destructor
     ~Module() override;
 
     /// @returns the dependency-ordered global declarations for the module
-    const std::vector<const ast::Node*>& DependencyOrderedDeclarations() const {
+    const utils::Vector<const ast::Node*, 64>& DependencyOrderedDeclarations() const {
         return dep_ordered_decls_;
     }
 
@@ -48,7 +47,7 @@ class Module final : public Castable<Module, Node> {
     const ast::Extensions& Extensions() const { return extensions_; }
 
   private:
-    const std::vector<const ast::Node*> dep_ordered_decls_;
+    const utils::Vector<const ast::Node*, 64> dep_ordered_decls_;
     ast::Extensions extensions_;
 };
 
diff --git a/src/tint/transform/combine_samplers.cc b/src/tint/transform/combine_samplers.cc
index 2f5101edf1..ca9005b3cf 100644
--- a/src/tint/transform/combine_samplers.cc
+++ b/src/tint/transform/combine_samplers.cc
@@ -167,7 +167,7 @@ struct CombineSamplers::State {
         ctx.ReplaceAll([&](const ast::Function* src) -> const ast::Function* {
             if (auto* func = sem.Get(src)) {
                 auto pairs = func->TextureSamplerPairs();
-                if (pairs.empty()) {
+                if (pairs.IsEmpty()) {
                     return nullptr;
                 }
                 utils::Vector<const ast::Parameter*, 8> params;
diff --git a/src/tint/transform/disable_uniformity_analysis.cc b/src/tint/transform/disable_uniformity_analysis.cc
index 7a3002377d..918b1f163a 100644
--- a/src/tint/transform/disable_uniformity_analysis.cc
+++ b/src/tint/transform/disable_uniformity_analysis.cc
@@ -28,7 +28,7 @@ DisableUniformityAnalysis::DisableUniformityAnalysis() = default;
 DisableUniformityAnalysis::~DisableUniformityAnalysis() = default;
 
 bool DisableUniformityAnalysis::ShouldRun(const Program* program, const DataMap&) const {
-    return !program->Sem().Module()->Extensions().contains(
+    return !program->Sem().Module()->Extensions().Contains(
         ast::Extension::kChromiumDisableUniformityAnalysis);
 }
 
diff --git a/src/tint/transform/spirv_atomic.cc b/src/tint/transform/spirv_atomic.cc
index b79f60f214..bece2d6538 100644
--- a/src/tint/transform/spirv_atomic.cc
+++ b/src/tint/transform/spirv_atomic.cc
@@ -48,7 +48,7 @@ struct SpirvAtomic::State {
     ProgramBuilder& b = *ctx.dst;
     std::unordered_map<const ast::Struct*, ForkedStruct> forked_structs;
     std::unordered_set<const sem::Variable*> atomic_variables;
-    utils::UniqueVector<const sem::Expression*> atomic_expressions;
+    utils::UniqueVector<const sem::Expression*, 8> atomic_expressions;
 
   public:
     /// Constructor
@@ -92,7 +92,7 @@ struct SpirvAtomic::State {
 
                     // Keep track of this expression. We'll need to modify the source variable /
                     // structure to be atomic.
-                    atomic_expressions.add(ctx.src->Sem().Get(args[0]));
+                    atomic_expressions.Add(ctx.src->Sem().Get(args[0]));
                 }
 
                 // Remove the stub from the output program
@@ -153,7 +153,7 @@ struct SpirvAtomic::State {
     }
 
     void ProcessAtomicExpressions() {
-        for (size_t i = 0; i < atomic_expressions.size(); i++) {
+        for (size_t i = 0; i < atomic_expressions.Length(); i++) {
             Switch(
                 atomic_expressions[i],  //
                 [&](const sem::VariableUser* user) {
@@ -162,7 +162,7 @@ struct SpirvAtomic::State {
                         ctx.Replace(v->type, AtomicTypeFor(user->Variable()->Type()));
                     }
                     if (auto* ctor = user->Variable()->Constructor()) {
-                        atomic_expressions.add(ctor);
+                        atomic_expressions.Add(ctor);
                     }
                 },
                 [&](const sem::StructMemberAccess* access) {
@@ -170,14 +170,14 @@ struct SpirvAtomic::State {
                     // atomic.
                     auto* member = access->Member();
                     Fork(member->Struct()->Declaration()).atomic_members.emplace(member->Index());
-                    atomic_expressions.add(access->Object());
+                    atomic_expressions.Add(access->Object());
                 },
                 [&](const sem::IndexAccessorExpression* index) {
-                    atomic_expressions.add(index->Object());
+                    atomic_expressions.Add(index->Object());
                 },
                 [&](const sem::Expression* e) {
                     if (auto* unary = e->Declaration()->As<ast::UnaryOpExpression>()) {
-                        atomic_expressions.add(ctx.src->Sem().Get(unary->expr));
+                        atomic_expressions.Add(ctx.src->Sem().Get(unary->expr));
                     }
                 });
         }
diff --git a/src/tint/transform/zero_init_workgroup_memory.cc b/src/tint/transform/zero_init_workgroup_memory.cc
index 963ca0124b..94df3b9885 100644
--- a/src/tint/transform/zero_init_workgroup_memory.cc
+++ b/src/tint/transform/zero_init_workgroup_memory.cc
@@ -75,7 +75,7 @@ struct ZeroInitWorkgroupMemory::State {
     };
 
     /// A list of unique ArrayIndex
-    using ArrayIndices = utils::UniqueVector<ArrayIndex, ArrayIndex::Hasher>;
+    using ArrayIndices = utils::UniqueVector<ArrayIndex, 4, ArrayIndex::Hasher>;
 
     /// Expression holds information about an expression that is being built for a
     /// statement will zero workgroup values.
@@ -193,7 +193,7 @@ struct ZeroInitWorkgroupMemory::State {
             ArrayIndices array_indices;
             for (auto& s : stmts) {
                 for (auto& idx : s.array_indices) {
-                    array_indices.add(idx);
+                    array_indices.Add(idx);
                 }
             }
 
@@ -311,7 +311,7 @@ struct ZeroInitWorkgroupMemory::State {
                 auto division = num_values;
                 auto a = get_expr(modulo);
                 auto array_indices = a.array_indices;
-                array_indices.add(ArrayIndex{modulo, division});
+                array_indices.Add(ArrayIndex{modulo, division});
                 auto index = utils::GetOrCreate(array_index_names, ArrayIndex{modulo, division},
                                                 [&] { return b.Symbols().New("i"); });
                 return Expression{b.IndexAccessor(a.expr, index), a.num_iterations, array_indices};
diff --git a/src/tint/utils/unique_vector.h b/src/tint/utils/unique_vector.h
index d4018b247c..bda090cea0 100644
--- a/src/tint/utils/unique_vector.h
+++ b/src/tint/utils/unique_vector.h
@@ -21,17 +21,15 @@
 #include <utility>
 #include <vector>
 
+#include "src/tint/utils/hashset.h"
+#include "src/tint/utils/vector.h"
+
 namespace tint::utils {
 
 /// UniqueVector is an ordered container that only contains unique items.
 /// Attempting to add a duplicate is a no-op.
-template <typename T, typename HASH = std::hash<T>, typename EQUAL = std::equal_to<T>>
+template <typename T, size_t N, typename HASH = std::hash<T>, typename EQUAL = std::equal_to<T>>
 struct UniqueVector {
-    /// The iterator returned by begin() and end()
-    using ConstIterator = typename std::vector<T>::const_iterator;
-    /// The iterator returned by rbegin() and rend()
-    using ConstReverseIterator = typename std::vector<T>::const_reverse_iterator;
-
     /// Constructor
     UniqueVector() = default;
 
@@ -40,7 +38,7 @@ struct UniqueVector {
     /// elements will be removed.
     explicit UniqueVector(std::vector<T>&& v) {
         for (auto& el : v) {
-            add(el);
+            Add(el);
         }
     }
 
@@ -48,10 +46,9 @@ struct UniqueVector {
     /// already contain the given item.
     /// @param item the item to append to the end of the vector
     /// @returns true if the item was added, otherwise false.
-    bool add(const T& item) {
-        if (set.count(item) == 0) {
-            vector.emplace_back(item);
-            set.emplace(item);
+    bool Add(const T& item) {
+        if (set.Add(item)) {
+            vector.Push(item);
             return true;
         }
         return false;
@@ -59,7 +56,7 @@ struct UniqueVector {
 
     /// @returns true if the vector contains `item`
     /// @param item the item
-    bool contains(const T& item) const { return set.count(item); }
+    bool Contains(const T& item) const { return set.Contains(item); }
 
     /// @param i the index of the element to retrieve
     /// @returns the element at the index `i`
@@ -70,48 +67,50 @@ struct UniqueVector {
     const T& operator[](size_t i) const { return vector[i]; }
 
     /// @returns true if the vector is empty
-    bool empty() const { return vector.empty(); }
+    bool IsEmpty() const { return vector.IsEmpty(); }
 
     /// @returns the number of items in the vector
-    size_t size() const { return vector.size(); }
+    size_t Length() const { return vector.Length(); }
 
     /// @returns the pointer to the first element in the vector, or nullptr if the vector is empty.
-    const T* data() const { return vector.empty() ? nullptr : vector.data(); }
+    const T* Data() const { return vector.IsEmpty() ? nullptr : &vector[0]; }
 
     /// @returns an iterator to the beginning of the vector
-    ConstIterator begin() const { return vector.begin(); }
+    auto begin() const { return vector.begin(); }
 
     /// @returns an iterator to the end of the vector
-    ConstIterator end() const { return vector.end(); }
+    auto end() const { return vector.end(); }
 
     /// @returns an iterator to the beginning of the reversed vector
-    ConstReverseIterator rbegin() const { return vector.rbegin(); }
+    auto rbegin() const { return vector.rbegin(); }
 
     /// @returns an iterator to the end of the reversed vector
-    ConstReverseIterator rend() const { return vector.rend(); }
+    auto rend() const { return vector.rend(); }
 
     /// @returns a const reference to the internal vector
-    operator const std::vector<T>&() const { return vector; }
+    operator const Vector<T, N>&() const { return vector; }
+
+    /// @returns the std::move()'d vector.
+    /// @note The UniqueVector must not be used after calling this method
+    VectorRef<T> Release() { return std::move(vector); }
 
     /// Pre-allocates `count` elements in the vector and set
     /// @param count the number of elements to pre-allocate
-    void reserve(size_t count) {
-        vector.reserve(count);
-        set.reserve(count);
+    void Reserve(size_t count) {
+        vector.Reserve(count);
+        set.Reserve(count);
     }
 
     /// Removes the last element from the vector
     /// @returns the popped element
-    T pop_back() {
-        auto el = std::move(vector.back());
-        set.erase(el);
-        vector.pop_back();
-        return el;
+    T Pop() {
+        set.Remove(vector.Back());
+        return vector.Pop();
     }
 
   private:
-    std::vector<T> vector;
-    std::unordered_set<T, HASH, EQUAL> set;
+    Vector<T, N> vector;
+    Hashset<T, N, HASH, EQUAL> set;
 };
 
 }  // namespace tint::utils
diff --git a/src/tint/utils/unique_vector_test.cc b/src/tint/utils/unique_vector_test.cc
index 035ebf8de7..9b015c2092 100644
--- a/src/tint/utils/unique_vector_test.cc
+++ b/src/tint/utils/unique_vector_test.cc
@@ -13,6 +13,9 @@
 // limitations under the License.
 
 #include "src/tint/utils/unique_vector.h"
+
+#include <vector>
+
 #include "src/tint/utils/reverse.h"
 
 #include "gtest/gtest.h"
@@ -21,16 +24,16 @@ namespace tint::utils {
 namespace {
 
 TEST(UniqueVectorTest, Empty) {
-    UniqueVector<int> unique_vec;
-    EXPECT_EQ(unique_vec.size(), 0u);
-    EXPECT_EQ(unique_vec.empty(), true);
+    UniqueVector<int, 4> unique_vec;
+    EXPECT_EQ(unique_vec.Length(), 0u);
+    EXPECT_EQ(unique_vec.IsEmpty(), true);
     EXPECT_EQ(unique_vec.begin(), unique_vec.end());
 }
 
 TEST(UniqueVectorTest, MoveConstructor) {
-    UniqueVector<int> unique_vec(std::vector<int>{0, 3, 2, 1, 2});
-    EXPECT_EQ(unique_vec.size(), 4u);
-    EXPECT_EQ(unique_vec.empty(), false);
+    UniqueVector<int, 4> unique_vec(std::vector<int>{0, 3, 2, 1, 2});
+    EXPECT_EQ(unique_vec.Length(), 4u);
+    EXPECT_EQ(unique_vec.IsEmpty(), false);
     EXPECT_EQ(unique_vec[0], 0);
     EXPECT_EQ(unique_vec[1], 3);
     EXPECT_EQ(unique_vec[2], 2);
@@ -38,12 +41,12 @@ TEST(UniqueVectorTest, MoveConstructor) {
 }
 
 TEST(UniqueVectorTest, AddUnique) {
-    UniqueVector<int> unique_vec;
-    unique_vec.add(0);
-    unique_vec.add(1);
-    unique_vec.add(2);
-    EXPECT_EQ(unique_vec.size(), 3u);
-    EXPECT_EQ(unique_vec.empty(), false);
+    UniqueVector<int, 4> unique_vec;
+    unique_vec.Add(0);
+    unique_vec.Add(1);
+    unique_vec.Add(2);
+    EXPECT_EQ(unique_vec.Length(), 3u);
+    EXPECT_EQ(unique_vec.IsEmpty(), false);
     int i = 0;
     for (auto n : unique_vec) {
         EXPECT_EQ(n, i);
@@ -59,15 +62,15 @@ TEST(UniqueVectorTest, AddUnique) {
 }
 
 TEST(UniqueVectorTest, AddDuplicates) {
-    UniqueVector<int> unique_vec;
-    unique_vec.add(0);
-    unique_vec.add(0);
-    unique_vec.add(0);
-    unique_vec.add(1);
-    unique_vec.add(1);
-    unique_vec.add(2);
-    EXPECT_EQ(unique_vec.size(), 3u);
-    EXPECT_EQ(unique_vec.empty(), false);
+    UniqueVector<int, 4> unique_vec;
+    unique_vec.Add(0);
+    unique_vec.Add(0);
+    unique_vec.Add(0);
+    unique_vec.Add(1);
+    unique_vec.Add(1);
+    unique_vec.Add(2);
+    EXPECT_EQ(unique_vec.Length(), 3u);
+    EXPECT_EQ(unique_vec.IsEmpty(), false);
     int i = 0;
     for (auto n : unique_vec) {
         EXPECT_EQ(n, i);
@@ -83,17 +86,17 @@ TEST(UniqueVectorTest, AddDuplicates) {
 }
 
 TEST(UniqueVectorTest, AsVector) {
-    UniqueVector<int> unique_vec;
-    unique_vec.add(0);
-    unique_vec.add(0);
-    unique_vec.add(0);
-    unique_vec.add(1);
-    unique_vec.add(1);
-    unique_vec.add(2);
+    UniqueVector<int, 4> unique_vec;
+    unique_vec.Add(0);
+    unique_vec.Add(0);
+    unique_vec.Add(0);
+    unique_vec.Add(1);
+    unique_vec.Add(1);
+    unique_vec.Add(2);
 
-    const std::vector<int>& vec = unique_vec;
-    EXPECT_EQ(vec.size(), 3u);
-    EXPECT_EQ(unique_vec.empty(), false);
+    const utils::Vector<int, 4>& vec = unique_vec;
+    EXPECT_EQ(vec.Length(), 3u);
+    EXPECT_EQ(unique_vec.IsEmpty(), false);
     int i = 0;
     for (auto n : vec) {
         EXPECT_EQ(n, i);
@@ -106,46 +109,46 @@ TEST(UniqueVectorTest, AsVector) {
 }
 
 TEST(UniqueVectorTest, PopBack) {
-    UniqueVector<int> unique_vec;
-    unique_vec.add(0);
-    unique_vec.add(2);
-    unique_vec.add(1);
+    UniqueVector<int, 4> unique_vec;
+    unique_vec.Add(0);
+    unique_vec.Add(2);
+    unique_vec.Add(1);
 
-    EXPECT_EQ(unique_vec.pop_back(), 1);
-    EXPECT_EQ(unique_vec.size(), 2u);
-    EXPECT_EQ(unique_vec.empty(), false);
+    EXPECT_EQ(unique_vec.Pop(), 1);
+    EXPECT_EQ(unique_vec.Length(), 2u);
+    EXPECT_EQ(unique_vec.IsEmpty(), false);
     EXPECT_EQ(unique_vec[0], 0);
     EXPECT_EQ(unique_vec[1], 2);
 
-    EXPECT_EQ(unique_vec.pop_back(), 2);
-    EXPECT_EQ(unique_vec.size(), 1u);
-    EXPECT_EQ(unique_vec.empty(), false);
+    EXPECT_EQ(unique_vec.Pop(), 2);
+    EXPECT_EQ(unique_vec.Length(), 1u);
+    EXPECT_EQ(unique_vec.IsEmpty(), false);
     EXPECT_EQ(unique_vec[0], 0);
 
-    unique_vec.add(1);
+    unique_vec.Add(1);
 
-    EXPECT_EQ(unique_vec.size(), 2u);
-    EXPECT_EQ(unique_vec.empty(), false);
+    EXPECT_EQ(unique_vec.Length(), 2u);
+    EXPECT_EQ(unique_vec.IsEmpty(), false);
     EXPECT_EQ(unique_vec[0], 0);
     EXPECT_EQ(unique_vec[1], 1);
 
-    EXPECT_EQ(unique_vec.pop_back(), 1);
-    EXPECT_EQ(unique_vec.size(), 1u);
-    EXPECT_EQ(unique_vec.empty(), false);
+    EXPECT_EQ(unique_vec.Pop(), 1);
+    EXPECT_EQ(unique_vec.Length(), 1u);
+    EXPECT_EQ(unique_vec.IsEmpty(), false);
     EXPECT_EQ(unique_vec[0], 0);
 
-    EXPECT_EQ(unique_vec.pop_back(), 0);
-    EXPECT_EQ(unique_vec.size(), 0u);
-    EXPECT_EQ(unique_vec.empty(), true);
+    EXPECT_EQ(unique_vec.Pop(), 0);
+    EXPECT_EQ(unique_vec.Length(), 0u);
+    EXPECT_EQ(unique_vec.IsEmpty(), true);
 }
 
 TEST(UniqueVectorTest, Data) {
-    UniqueVector<int> unique_vec;
-    EXPECT_EQ(unique_vec.data(), nullptr);
+    UniqueVector<int, 4> unique_vec;
+    EXPECT_EQ(unique_vec.Data(), nullptr);
 
-    unique_vec.add(42);
-    EXPECT_EQ(unique_vec.data(), &unique_vec[0]);
-    EXPECT_EQ(*unique_vec.data(), 42);
+    unique_vec.Add(42);
+    EXPECT_EQ(unique_vec.Data(), &unique_vec[0]);
+    EXPECT_EQ(*unique_vec.Data(), 42);
 }
 
 }  // namespace