Avoid redundant creation of ComputePipelineBase in GetCachedComputePipeline

This patch removes a redundant creation of ComputePipelineBase object
in GetCachedComputePipeline(). Instead, we directly compute the blueprint
hash from the uninitialized backend compute pipeline object.

BUG=dawn:529
TEST=dawn_end2end_tests

Change-Id: I9b982664aa140ab385418a202270b9988cfcb9f3
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/66221
Reviewed-by: Austin Eng <enga@chromium.org>
Commit-Queue: Jiawei Shao <jiawei.shao@intel.com>
diff --git a/src/dawn_native/ComputePipeline.cpp b/src/dawn_native/ComputePipeline.cpp
index f789235..d79d4a0 100644
--- a/src/dawn_native/ComputePipeline.cpp
+++ b/src/dawn_native/ComputePipeline.cpp
@@ -46,6 +46,7 @@
                        {{SingleShaderStage::Compute, descriptor->compute.module,
                          descriptor->compute.entryPoint, descriptor->compute.constantCount,
                          descriptor->compute.constants}}) {
+        SetContentHash(ComputeContentHash());
     }
 
     ComputePipelineBase::ComputePipelineBase(DeviceBase* device, ObjectBase::ErrorTag tag)
@@ -59,13 +60,21 @@
         }
     }
 
-    MaybeError ComputePipelineBase::Initialize() {
-        return {};
-    }
-
     // static
     ComputePipelineBase* ComputePipelineBase::MakeError(DeviceBase* device) {
-        return new ComputePipelineBase(device, ObjectBase::kError);
+        class ErrorComputePipeline final : public ComputePipelineBase {
+          public:
+            ErrorComputePipeline(DeviceBase* device)
+                : ComputePipelineBase(device, ObjectBase::kError) {
+            }
+
+            MaybeError Initialize() override {
+                UNREACHABLE();
+                return {};
+            }
+        };
+
+        return new ErrorComputePipeline(device);
     }
 
     ObjectType ComputePipelineBase::GetType() const {
diff --git a/src/dawn_native/ComputePipeline.h b/src/dawn_native/ComputePipeline.h
index 1c134e6..6352c19 100644
--- a/src/dawn_native/ComputePipeline.h
+++ b/src/dawn_native/ComputePipeline.h
@@ -43,11 +43,6 @@
 
       private:
         ComputePipelineBase(DeviceBase* device, ObjectBase::ErrorTag tag);
-
-        // CreateComputePipelineAsyncTask is declared as a friend of ComputePipelineBase as it
-        // needs to call the private member function ComputePipelineBase::Initialize().
-        friend class CreateComputePipelineAsyncTask;
-        virtual MaybeError Initialize();
     };
 
 }  // namespace dawn_native
diff --git a/src/dawn_native/CreatePipelineAsyncTask.cpp b/src/dawn_native/CreatePipelineAsyncTask.cpp
index 6ada64e..f806bb5 100644
--- a/src/dawn_native/CreatePipelineAsyncTask.cpp
+++ b/src/dawn_native/CreatePipelineAsyncTask.cpp
@@ -103,11 +103,9 @@
 
     CreateComputePipelineAsyncTask::CreateComputePipelineAsyncTask(
         Ref<ComputePipelineBase> nonInitializedComputePipeline,
-        size_t blueprintHash,
         WGPUCreateComputePipelineAsyncCallback callback,
         void* userdata)
         : mComputePipeline(std::move(nonInitializedComputePipeline)),
-          mBlueprintHash(blueprintHash),
           mCallback(callback),
           mUserdata(userdata) {
         ASSERT(mComputePipeline != nullptr);
@@ -122,7 +120,7 @@
         }
 
         mComputePipeline->GetDevice()->AddComputePipelineAsyncCallbackTask(
-            mComputePipeline, errorMessage, mCallback, mUserdata, mBlueprintHash);
+            mComputePipeline, errorMessage, mCallback, mUserdata);
     }
 
     void CreateComputePipelineAsyncTask::RunAsync(
diff --git a/src/dawn_native/CreatePipelineAsyncTask.h b/src/dawn_native/CreatePipelineAsyncTask.h
index 6a36ff0..3bac477 100644
--- a/src/dawn_native/CreatePipelineAsyncTask.h
+++ b/src/dawn_native/CreatePipelineAsyncTask.h
@@ -72,7 +72,6 @@
     class CreateComputePipelineAsyncTask {
       public:
         CreateComputePipelineAsyncTask(Ref<ComputePipelineBase> nonInitializedComputePipeline,
-                                       size_t blueprintHash,
                                        WGPUCreateComputePipelineAsyncCallback callback,
                                        void* userdata);
 
@@ -82,7 +81,6 @@
 
       private:
         Ref<ComputePipelineBase> mComputePipeline;
-        size_t mBlueprintHash;
         WGPUCreateComputePipelineAsyncCallback mCallback;
         void* mUserdata;
     };
diff --git a/src/dawn_native/Device.cpp b/src/dawn_native/Device.cpp
index a867110..03fe3b9 100644
--- a/src/dawn_native/Device.cpp
+++ b/src/dawn_native/Device.cpp
@@ -678,20 +678,15 @@
         return mEmptyBindGroupLayout.Get();
     }
 
