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 <rharrison@chromium.org>
Reviewed-by: Kai Ninomiya <kainino@chromium.org>
Reviewed-by: Corentin Wallez <cwallez@chromium.org>
This commit is contained in:
Ryan Harrison 2020-01-17 20:28:58 +00:00 committed by Commit Bot service account
parent f70786126c
commit 352a589fe0
4 changed files with 66 additions and 18 deletions

View File

@ -32,7 +32,7 @@ namespace dawn_native { namespace metal {
MaybeError ComputePipeline::Initialize(const ComputePipelineDescriptor* descriptor) { MaybeError ComputePipeline::Initialize(const ComputePipelineDescriptor* descriptor) {
auto mtlDevice = ToBackend(GetDevice())->GetMTLDevice(); auto mtlDevice = ToBackend(GetDevice())->GetMTLDevice();
const ShaderModule* computeModule = ToBackend(descriptor->computeStage.module); ShaderModule* computeModule = ToBackend(descriptor->computeStage.module);
const char* computeEntryPoint = descriptor->computeStage.entryPoint; const char* computeEntryPoint = descriptor->computeStage.entryPoint;
ShaderModule::MetalFunctionData computeData; ShaderModule::MetalFunctionData computeData;
DAWN_TRY(computeModule->GetFunction(computeEntryPoint, SingleShaderStage::Compute, DAWN_TRY(computeModule->GetFunction(computeEntryPoint, SingleShaderStage::Compute,

View File

@ -330,7 +330,7 @@ namespace dawn_native { namespace metal {
MTLRenderPipelineDescriptor* descriptorMTL = [MTLRenderPipelineDescriptor new]; MTLRenderPipelineDescriptor* descriptorMTL = [MTLRenderPipelineDescriptor new];
const ShaderModule* vertexModule = ToBackend(descriptor->vertexStage.module); ShaderModule* vertexModule = ToBackend(descriptor->vertexStage.module);
const char* vertexEntryPoint = descriptor->vertexStage.entryPoint; const char* vertexEntryPoint = descriptor->vertexStage.entryPoint;
ShaderModule::MetalFunctionData vertexData; ShaderModule::MetalFunctionData vertexData;
DAWN_TRY(vertexModule->GetFunction(vertexEntryPoint, SingleShaderStage::Vertex, DAWN_TRY(vertexModule->GetFunction(vertexEntryPoint, SingleShaderStage::Vertex,
@ -341,7 +341,7 @@ namespace dawn_native { namespace metal {
mStagesRequiringStorageBufferLength |= wgpu::ShaderStage::Vertex; mStagesRequiringStorageBufferLength |= wgpu::ShaderStage::Vertex;
} }
const ShaderModule* fragmentModule = ToBackend(descriptor->fragmentStage->module); ShaderModule* fragmentModule = ToBackend(descriptor->fragmentStage->module);
const char* fragmentEntryPoint = descriptor->fragmentStage->entryPoint; const char* fragmentEntryPoint = descriptor->fragmentStage->entryPoint;
ShaderModule::MetalFunctionData fragmentData; ShaderModule::MetalFunctionData fragmentData;
DAWN_TRY(fragmentModule->GetFunction(fragmentEntryPoint, SingleShaderStage::Fragment, DAWN_TRY(fragmentModule->GetFunction(fragmentEntryPoint, SingleShaderStage::Fragment,

View File

@ -46,7 +46,7 @@ namespace dawn_native { namespace metal {
MaybeError GetFunction(const char* functionName, MaybeError GetFunction(const char* functionName,
SingleShaderStage functionStage, SingleShaderStage functionStage,
const PipelineLayout* layout, const PipelineLayout* layout,
MetalFunctionData* out) const; MetalFunctionData* out);
private: private:
ShaderModule(Device* device, const ShaderModuleDescriptor* descriptor); ShaderModule(Device* device, const ShaderModuleDescriptor* descriptor);

View File

@ -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() { shaderc_spvc::CompileOptions GetMSLCompileOptions() {
// If these options are changed, the values in DawnSPIRVCrossGLSLFastFuzzer.cpp need to // If these options are changed, the values in DawnSPIRVCrossGLSLFastFuzzer.cpp need to
// be updated. // be updated.
@ -94,7 +108,7 @@ namespace dawn_native { namespace metal {
MaybeError ShaderModule::GetFunction(const char* functionName, MaybeError ShaderModule::GetFunction(const char* functionName,
SingleShaderStage functionStage, SingleShaderStage functionStage,
const PipelineLayout* layout, const PipelineLayout* layout,
ShaderModule::MetalFunctionData* out) const { ShaderModule::MetalFunctionData* out) {
ASSERT(!IsError()); ASSERT(!IsError());
ASSERT(out); ASSERT(out);
std::unique_ptr<spirv_cross::CompilerMSL> compiler_impl; std::unique_ptr<spirv_cross::CompilerMSL> compiler_impl;
@ -137,30 +151,58 @@ namespace dawn_native { namespace metal {
for (uint32_t binding : IterateBitSet(bgInfo.mask)) { for (uint32_t binding : IterateBitSet(bgInfo.mask)) {
for (auto stage : IterateStages(bgInfo.visibilities[binding])) { for (auto stage : IterateStages(bgInfo.visibilities[binding])) {
uint32_t index = layout->GetBindingIndexInfo(stage)[group][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; spirv_cross::MSLResourceBinding mslBinding;
mslBinding.stage = SpirvExecutionModelForStage(stage); mslBinding.stage = SpirvExecutionModelForStage(stage);
mslBinding.desc_set = group; mslBinding.desc_set = group;
mslBinding.binding = binding; mslBinding.binding = binding;
mslBinding.msl_buffer = mslBinding.msl_texture = mslBinding.msl_sampler = index; mslBinding.msl_buffer = mslBinding.msl_texture = mslBinding.msl_sampler =
index;
compiler->add_msl_resource_binding(mslBinding); compiler->add_msl_resource_binding(mslBinding);
} }
} }
} }
}
{ {
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); spv::ExecutionModel executionModel = SpirvExecutionModelForStage(functionStage);
auto size = compiler->get_entry_point(functionName, executionModel).workgroup_size; auto size = compiler->get_entry_point(functionName, executionModel).workgroup_size;
out->localWorkgroupSize = MTLSizeMake(size.x, size.y, size.z); out->localWorkgroupSize = MTLSizeMake(size.x, size.y, size.z);
} }
}
{ {
// SPIRV-Cross also supports re-ordering attributes but it seems to do the correct thing // SPIRV-Cross also supports re-ordering attributes but it seems to do the correct thing
// by default. // by default.
std::string msl = compiler->compile(); NSString* mslSource;
NSString* mslSource = [NSString stringWithFormat:@"%s", msl.c_str()]; 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(); auto mtlDevice = ToBackend(GetDevice())->GetMTLDevice();
NSError* error = nil; NSError* error = nil;
id<MTLLibrary> library = [mtlDevice newLibraryWithSource:mslSource id<MTLLibrary> library = [mtlDevice newLibraryWithSource:mslSource
@ -187,7 +229,13 @@ namespace dawn_native { namespace metal {
[library release]; [library release];
} }
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(); out->needsStorageBufferLength = compiler->needs_buffer_size_buffer();
}
return {}; return {};
} }