From 352a589fe08bc7cf9df8a54a606d8a34ef6898f0 Mon Sep 17 00:00:00 2001 From: Ryan Harrison Date: Fri, 17 Jan 2020 20:28:58 +0000 Subject: [PATCH] Add code path to use spvc in Metal backend BUG=dawn:291 Change-Id: Idf20496bac733b14db3b7df7eb86ff0a23a9d826 Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/15161 Commit-Queue: Ryan Harrison Reviewed-by: Kai Ninomiya Reviewed-by: Corentin Wallez --- src/dawn_native/metal/ComputePipelineMTL.mm | 2 +- src/dawn_native/metal/RenderPipelineMTL.mm | 4 +- src/dawn_native/metal/ShaderModuleMTL.h | 2 +- src/dawn_native/metal/ShaderModuleMTL.mm | 76 +++++++++++++++++---- 4 files changed, 66 insertions(+), 18 deletions(-) diff --git a/src/dawn_native/metal/ComputePipelineMTL.mm b/src/dawn_native/metal/ComputePipelineMTL.mm index 0a66866ec7..5e08cf6546 100644 --- a/src/dawn_native/metal/ComputePipelineMTL.mm +++ b/src/dawn_native/metal/ComputePipelineMTL.mm @@ -32,7 +32,7 @@ namespace dawn_native { namespace metal { MaybeError ComputePipeline::Initialize(const ComputePipelineDescriptor* descriptor) { auto mtlDevice = ToBackend(GetDevice())->GetMTLDevice(); - const ShaderModule* computeModule = ToBackend(descriptor->computeStage.module); + ShaderModule* computeModule = ToBackend(descriptor->computeStage.module); const char* computeEntryPoint = descriptor->computeStage.entryPoint; ShaderModule::MetalFunctionData computeData; DAWN_TRY(computeModule->GetFunction(computeEntryPoint, SingleShaderStage::Compute, diff --git a/src/dawn_native/metal/RenderPipelineMTL.mm b/src/dawn_native/metal/RenderPipelineMTL.mm index c0e7af7aa6..7177fff222 100644 --- a/src/dawn_native/metal/RenderPipelineMTL.mm +++ b/src/dawn_native/metal/RenderPipelineMTL.mm @@ -330,7 +330,7 @@ namespace dawn_native { namespace metal { MTLRenderPipelineDescriptor* descriptorMTL = [MTLRenderPipelineDescriptor new]; - const ShaderModule* vertexModule = ToBackend(descriptor->vertexStage.module); + ShaderModule* vertexModule = ToBackend(descriptor->vertexStage.module); const char* vertexEntryPoint = descriptor->vertexStage.entryPoint; ShaderModule::MetalFunctionData vertexData; DAWN_TRY(vertexModule->GetFunction(vertexEntryPoint, SingleShaderStage::Vertex, @@ -341,7 +341,7 @@ namespace dawn_native { namespace metal { mStagesRequiringStorageBufferLength |= wgpu::ShaderStage::Vertex; } - const ShaderModule* fragmentModule = ToBackend(descriptor->fragmentStage->module); + ShaderModule* fragmentModule = ToBackend(descriptor->fragmentStage->module); const char* fragmentEntryPoint = descriptor->fragmentStage->entryPoint; ShaderModule::MetalFunctionData fragmentData; DAWN_TRY(fragmentModule->GetFunction(fragmentEntryPoint, SingleShaderStage::Fragment, diff --git a/src/dawn_native/metal/ShaderModuleMTL.h b/src/dawn_native/metal/ShaderModuleMTL.h index 45df04ff32..ef8dd38f4e 100644 --- a/src/dawn_native/metal/ShaderModuleMTL.h +++ b/src/dawn_native/metal/ShaderModuleMTL.h @@ -46,7 +46,7 @@ namespace dawn_native { namespace metal { MaybeError GetFunction(const char* functionName, SingleShaderStage functionStage, const PipelineLayout* layout, - MetalFunctionData* out) const; + MetalFunctionData* out); private: ShaderModule(Device* device, const ShaderModuleDescriptor* descriptor); diff --git a/src/dawn_native/metal/ShaderModuleMTL.mm b/src/dawn_native/metal/ShaderModuleMTL.mm index 7543569ba8..d817321626 100644 --- a/src/dawn_native/metal/ShaderModuleMTL.mm +++ b/src/dawn_native/metal/ShaderModuleMTL.mm @@ -39,6 +39,20 @@ namespace dawn_native { namespace metal { } } + shaderc_spvc_execution_model ToSpvcExecutionModel(SingleShaderStage stage) { + switch (stage) { + case SingleShaderStage::Vertex: + return shaderc_spvc_execution_model_vertex; + case SingleShaderStage::Fragment: + return shaderc_spvc_execution_model_fragment; + case SingleShaderStage::Compute: + return shaderc_spvc_execution_model_glcompute; + default: + UNREACHABLE(); + return shaderc_spvc_execution_model_invalid; + } + } + shaderc_spvc::CompileOptions GetMSLCompileOptions() { // If these options are changed, the values in DawnSPIRVCrossGLSLFastFuzzer.cpp need to // be updated. @@ -94,7 +108,7 @@ namespace dawn_native { namespace metal { MaybeError ShaderModule::GetFunction(const char* functionName, SingleShaderStage functionStage, const PipelineLayout* layout, - ShaderModule::MetalFunctionData* out) const { + ShaderModule::MetalFunctionData* out) { ASSERT(!IsError()); ASSERT(out); std::unique_ptr compiler_impl; @@ -137,30 +151,58 @@ namespace dawn_native { namespace metal { for (uint32_t binding : IterateBitSet(bgInfo.mask)) { for (auto stage : IterateStages(bgInfo.visibilities[binding])) { uint32_t index = layout->GetBindingIndexInfo(stage)[group][binding]; + if (GetDevice()->IsToggleEnabled(Toggle::UseSpvc)) { + shaderc_spvc_msl_resource_binding mslBinding; + mslBinding.stage = ToSpvcExecutionModel(stage); + mslBinding.desc_set = group; + mslBinding.binding = binding; + mslBinding.msl_buffer = mslBinding.msl_texture = mslBinding.msl_sampler = + index; + DAWN_TRY(CheckSpvcSuccess(mSpvcContext.AddMSLResourceBinding(mslBinding), + "Unable to add MSL Resource Binding")); + } else { + spirv_cross::MSLResourceBinding mslBinding; + mslBinding.stage = SpirvExecutionModelForStage(stage); + mslBinding.desc_set = group; + mslBinding.binding = binding; + mslBinding.msl_buffer = mslBinding.msl_texture = mslBinding.msl_sampler = + index; - spirv_cross::MSLResourceBinding mslBinding; - mslBinding.stage = SpirvExecutionModelForStage(stage); - mslBinding.desc_set = group; - mslBinding.binding = binding; - mslBinding.msl_buffer = mslBinding.msl_texture = mslBinding.msl_sampler = index; - - compiler->add_msl_resource_binding(mslBinding); + compiler->add_msl_resource_binding(mslBinding); + } } } } { - spv::ExecutionModel executionModel = SpirvExecutionModelForStage(functionStage); - auto size = compiler->get_entry_point(functionName, executionModel).workgroup_size; - out->localWorkgroupSize = MTLSizeMake(size.x, size.y, size.z); + if (GetDevice()->IsToggleEnabled(Toggle::UseSpvc)) { + shaderc_spvc_execution_model executionModel = ToSpvcExecutionModel(functionStage); + shaderc_spvc_workgroup_size size; + DAWN_TRY(CheckSpvcSuccess( + mSpvcContext.GetWorkgroupSize(functionName, executionModel, &size), + "Unable to get workgroup size for shader")); + out->localWorkgroupSize = MTLSizeMake(size.x, size.y, size.z); + } else { + spv::ExecutionModel executionModel = SpirvExecutionModelForStage(functionStage); + auto size = compiler->get_entry_point(functionName, executionModel).workgroup_size; + out->localWorkgroupSize = MTLSizeMake(size.x, size.y, size.z); + } } { // SPIRV-Cross also supports re-ordering attributes but it seems to do the correct thing // by default. - std::string msl = compiler->compile(); - NSString* mslSource = [NSString stringWithFormat:@"%s", msl.c_str()]; + NSString* mslSource; + if (GetDevice()->IsToggleEnabled(Toggle::UseSpvc)) { + shaderc_spvc::CompilationResult result; + DAWN_TRY(CheckSpvcSuccess(mSpvcContext.CompileShader(&result), + "Unable to compile shader")); + mslSource = [NSString stringWithFormat:@"%s", result.GetStringOutput().c_str()]; + } else { + std::string msl = compiler->compile(); + mslSource = [NSString stringWithFormat:@"%s", msl.c_str()]; + } auto mtlDevice = ToBackend(GetDevice())->GetMTLDevice(); NSError* error = nil; id library = [mtlDevice newLibraryWithSource:mslSource @@ -187,7 +229,13 @@ namespace dawn_native { namespace metal { [library release]; } - out->needsStorageBufferLength = compiler->needs_buffer_size_buffer(); + if (GetDevice()->IsToggleEnabled(Toggle::UseSpvc)) { + DAWN_TRY( + CheckSpvcSuccess(mSpvcContext.NeedsBufferSizeBuffer(&out->needsStorageBufferLength), + "Unable to determine if shader needs buffer size buffer")); + } else { + out->needsStorageBufferLength = compiler->needs_buffer_size_buffer(); + } return {}; }