-    std::pair<Ref<ComputePipelineBase>, size_t> DeviceBase::GetCachedComputePipeline(
-        const ComputePipelineDescriptor* descriptor) {
-        ComputePipelineBase blueprint(this, descriptor);
-
-        const size_t blueprintHash = blueprint.ComputeContentHash();
-        blueprint.SetContentHash(blueprintHash);
-
-        Ref<ComputePipelineBase> result;
-        auto iter = mCaches->computePipelines.find(&blueprint);
+    Ref<ComputePipelineBase> DeviceBase::GetCachedComputePipeline(
+        ComputePipelineBase* uninitializedComputePipeline) {
+        Ref<ComputePipelineBase> cachedPipeline;
+        auto iter = mCaches->computePipelines.find(uninitializedComputePipeline);
         if (iter != mCaches->computePipelines.end()) {
-            result = *iter;
+            cachedPipeline = *iter;
         }
 
-        return std::make_pair(result, blueprintHash);
+        return cachedPipeline;
     }
 
     Ref<RenderPipelineBase> DeviceBase::GetCachedRenderPipeline(
@@ -705,9 +700,7 @@
     }
 
     Ref<ComputePipelineBase> DeviceBase::AddOrGetCachedComputePipeline(
-        Ref<ComputePipelineBase> computePipeline,
-        size_t blueprintHash) {
-        computePipeline->SetContentHash(blueprintHash);
+        Ref<ComputePipelineBase> computePipeline) {
         auto insertion = mCaches->computePipelines.insert(computePipeline.Get());
         if (insertion.second) {
             computePipeline->SetIsCachedReference();
@@ -1226,15 +1219,16 @@
         DAWN_TRY_ASSIGN(layoutRef, ValidateLayoutAndGetComputePipelineDescriptorWithDefaults(
                                        this, *descriptor, &appliedDescriptor));
 
-        auto pipelineAndBlueprintFromCache = GetCachedComputePipeline(&appliedDescriptor);
-        if (pipelineAndBlueprintFromCache.first.Get() != nullptr) {
-            return std::move(pipelineAndBlueprintFromCache.first);
+        Ref<ComputePipelineBase> uninitializedComputePipeline =
+            CreateUninitializedComputePipelineImpl(&appliedDescriptor);
+        Ref<ComputePipelineBase> cachedComputePipeline =
+            GetCachedComputePipeline(uninitializedComputePipeline.Get());
+        if (cachedComputePipeline.Get() != nullptr) {
+            return cachedComputePipeline;
         }
 
-        Ref<ComputePipelineBase> backendObj;
-        DAWN_TRY_ASSIGN(backendObj, CreateComputePipelineImpl(&appliedDescriptor));
-        size_t blueprintHash = pipelineAndBlueprintFromCache.second;
-        return AddOrGetCachedComputePipeline(backendObj, blueprintHash);
+        DAWN_TRY(uninitializedComputePipeline->Initialize());
+        return AddOrGetCachedComputePipeline(std::move(uninitializedComputePipeline));
     }
 
     MaybeError DeviceBase::CreateComputePipelineAsync(
@@ -1251,38 +1245,42 @@
         DAWN_TRY_ASSIGN(layoutRef, ValidateLayoutAndGetComputePipelineDescriptorWithDefaults(
                                        this, *descriptor, &appliedDescriptor));
 
+        Ref<ComputePipelineBase> uninitializedComputePipeline =
+            CreateUninitializedComputePipelineImpl(&appliedDescriptor);
+
         // Call the callback directly when we can get a cached compute pipeline object.
-        auto pipelineAndBlueprintFromCache = GetCachedComputePipeline(&appliedDescriptor);
-        if (pipelineAndBlueprintFromCache.first.Get() != nullptr) {
-            Ref<ComputePipelineBase> result = std::move(pipelineAndBlueprintFromCache.first);
+        Ref<ComputePipelineBase> cachedComputePipeline =
+            GetCachedComputePipeline(uninitializedComputePipeline.Get());
+        if (cachedComputePipeline.Get() != nullptr) {
             callback(WGPUCreatePipelineAsyncStatus_Success,
-                     reinterpret_cast<WGPUComputePipeline>(result.Detach()), "", userdata);
+                     reinterpret_cast<WGPUComputePipeline>(cachedComputePipeline.Detach()), "",
+                     userdata);
         } else {
-            // Otherwise we will create the pipeline object in CreateComputePipelineAsyncImpl(),
-            // where the pipeline object may be created asynchronously and the result will be saved
-            // to mCreatePipelineAsyncTracker.
-            const size_t blueprintHash = pipelineAndBlueprintFromCache.second;
-            CreateComputePipelineAsyncImpl(&appliedDescriptor, blueprintHash, callback, userdata);
+            // Otherwise we will create the pipeline object in InitializeComputePipelineAsyncImpl(),
+            // where the pipeline object may be initialized asynchronously and the result will be
+            // saved to mCreatePipelineAsyncTracker.
+            InitializeComputePipelineAsyncImpl(std::move(uninitializedComputePipeline), callback,
+                                               userdata);
         }
 
         return {};
     }
 
-    // This function is overwritten with the async version on the backends that supports creating
-    // compute pipeline asynchronously.
-    void DeviceBase::CreateComputePipelineAsyncImpl(const ComputePipelineDescriptor* descriptor,
-                                                    size_t blueprintHash,
-                                                    WGPUCreateComputePipelineAsyncCallback callback,
-                                                    void* userdata) {
+    // This function is overwritten with the async version on the backends that supports
+    //  initializing compute pipelines asynchronously.
+    void DeviceBase::InitializeComputePipelineAsyncImpl(
+        Ref<ComputePipelineBase> computePipeline,
+        WGPUCreateComputePipelineAsyncCallback callback,
+        void* userdata) {
         Ref<ComputePipelineBase> result;
         std::string errorMessage;
 
-        auto resultOrError = CreateComputePipelineImpl(descriptor);
-        if (resultOrError.IsError()) {
-            std::unique_ptr<ErrorData> error = resultOrError.AcquireError();
+        MaybeError maybeError = computePipeline->Initialize();
+        if (maybeError.IsError()) {
+            std::unique_ptr<ErrorData> error = maybeError.AcquireError();
             errorMessage = error->GetMessage();
         } else {
-            result = AddOrGetCachedComputePipeline(resultOrError.AcquireSuccess(), blueprintHash);
+            result = AddOrGetCachedComputePipeline(std::move(computePipeline));
         }
 
         std::unique_ptr<CreateComputePipelineAsyncCallbackTask> callbackTask =
@@ -1305,7 +1303,7 @@
             std::unique_ptr<ErrorData> error = maybeError.AcquireError();
             errorMessage = error->GetMessage();
         } else {
-            result = AddOrGetCachedRenderPipeline(renderPipeline);
+            result = AddOrGetCachedRenderPipeline(std::move(renderPipeline));
         }
 
         std::unique_ptr<CreateRenderPipelineAsyncCallbackTask> callbackTask =
@@ -1367,7 +1365,7 @@
                                        this, *descriptor, &appliedDescriptor));
 
         Ref<RenderPipelineBase> uninitializedRenderPipeline =
-            CreateUninitializedRenderPipeline(&appliedDescriptor);
+            CreateUninitializedRenderPipelineImpl(&appliedDescriptor);
 
         Ref<RenderPipelineBase> cachedRenderPipeline =
             GetCachedRenderPipeline(uninitializedRenderPipeline.Get());
@@ -1395,7 +1393,7 @@
                                        this, *descriptor, &appliedDescriptor));
 
         Ref<RenderPipelineBase> uninitializedRenderPipeline =
-            CreateUninitializedRenderPipeline(&appliedDescriptor);
+            CreateUninitializedRenderPipelineImpl(&appliedDescriptor);
 
         // Call the callback directly when we can get a cached render pipeline object.
         Ref<RenderPipelineBase> cachedRenderPipeline =
@@ -1578,53 +1576,31 @@
         return mWorkerTaskPool.get();
     }
 
