blob: 20b2080ee495124fd18770ca6395bab9ae2bc5ee [file] [log] [blame]
Corentin Wallez4a9ef4e2018-07-18 11:40:26 +02001// Copyright 2017 The Dawn Authors
Corentin Wallez0ba55502017-06-14 15:46:59 -04002//
3// Licensed under the Apache License, Version 2.0 (the "License");
4// you may not use this file except in compliance with the License.
5// You may obtain a copy of the License at
6//
7// http://www.apache.org/licenses/LICENSE-2.0
8//
9// Unless required by applicable law or agreed to in writing, software
10// distributed under the License is distributed on an "AS IS" BASIS,
11// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
12// See the License for the specific language governing permissions and
13// limitations under the License.
14
Corentin Wallez30965a72018-07-24 16:42:33 +020015#ifndef DAWNNATIVE_METAL_COMPUTEPIPELINEMTL_H_
16#define DAWNNATIVE_METAL_COMPUTEPIPELINEMTL_H_
Corentin Wallez0ba55502017-06-14 15:46:59 -040017
Corentin Wallezd37523f2018-07-24 13:53:51 +020018#include "dawn_native/ComputePipeline.h"
Corentin Wallez0ba55502017-06-14 15:46:59 -040019
Corentin Wallez0055d952020-11-16 23:07:56 +000020#include "common/NSRef.h"
21
Corentin Wallez0ba55502017-06-14 15:46:59 -040022#import <Metal/Metal.h>
23
Corentin Wallez49a65d02018-07-24 16:45:45 +020024namespace dawn_native { namespace metal {
Corentin Wallez0ba55502017-06-14 15:46:59 -040025
Corentin Wallez8e335a52018-08-27 23:12:56 +020026 class Device;
27
Rafael Cintronc64242d2020-04-06 18:20:02 +000028 class ComputePipeline final : public ComputePipelineBase {
Corentin Wallezf58d84d2017-11-24 14:12:44 -050029 public:
Jiawei Shao8fd1eb52021-10-13 00:43:05 +000030 static Ref<ComputePipeline> CreateUninitialized(
Corentin Wallez50f99582021-03-31 18:36:32 +000031 Device* device,
32 const ComputePipelineDescriptor* descriptor);
Jiawei Shao8fd1eb52021-10-13 00:43:05 +000033 static void InitializeAsync(Ref<ComputePipelineBase> computePipeline,
34 WGPUCreateComputePipelineAsyncCallback callback,
35 void* userdata);
Corentin Wallez0ba55502017-06-14 15:46:59 -040036
Corentin Wallezf58d84d2017-11-24 14:12:44 -050037 void Encode(id<MTLComputeCommandEncoder> encoder);
38 MTLSize GetLocalWorkGroupSize() const;
Corentin Wallezbfc9cee2019-08-02 18:15:08 +000039 bool RequiresStorageBufferLength() const;
Corentin Wallez0ba55502017-06-14 15:46:59 -040040
Corentin Wallezf58d84d2017-11-24 14:12:44 -050041 private:
Ryan Harrison5c413af2019-12-12 17:51:39 +000042 using ComputePipelineBase::ComputePipelineBase;
Jiawei Shao42448da2021-09-11 09:04:34 +000043 MaybeError Initialize() override;
Ryan Harrison5c413af2019-12-12 17:51:39 +000044
Corentin Wallez0055d952020-11-16 23:07:56 +000045 NSPRef<id<MTLComputePipelineState>> mMtlComputePipelineState;
Corentin Wallezf58d84d2017-11-24 14:12:44 -050046 MTLSize mLocalWorkgroupSize;
Corentin Wallezbfc9cee2019-08-02 18:15:08 +000047 bool mRequiresStorageBufferLength;
Austin Eng5528d0e2021-09-15 18:16:50 +000048 std::vector<uint32_t> mWorkgroupAllocations;
Corentin Wallez0ba55502017-06-14 15:46:59 -040049 };
50
Corentin Wallez49a65d02018-07-24 16:45:45 +020051}} // namespace dawn_native::metal
Corentin Wallez0ba55502017-06-14 15:46:59 -040052
Corentin Wallez30965a72018-07-24 16:42:33 +020053#endif // DAWNNATIVE_METAL_COMPUTEPIPELINEMTL_H_