-    Ref<RenderPipelineBase> DeviceBase::CreateUninitializedRenderPipeline(
-        const RenderPipelineDescriptor* descriptor) {
-        return CreateUninitializedRenderPipelineImpl(descriptor);
-    }
-
     void DeviceBase::AddComputePipelineAsyncCallbackTask(
         Ref<ComputePipelineBase> pipeline,
         std::string errorMessage,
         WGPUCreateComputePipelineAsyncCallback callback,
-        void* userdata,
-        size_t blueprintHash) {
+        void* userdata) {
         // CreateComputePipelineAsyncWaitableCallbackTask is declared as an internal class as it
         // needs to call the private member function DeviceBase::AddOrGetCachedComputePipeline().
         struct CreateComputePipelineAsyncWaitableCallbackTask final
             : CreateComputePipelineAsyncCallbackTask {
-            CreateComputePipelineAsyncWaitableCallbackTask(
-                Ref<ComputePipelineBase> pipeline,
-                std::string errorMessage,
-                WGPUCreateComputePipelineAsyncCallback callback,
-                void* userdata,
-                size_t blueprintHash)
-                : CreateComputePipelineAsyncCallbackTask(std::move(pipeline),
-                                                         errorMessage,
-                                                         callback,
-                                                         userdata),
-                  mBlueprintHash(blueprintHash) {
-            }
-
+            using CreateComputePipelineAsyncCallbackTask::CreateComputePipelineAsyncCallbackTask;
             void Finish() final {
                 // TODO(dawn:529): call AddOrGetCachedComputePipeline() asynchronously in
                 // CreateComputePipelineAsyncTaskImpl::Run() when the front-end pipeline cache is
                 // thread-safe.
                 if (mPipeline.Get() != nullptr) {
-                    mPipeline = mPipeline->GetDevice()->AddOrGetCachedComputePipeline(
-                        mPipeline, mBlueprintHash);
+                    mPipeline = mPipeline->GetDevice()->AddOrGetCachedComputePipeline(mPipeline);
                 }
 
                 CreateComputePipelineAsyncCallbackTask::Finish();
             }
-
-          private:
-            size_t mBlueprintHash;
         };
 
         mCallbackTaskManager->AddCallbackTask(
             std::make_unique<CreateComputePipelineAsyncWaitableCallbackTask>(
-                std::move(pipeline), errorMessage, callback, userdata, blueprintHash));
+                std::move(pipeline), errorMessage, callback, userdata));
     }
 
     void DeviceBase::AddRenderPipelineAsyncCallbackTask(
diff --git a/src/dawn_native/Device.h b/src/dawn_native/Device.h
index c2969c3..92c1cd3 100644
--- a/src/dawn_native/Device.h
+++ b/src/dawn_native/Device.h
@@ -342,8 +342,7 @@
         void AddComputePipelineAsyncCallbackTask(Ref<ComputePipelineBase> pipeline,
                                                  std::string errorMessage,
                                                  WGPUCreateComputePipelineAsyncCallback callback,
-                                                 void* userdata,
-                                                 size_t blueprintHash);
+                                                 void* userdata);
         void AddRenderPipelineAsyncCallbackTask(Ref<RenderPipelineBase> pipeline,
                                                 std::string errorMessage,
                                                 WGPUCreateRenderPipelineAsyncCallback callback,
@@ -376,14 +375,10 @@
             PipelineCompatibilityToken pipelineCompatibilityToken) = 0;
         virtual ResultOrError<Ref<BufferBase>> CreateBufferImpl(
             const BufferDescriptor* descriptor) = 0;
-        virtual ResultOrError<Ref<ComputePipelineBase>> CreateComputePipelineImpl(
-            const ComputePipelineDescriptor* descriptor) = 0;
         virtual ResultOrError<Ref<PipelineLayoutBase>> CreatePipelineLayoutImpl(
             const PipelineLayoutDescriptor* descriptor) = 0;
         virtual ResultOrError<Ref<QuerySetBase>> CreateQuerySetImpl(
             const QuerySetDescriptor* descriptor) = 0;
-        virtual Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
-            const RenderPipelineDescriptor* descriptor) = 0;
         virtual ResultOrError<Ref<SamplerBase>> CreateSamplerImpl(
             const SamplerDescriptor* descriptor) = 0;
         virtual ResultOrError<Ref<ShaderModuleBase>> CreateShaderModuleImpl(
@@ -401,6 +396,10 @@
         virtual ResultOrError<Ref<TextureViewBase>> CreateTextureViewImpl(
             TextureBase* texture,
             const TextureViewDescriptor* descriptor) = 0;
+        virtual Ref<ComputePipelineBase> CreateUninitializedComputePipelineImpl(
+            const ComputePipelineDescriptor* descriptor) = 0;
+        virtual Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
+            const RenderPipelineDescriptor* descriptor) = 0;
         virtual void SetLabelImpl();
 
         virtual MaybeError TickImpl() = 0;
@@ -408,21 +407,18 @@
 
         ResultOrError<Ref<BindGroupLayoutBase>> CreateEmptyBindGroupLayout();
 
-        std::pair<Ref<ComputePipelineBase>, size_t> GetCachedComputePipeline(
-            const ComputePipelineDescriptor* descriptor);
+        Ref<ComputePipelineBase> GetCachedComputePipeline(
+            ComputePipelineBase* uninitializedComputePipeline);
         Ref<RenderPipelineBase> GetCachedRenderPipeline(
             RenderPipelineBase* uninitializedRenderPipeline);
         Ref<ComputePipelineBase> AddOrGetCachedComputePipeline(
-            Ref<ComputePipelineBase> computePipeline,
-            size_t blueprintHash);
+            Ref<ComputePipelineBase> computePipeline);
         Ref<RenderPipelineBase> AddOrGetCachedRenderPipeline(
             Ref<RenderPipelineBase> renderPipeline);
-        virtual void CreateComputePipelineAsyncImpl(const ComputePipelineDescriptor* descriptor,
-                                                    size_t blueprintHash,
-                                                    WGPUCreateComputePipelineAsyncCallback callback,
-                                                    void* userdata);
-        Ref<RenderPipelineBase> CreateUninitializedRenderPipeline(
-            const RenderPipelineDescriptor* descriptor);
+        virtual void InitializeComputePipelineAsyncImpl(
+            Ref<ComputePipelineBase> computePipeline,
+            WGPUCreateComputePipelineAsyncCallback callback,
+            void* userdata);
         virtual void InitializeRenderPipelineAsyncImpl(
             Ref<RenderPipelineBase> renderPipeline,
             WGPUCreateRenderPipelineAsyncCallback callback,
diff --git a/src/dawn_native/Pipeline.h b/src/dawn_native/Pipeline.h
index c73d389..7444293 100644
--- a/src/dawn_native/Pipeline.h
+++ b/src/dawn_native/Pipeline.h
@@ -66,6 +66,9 @@
         // Implementation of the API entrypoint. Do not use in a reentrant manner.
         BindGroupLayoutBase* APIGetBindGroupLayout(uint32_t groupIndex);
 
+        // Initialize() should only be called once by the frontend.
+        virtual MaybeError Initialize() = 0;
+
       protected:
         PipelineBase(DeviceBase* device,
                      PipelineLayoutBase* layout,
diff --git a/src/dawn_native/RenderPipeline.h b/src/dawn_native/RenderPipeline.h
index 464ee5d..b38ec97 100644
--- a/src/dawn_native/RenderPipeline.h
+++ b/src/dawn_native/RenderPipeline.h
@@ -105,9 +105,6 @@
             bool operator()(const RenderPipelineBase* a, const RenderPipelineBase* b) const;
         };
 
-        // Initialize() should only be called once by the frontend.
-        virtual MaybeError Initialize() = 0;
-
       private:
         RenderPipelineBase(DeviceBase* device, ObjectBase::ErrorTag tag);
 
diff --git a/src/dawn_native/d3d12/ComputePipelineD3D12.cpp b/src/dawn_native/d3d12/ComputePipelineD3D12.cpp
index 0925b92..29ae08a 100644
--- a/src/dawn_native/d3d12/ComputePipelineD3D12.cpp
+++ b/src/dawn_native/d3d12/ComputePipelineD3D12.cpp
@@ -24,12 +24,10 @@
 
 namespace dawn_native { namespace d3d12 {
 
-    ResultOrError<Ref<ComputePipeline>> ComputePipeline::Create(
+    Ref<ComputePipeline> ComputePipeline::CreateUninitialized(
         Device* device,
         const ComputePipelineDescriptor* descriptor) {
-        Ref<ComputePipeline> pipeline = AcquireRef(new ComputePipeline(device, descriptor));
-        DAWN_TRY(pipeline->Initialize());
-        return pipeline;
+        return AcquireRef(new ComputePipeline(device, descriptor));
     }
 
     MaybeError ComputePipeline::Initialize() {
@@ -77,14 +75,11 @@
                      GetLabel());
     }
 
-    void ComputePipeline::CreateAsync(Device* device,
-                                      const ComputePipelineDescriptor* descriptor,
-                                      size_t blueprintHash,
-                                      WGPUCreateComputePipelineAsyncCallback callback,
-                                      void* userdata) {
-        Ref<ComputePipeline> pipeline = AcquireRef(new ComputePipeline(device, descriptor));
+    void ComputePipeline::InitializeAsync(Ref<ComputePipelineBase> computePipeline,
+                                          WGPUCreateComputePipelineAsyncCallback callback,
+                                          void* userdata) {
         std::unique_ptr<CreateComputePipelineAsyncTask> asyncTask =
-            std::make_unique<CreateComputePipelineAsyncTask>(pipeline, blueprintHash, callback,
+            std::make_unique<CreateComputePipelineAsyncTask>(std::move(computePipeline), callback,
                                                              userdata);
         CreateComputePipelineAsyncTask::RunAsync(std::move(asyncTask));
     }
diff --git a/src/dawn_native/d3d12/ComputePipelineD3D12.h b/src/dawn_native/d3d12/ComputePipelineD3D12.h
index d945ee2..7c7a02d 100644
--- a/src/dawn_native/d3d12/ComputePipelineD3D12.h
+++ b/src/dawn_native/d3d12/ComputePipelineD3D12.h
@@ -25,25 +25,24 @@
 
     class ComputePipeline final : public ComputePipelineBase {
       public:
-        static ResultOrError<Ref<ComputePipeline>> Create(
+        static Ref<ComputePipeline> CreateUninitialized(
             Device* device,
             const ComputePipelineDescriptor* descriptor);
-        static void CreateAsync(Device* device,
-                                const ComputePipelineDescriptor* descriptor,
-                                size_t blueprintHash,
-                                WGPUCreateComputePipelineAsyncCallback callback,
-                                void* userdata);
+        static void InitializeAsync(Ref<ComputePipelineBase> computePipeline,
+                                    WGPUCreateComputePipelineAsyncCallback callback,
+                                    void* userdata);
         ComputePipeline() = delete;
 
         ID3D12PipelineState* GetPipelineState() const;
 
+        MaybeError Initialize() override;
+
         // Dawn API
         void SetLabelImpl() override;
 
       private:
         ~ComputePipeline() override;
         using ComputePipelineBase::ComputePipelineBase;
-        MaybeError Initialize() override;
         ComPtr<ID3D12PipelineState> mPipelineState;
     };
 
diff --git a/src/dawn_native/d3d12/DeviceD3D12.cpp b/src/dawn_native/d3d12/DeviceD3D12.cpp
index 4e019c2..0c2d34c 100644
--- a/src/dawn_native/d3d12/DeviceD3D12.cpp
+++ b/src/dawn_native/d3d12/DeviceD3D12.cpp
@@ -334,9 +334,9 @@
         const CommandBufferDescriptor* descriptor) {
         return CommandBuffer::Create(encoder, descriptor);
     }
-    ResultOrError<Ref<ComputePipelineBase>> Device::CreateComputePipelineImpl(
+    Ref<ComputePipelineBase> Device::CreateUninitializedComputePipelineImpl(
         const ComputePipelineDescriptor* descriptor) {
-        return ComputePipeline::Create(this, descriptor);
+        return ComputePipeline::CreateUninitialized(this, descriptor);
     }
     ResultOrError<Ref<PipelineLayoutBase>> Device::CreatePipelineLayoutImpl(
         const PipelineLayoutDescriptor* descriptor) {
@@ -376,16 +376,15 @@
         const TextureViewDescriptor* descriptor) {
         return TextureView::Create(texture, descriptor);
     }
-    void Device::CreateComputePipelineAsyncImpl(const ComputePipelineDescriptor* descriptor,
-                                                size_t blueprintHash,
-                                                WGPUCreateComputePipelineAsyncCallback callback,
-                                                void* userdata) {
-        ComputePipeline::CreateAsync(this, descriptor, blueprintHash, callback, userdata);
+    void Device::InitializeComputePipelineAsyncImpl(Ref<ComputePipelineBase> computePipeline,
+                                                    WGPUCreateComputePipelineAsyncCallback callback,
+                                                    void* userdata) {
+        ComputePipeline::InitializeAsync(std::move(computePipeline), callback, userdata);
     }
     void Device::InitializeRenderPipelineAsyncImpl(Ref<RenderPipelineBase> renderPipeline,
                                                    WGPUCreateRenderPipelineAsyncCallback callback,
                                                    void* userdata) {
-        RenderPipeline::InitializeAsync(renderPipeline, callback, userdata);
+        RenderPipeline::InitializeAsync(std::move(renderPipeline), callback, userdata);
     }
 
     ResultOrError<std::unique_ptr<StagingBufferBase>> Device::CreateStagingBuffer(size_t size) {
diff --git a/src/dawn_native/d3d12/DeviceD3D12.h b/src/dawn_native/d3d12/DeviceD3D12.h
index 03856be..0e78cf9 100644
--- a/src/dawn_native/d3d12/DeviceD3D12.h
+++ b/src/dawn_native/d3d12/DeviceD3D12.h
@@ -149,14 +149,10 @@
             PipelineCompatibilityToken pipelineCompatibilityToken) override;
         ResultOrError<Ref<BufferBase>> CreateBufferImpl(
             const BufferDescriptor* descriptor) override;
-        ResultOrError<Ref<ComputePipelineBase>> CreateComputePipelineImpl(
-            const ComputePipelineDescriptor* descriptor) override;
         ResultOrError<Ref<PipelineLayoutBase>> CreatePipelineLayoutImpl(
             const PipelineLayoutDescriptor* descriptor) override;
         ResultOrError<Ref<QuerySetBase>> CreateQuerySetImpl(
             const QuerySetDescriptor* descriptor) override;
-        Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
-            const RenderPipelineDescriptor* descriptor) override;
         ResultOrError<Ref<SamplerBase>> CreateSamplerImpl(
             const SamplerDescriptor* descriptor) override;
         ResultOrError<Ref<ShaderModuleBase>> CreateShaderModuleImpl(
@@ -173,10 +169,13 @@
         ResultOrError<Ref<TextureViewBase>> CreateTextureViewImpl(
             TextureBase* texture,
             const TextureViewDescriptor* descriptor) override;
-        void CreateComputePipelineAsyncImpl(const ComputePipelineDescriptor* descriptor,
-                                            size_t blueprintHash,
-                                            WGPUCreateComputePipelineAsyncCallback callback,
-                                            void* userdata) override;
+        Ref<ComputePipelineBase> CreateUninitializedComputePipelineImpl(
+            const ComputePipelineDescriptor* descriptor) override;
+        Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
+            const RenderPipelineDescriptor* descriptor) override;
+        void InitializeComputePipelineAsyncImpl(Ref<ComputePipelineBase> computePipeline,
+                                                WGPUCreateComputePipelineAsyncCallback callback,
+                                                void* userdata) override;
         void InitializeRenderPipelineAsyncImpl(Ref<RenderPipelineBase> renderPipeline,
                                                WGPUCreateRenderPipelineAsyncCallback callback,
                                                void* userdata) override;
diff --git a/src/dawn_native/metal/ComputePipelineMTL.h b/src/dawn_native/metal/ComputePipelineMTL.h
index 4ecb450..20b2080 100644
--- a/src/dawn_native/metal/ComputePipelineMTL.h
+++ b/src/dawn_native/metal/ComputePipelineMTL.h
@@ -27,14 +27,12 @@
 
     class ComputePipeline final : public ComputePipelineBase {
       public:
-        static ResultOrError<Ref<ComputePipeline>> Create(
+        static Ref<ComputePipeline> CreateUninitialized(
             Device* device,
             const ComputePipelineDescriptor* descriptor);
-        static void CreateAsync(Device* device,
-                                const ComputePipelineDescriptor* descriptor,
-                                size_t blueprintHash,
-                                WGPUCreateComputePipelineAsyncCallback callback,
-                                void* userdata);
+        static void InitializeAsync(Ref<ComputePipelineBase> computePipeline,
+                                    WGPUCreateComputePipelineAsyncCallback callback,
+                                    void* userdata);
 
         void Encode(id<MTLComputeCommandEncoder> encoder);
         MTLSize GetLocalWorkGroupSize() const;
diff --git a/src/dawn_native/metal/ComputePipelineMTL.mm b/src/dawn_native/metal/ComputePipelineMTL.mm
index 8879fb2..18edc56 100644
--- a/src/dawn_native/metal/ComputePipelineMTL.mm
+++ b/src/dawn_native/metal/ComputePipelineMTL.mm
@@ -21,12 +21,10 @@
 namespace dawn_native { namespace metal {
 
     // static
-    ResultOrError<Ref<ComputePipeline>> ComputePipeline::Create(
+    Ref<ComputePipeline> ComputePipeline::CreateUninitialized(
         Device* device,
         const ComputePipelineDescriptor* descriptor) {
-        Ref<ComputePipeline> pipeline = AcquireRef(new ComputePipeline(device, descriptor));
-        DAWN_TRY(pipeline->Initialize());
-        return pipeline;
+        return AcquireRef(new ComputePipeline(device, descriptor));
     }
 
     MaybeError ComputePipeline::Initialize() {
@@ -76,14 +74,11 @@
         return mRequiresStorageBufferLength;
     }
 
-    void ComputePipeline::CreateAsync(Device* device,
-                                      const ComputePipelineDescriptor* descriptor,
-                                      size_t blueprintHash,
-                                      WGPUCreateComputePipelineAsyncCallback callback,
-                                      void* userdata) {
-        Ref<ComputePipeline> pipeline = AcquireRef(new ComputePipeline(device, descriptor));
+    void ComputePipeline::InitializeAsync(Ref<ComputePipelineBase> computePipeline,
+                                          WGPUCreateComputePipelineAsyncCallback callback,
+                                          void* userdata) {
         std::unique_ptr<CreateComputePipelineAsyncTask> asyncTask =
-            std::make_unique<CreateComputePipelineAsyncTask>(pipeline, blueprintHash, callback,
+            std::make_unique<CreateComputePipelineAsyncTask>(std::move(computePipeline), callback,
                                                              userdata);
         CreateComputePipelineAsyncTask::RunAsync(std::move(asyncTask));
     }
diff --git a/src/dawn_native/metal/DeviceMTL.h b/src/dawn_native/metal/DeviceMTL.h
index 5d16d8e..d7d09a2 100644
--- a/src/dawn_native/metal/DeviceMTL.h
+++ b/src/dawn_native/metal/DeviceMTL.h
@@ -89,14 +89,10 @@
         ResultOrError<Ref<CommandBufferBase>> CreateCommandBuffer(
             CommandEncoder* encoder,
             const CommandBufferDescriptor* descriptor) override;
-        ResultOrError<Ref<ComputePipelineBase>> CreateComputePipelineImpl(
-            const ComputePipelineDescriptor* descriptor) override;
         ResultOrError<Ref<PipelineLayoutBase>> CreatePipelineLayoutImpl(
             const PipelineLayoutDescriptor* descriptor) override;
         ResultOrError<Ref<QuerySetBase>> CreateQuerySetImpl(
             const QuerySetDescriptor* descriptor) override;
-        Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
-            const RenderPipelineDescriptor* descriptor) override;
         ResultOrError<Ref<SamplerBase>> CreateSamplerImpl(
             const SamplerDescriptor* descriptor) override;
         ResultOrError<Ref<ShaderModuleBase>> CreateShaderModuleImpl(
@@ -113,10 +109,13 @@
         ResultOrError<Ref<TextureViewBase>> CreateTextureViewImpl(
             TextureBase* texture,
             const TextureViewDescriptor* descriptor) override;
-        void CreateComputePipelineAsyncImpl(const ComputePipelineDescriptor* descriptor,
-                                            size_t blueprintHash,
-                                            WGPUCreateComputePipelineAsyncCallback callback,
-                                            void* userdata) override;
+        Ref<ComputePipelineBase> CreateUninitializedComputePipelineImpl(
+            const ComputePipelineDescriptor* descriptor) override;
+        Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
+            const RenderPipelineDescriptor* descriptor) override;
+        void InitializeComputePipelineAsyncImpl(Ref<ComputePipelineBase> computePipeline,
+                                                WGPUCreateComputePipelineAsyncCallback callback,
+                                                void* userdata) override;
         void InitializeRenderPipelineAsyncImpl(Ref<RenderPipelineBase> renderPipeline,
                                                WGPUCreateRenderPipelineAsyncCallback callback,
                                                void* userdata) override;
diff --git a/src/dawn_native/metal/DeviceMTL.mm b/src/dawn_native/metal/DeviceMTL.mm
index acc351f..10445e5 100644
--- a/src/dawn_native/metal/DeviceMTL.mm
+++ b/src/dawn_native/metal/DeviceMTL.mm
@@ -243,9 +243,9 @@
         const CommandBufferDescriptor* descriptor) {
         return CommandBuffer::Create(encoder, descriptor);
     }
-    ResultOrError<Ref<ComputePipelineBase>> Device::CreateComputePipelineImpl(
+    Ref<ComputePipelineBase> Device::CreateUninitializedComputePipelineImpl(
         const ComputePipelineDescriptor* descriptor) {
-        return ComputePipeline::Create(this, descriptor);
+        return ComputePipeline::CreateUninitialized(this, descriptor);
     }
     ResultOrError<Ref<PipelineLayoutBase>> Device::CreatePipelineLayoutImpl(
         const PipelineLayoutDescriptor* descriptor) {
@@ -285,16 +285,15 @@
         const TextureViewDescriptor* descriptor) {
         return TextureView::Create(texture, descriptor);
     }
-    void Device::CreateComputePipelineAsyncImpl(const ComputePipelineDescriptor* descriptor,
-                                                size_t blueprintHash,
-                                                WGPUCreateComputePipelineAsyncCallback callback,
-                                                void* userdata) {
-        ComputePipeline::CreateAsync(this, descriptor, blueprintHash, callback, userdata);
+    void Device::InitializeComputePipelineAsyncImpl(Ref<ComputePipelineBase> computePipeline,
+                                                    WGPUCreateComputePipelineAsyncCallback callback,
+                                                    void* userdata) {
+        ComputePipeline::InitializeAsync(std::move(computePipeline), callback, userdata);
     }
     void Device::InitializeRenderPipelineAsyncImpl(Ref<RenderPipelineBase> renderPipeline,
                                                    WGPUCreateRenderPipelineAsyncCallback callback,
                                                    void* userdata) {
-        RenderPipeline::InitializeAsync(renderPipeline, callback, userdata);
+        RenderPipeline::InitializeAsync(std::move(renderPipeline), callback, userdata);
     }
 
     ResultOrError<ExecutionSerial> Device::CheckAndUpdateCompletedSerials() {
diff --git a/src/dawn_native/null/DeviceNull.cpp b/src/dawn_native/null/DeviceNull.cpp
index 17030dd..4455226 100644
--- a/src/dawn_native/null/DeviceNull.cpp
+++ b/src/dawn_native/null/DeviceNull.cpp
@@ -112,7 +112,7 @@
         const CommandBufferDescriptor* descriptor) {
         return AcquireRef(new CommandBuffer(encoder, descriptor));
     }
-    ResultOrError<Ref<ComputePipelineBase>> Device::CreateComputePipelineImpl(
+    Ref<ComputePipelineBase> Device::CreateUninitializedComputePipelineImpl(
         const ComputePipelineDescriptor* descriptor) {
         return AcquireRef(new ComputePipeline(this, descriptor));
     }
@@ -371,6 +371,11 @@
         return {};
     }
 
+    // ComputePipeline
+    MaybeError ComputePipeline::Initialize() {
+        return {};
+    }
+
     // RenderPipeline
     MaybeError RenderPipeline::Initialize() {
         return {};
diff --git a/src/dawn_native/null/DeviceNull.h b/src/dawn_native/null/DeviceNull.h
index 0f29139..9e9200c 100644
--- a/src/dawn_native/null/DeviceNull.h
+++ b/src/dawn_native/null/DeviceNull.h
@@ -43,7 +43,7 @@
     class BindGroupLayout;
     class Buffer;
     class CommandBuffer;
-    using ComputePipeline = ComputePipelineBase;
+    class ComputePipeline;
     class Device;
     using PipelineLayout = PipelineLayoutBase;
     class QuerySet;
@@ -129,7 +129,7 @@
             PipelineCompatibilityToken pipelineCompatibilityToken) override;
         ResultOrError<Ref<BufferBase>> CreateBufferImpl(
             const BufferDescriptor* descriptor) override;
-        ResultOrError<Ref<ComputePipelineBase>> CreateComputePipelineImpl(
+        Ref<ComputePipelineBase> CreateUninitializedComputePipelineImpl(
             const ComputePipelineDescriptor* descriptor) override;
         ResultOrError<Ref<PipelineLayoutBase>> CreatePipelineLayoutImpl(
             const PipelineLayoutDescriptor* descriptor) override;
@@ -261,6 +261,13 @@
                                    size_t size) override;
     };
 
+    class ComputePipeline final : public ComputePipelineBase {
+      public:
+        using ComputePipelineBase::ComputePipelineBase;
+
+        MaybeError Initialize() override;
+    };
+
     class RenderPipeline final : public RenderPipelineBase {
       public:
         using RenderPipelineBase::RenderPipelineBase;
diff --git a/src/dawn_native/opengl/ComputePipelineGL.cpp b/src/dawn_native/opengl/ComputePipelineGL.cpp
index e51e217..086ca7a 100644
--- a/src/dawn_native/opengl/ComputePipelineGL.cpp
+++ b/src/dawn_native/opengl/ComputePipelineGL.cpp
@@ -19,12 +19,10 @@
 namespace dawn_native { namespace opengl {
 
     // static
-    ResultOrError<Ref<ComputePipeline>> ComputePipeline::Create(
+    Ref<ComputePipeline> ComputePipeline::CreateUninitialized(
         Device* device,
         const ComputePipelineDescriptor* descriptor) {
-        Ref<ComputePipeline> pipeline = AcquireRef(new ComputePipeline(device, descriptor));
-        DAWN_TRY(pipeline->Initialize());
-        return pipeline;
+        return AcquireRef(new ComputePipeline(device, descriptor));
     }
 
     ComputePipeline::~ComputePipeline() {
diff --git a/src/dawn_native/opengl/ComputePipelineGL.h b/src/dawn_native/opengl/ComputePipelineGL.h
index dd5c0b3..444a289 100644
--- a/src/dawn_native/opengl/ComputePipelineGL.h
+++ b/src/dawn_native/opengl/ComputePipelineGL.h
@@ -27,16 +27,17 @@
 
     class ComputePipeline final : public ComputePipelineBase, public PipelineGL {
       public:
-        static ResultOrError<Ref<ComputePipeline>> Create(
+        static Ref<ComputePipeline> CreateUninitialized(
             Device* device,
             const ComputePipelineDescriptor* descriptor);
 
         void ApplyNow();
 
+        MaybeError Initialize() override;
+
       private:
         using ComputePipelineBase::ComputePipelineBase;
         ~ComputePipeline() override;
-        MaybeError Initialize() override;
     };
 
 }}  // namespace dawn_native::opengl
diff --git a/src/dawn_native/opengl/DeviceGL.cpp b/src/dawn_native/opengl/DeviceGL.cpp
index aa2fa18..08e544c 100644
--- a/src/dawn_native/opengl/DeviceGL.cpp
+++ b/src/dawn_native/opengl/DeviceGL.cpp
@@ -130,9 +130,9 @@
         const CommandBufferDescriptor* descriptor) {
         return AcquireRef(new CommandBuffer(encoder, descriptor));
     }
-    ResultOrError<Ref<ComputePipelineBase>> Device::CreateComputePipelineImpl(
+    Ref<ComputePipelineBase> Device::CreateUninitializedComputePipelineImpl(
         const ComputePipelineDescriptor* descriptor) {
-        return ComputePipeline::Create(this, descriptor);
+        return ComputePipeline::CreateUninitialized(this, descriptor);
     }
     ResultOrError<Ref<PipelineLayoutBase>> Device::CreatePipelineLayoutImpl(
         const PipelineLayoutDescriptor* descriptor) {
diff --git a/src/dawn_native/opengl/DeviceGL.h b/src/dawn_native/opengl/DeviceGL.h
index ee2ef46..b67f748 100644
--- a/src/dawn_native/opengl/DeviceGL.h
+++ b/src/dawn_native/opengl/DeviceGL.h
@@ -91,14 +91,10 @@
             PipelineCompatibilityToken pipelineCompatibilityToken) override;
         ResultOrError<Ref<BufferBase>> CreateBufferImpl(
             const BufferDescriptor* descriptor) override;
-        ResultOrError<Ref<ComputePipelineBase>> CreateComputePipelineImpl(
-            const ComputePipelineDescriptor* descriptor) override;
         ResultOrError<Ref<PipelineLayoutBase>> CreatePipelineLayoutImpl(
             const PipelineLayoutDescriptor* descriptor) override;
         ResultOrError<Ref<QuerySetBase>> CreateQuerySetImpl(
             const QuerySetDescriptor* descriptor) override;
-        Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
-            const RenderPipelineDescriptor* descriptor) override;
         ResultOrError<Ref<SamplerBase>> CreateSamplerImpl(
             const SamplerDescriptor* descriptor) override;
         ResultOrError<Ref<ShaderModuleBase>> CreateShaderModuleImpl(
@@ -115,6 +111,10 @@
         ResultOrError<Ref<TextureViewBase>> CreateTextureViewImpl(
             TextureBase* texture,
             const TextureViewDescriptor* descriptor) override;
+        Ref<ComputePipelineBase> CreateUninitializedComputePipelineImpl(
+            const ComputePipelineDescriptor* descriptor) override;
+        Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
+            const RenderPipelineDescriptor* descriptor) override;
 
         void InitTogglesFromDriver();
         ResultOrError<ExecutionSerial> CheckAndUpdateCompletedSerials() override;
diff --git a/src/dawn_native/vulkan/ComputePipelineVk.cpp b/src/dawn_native/vulkan/ComputePipelineVk.cpp
index 649ab54..c30f8c2 100644
--- a/src/dawn_native/vulkan/ComputePipelineVk.cpp
+++ b/src/dawn_native/vulkan/ComputePipelineVk.cpp
@@ -25,12 +25,10 @@
 namespace dawn_native { namespace vulkan {
 
     // static
-    ResultOrError<Ref<ComputePipeline>> ComputePipeline::Create(
+    Ref<ComputePipeline> ComputePipeline::CreateUninitialized(
         Device* device,
         const ComputePipelineDescriptor* descriptor) {
-        Ref<ComputePipeline> pipeline = AcquireRef(new ComputePipeline(device, descriptor));
-        DAWN_TRY(pipeline->Initialize());
-        return pipeline;
+        return AcquireRef(new ComputePipeline(device, descriptor));
     }
 
     MaybeError ComputePipeline::Initialize() {
@@ -102,14 +100,11 @@
         return mHandle;
     }
 
-    void ComputePipeline::CreateAsync(Device* device,
-                                      const ComputePipelineDescriptor* descriptor,
-                                      size_t blueprintHash,
-                                      WGPUCreateComputePipelineAsyncCallback callback,
-                                      void* userdata) {
-        Ref<ComputePipeline> pipeline = AcquireRef(new ComputePipeline(device, descriptor));
+    void ComputePipeline::InitializeAsync(Ref<ComputePipelineBase> computePipeline,
+                                          WGPUCreateComputePipelineAsyncCallback callback,
+                                          void* userdata) {
         std::unique_ptr<CreateComputePipelineAsyncTask> asyncTask =
-            std::make_unique<CreateComputePipelineAsyncTask>(pipeline, blueprintHash, callback,
+            std::make_unique<CreateComputePipelineAsyncTask>(std::move(computePipeline), callback,
                                                              userdata);
         CreateComputePipelineAsyncTask::RunAsync(std::move(asyncTask));
     }
diff --git a/src/dawn_native/vulkan/ComputePipelineVk.h b/src/dawn_native/vulkan/ComputePipelineVk.h
index 72e2716..1f2d20d 100644
--- a/src/dawn_native/vulkan/ComputePipelineVk.h
+++ b/src/dawn_native/vulkan/ComputePipelineVk.h
@@ -26,24 +26,23 @@
 
     class ComputePipeline final : public ComputePipelineBase {
       public:
-        static ResultOrError<Ref<ComputePipeline>> Create(
+        static Ref<ComputePipeline> CreateUninitialized(
             Device* device,
             const ComputePipelineDescriptor* descriptor);
-        static void CreateAsync(Device* device,
-                                const ComputePipelineDescriptor* descriptor,
-                                size_t blueprintHash,
-                                WGPUCreateComputePipelineAsyncCallback callback,
-                                void* userdata);
+        static void InitializeAsync(Ref<ComputePipelineBase> computePipeline,
+                                    WGPUCreateComputePipelineAsyncCallback callback,
+                                    void* userdata);
 
         VkPipeline GetHandle() const;
 
+        MaybeError Initialize() override;
+
         // Dawn API
         void SetLabelImpl() override;
 
       private:
         ~ComputePipeline() override;
         using ComputePipelineBase::ComputePipelineBase;
-        MaybeError Initialize() override;
 
         VkPipeline mHandle = VK_NULL_HANDLE;
     };
diff --git a/src/dawn_native/vulkan/DeviceVk.cpp b/src/dawn_native/vulkan/DeviceVk.cpp
index a30a20e..1360d3c 100644
--- a/src/dawn_native/vulkan/DeviceVk.cpp
+++ b/src/dawn_native/vulkan/DeviceVk.cpp
@@ -120,9 +120,9 @@
         const CommandBufferDescriptor* descriptor) {
         return CommandBuffer::Create(encoder, descriptor);
     }
-    ResultOrError<Ref<ComputePipelineBase>> Device::CreateComputePipelineImpl(
+    Ref<ComputePipelineBase> Device::CreateUninitializedComputePipelineImpl(
         const ComputePipelineDescriptor* descriptor) {
-        return ComputePipeline::Create(this, descriptor);
+        return ComputePipeline::CreateUninitialized(this, descriptor);
     }
     ResultOrError<Ref<PipelineLayoutBase>> Device::CreatePipelineLayoutImpl(
         const PipelineLayoutDescriptor* descriptor) {
@@ -162,16 +162,15 @@
         const TextureViewDescriptor* descriptor) {
         return TextureView::Create(texture, descriptor);
     }
-    void Device::CreateComputePipelineAsyncImpl(const ComputePipelineDescriptor* descriptor,
-                                                size_t blueprintHash,
-                                                WGPUCreateComputePipelineAsyncCallback callback,
-                                                void* userdata) {
-        ComputePipeline::CreateAsync(this, descriptor, blueprintHash, callback, userdata);
+    void Device::InitializeComputePipelineAsyncImpl(Ref<ComputePipelineBase> computePipeline,
+                                                    WGPUCreateComputePipelineAsyncCallback callback,
+                                                    void* userdata) {
+        ComputePipeline::InitializeAsync(std::move(computePipeline), callback, userdata);
     }
     void Device::InitializeRenderPipelineAsyncImpl(Ref<RenderPipelineBase> renderPipeline,
                                                    WGPUCreateRenderPipelineAsyncCallback callback,
                                                    void* userdata) {
-        RenderPipeline::InitializeAsync(renderPipeline, callback, userdata);
+        RenderPipeline::InitializeAsync(std::move(renderPipeline), callback, userdata);
     }
 
     MaybeError Device::TickImpl() {
diff --git a/src/dawn_native/vulkan/DeviceVk.h b/src/dawn_native/vulkan/DeviceVk.h
index 3378bc8..f77e0b4 100644
--- a/src/dawn_native/vulkan/DeviceVk.h
+++ b/src/dawn_native/vulkan/DeviceVk.h
@@ -113,14 +113,10 @@
             PipelineCompatibilityToken pipelineCompatibilityToken) override;
         ResultOrError<Ref<BufferBase>> CreateBufferImpl(
             const BufferDescriptor* descriptor) override;
-        ResultOrError<Ref<ComputePipelineBase>> CreateComputePipelineImpl(
-            const ComputePipelineDescriptor* descriptor) override;
         ResultOrError<Ref<PipelineLayoutBase>> CreatePipelineLayoutImpl(
             const PipelineLayoutDescriptor* descriptor) override;
         ResultOrError<Ref<QuerySetBase>> CreateQuerySetImpl(
             const QuerySetDescriptor* descriptor) override;
-        Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
-            const RenderPipelineDescriptor* descriptor) override;
         ResultOrError<Ref<SamplerBase>> CreateSamplerImpl(
             const SamplerDescriptor* descriptor) override;
         ResultOrError<Ref<ShaderModuleBase>> CreateShaderModuleImpl(
@@ -137,10 +133,13 @@
         ResultOrError<Ref<TextureViewBase>> CreateTextureViewImpl(
             TextureBase* texture,
             const TextureViewDescriptor* descriptor) override;
-        void CreateComputePipelineAsyncImpl(const ComputePipelineDescriptor* descriptor,
-                                            size_t blueprintHash,
-                                            WGPUCreateComputePipelineAsyncCallback callback,
-                                            void* userdata) override;
+        Ref<ComputePipelineBase> CreateUninitializedComputePipelineImpl(
+            const ComputePipelineDescriptor* descriptor) override;
+        Ref<RenderPipelineBase> CreateUninitializedRenderPipelineImpl(
+            const RenderPipelineDescriptor* descriptor) override;
+        void InitializeComputePipelineAsyncImpl(Ref<ComputePipelineBase> computePipeline,
+                                                WGPUCreateComputePipelineAsyncCallback callback,
+                                                void* userdata) override;
         void InitializeRenderPipelineAsyncImpl(Ref<RenderPipelineBase> renderPipeline,
                                                WGPUCreateRenderPipelineAsyncCallback callback,
                                                void* userdata) override;