diff --git a/Docs/MoltenVK_Runtime_UserGuide.md b/Docs/MoltenVK_Runtime_UserGuide.md index 8310e03ae..d657a6522 100644 --- a/Docs/MoltenVK_Runtime_UserGuide.md +++ b/Docs/MoltenVK_Runtime_UserGuide.md @@ -248,7 +248,9 @@ In addition to the core *Vulkan* API, **MoltenVK** also supports the following - `VK_KHR_swapchain` - `VK_KHR_swapchain_mutable_format` - `VK_KHR_variable_pointers` +- `VK_EXT_debug_marker` - `VK_EXT_debug_report` +- `VK_EXT_debug_utils` - `VK_EXT_host_query_reset` - `VK_EXT_memory_budget` - `VK_EXT_shader_viewport_index_layer` @@ -317,6 +319,13 @@ where `HEADER_FILE` is one of the following: These functions are exposed in this header for your own purposes such as interacting with *Metal* directly, or simply logging data values. +>***Note:*** The functions in `vk_mvk_moltenvk.h` are not supported by the *Vulkan SDK Loader and Layers* + framework. The opaque Vulkan objects used by the functions in `vk_mvk_moltenvk.h` (`VkInstance`, + `VkPhysicalDevice`, `VkShaderModule`, `VKImage`, ...), must have been retrieved directly from **MoltenVK**, + and not through the *Vulkan SDK Loader and Layers* framework. The *Vulkan SDK Loader and Layers* framework + often changes these opaque objects, and passing them from a higher layer directly to **MoltenVK** will + result in undefined behaviour. + ### Configuring MoltenVK diff --git a/Docs/Whats_New.md b/Docs/Whats_New.md index 024ee5a35..96ab18ecd 100644 --- a/Docs/Whats_New.md +++ b/Docs/Whats_New.md @@ -18,11 +18,30 @@ MoltenVK 1.0.36 Released TBD +- For shaders created directly from MSL, set function name from + `VkPipelineShaderStageCreateInfo::pName`. - On iOS GPU family 2 and earlier, support immutable depth-compare samplers as constexpr samplers hardcoded in MSL. -- Add MTLCommandBuffer completion timing performance tracking option. +- Add `MTLCommandBuffer` completion timing performance tracking option. - Expand `MVK_CONFIG_TRACE_VULKAN_CALLS` to optionally log Vulkan call timings. - Skip `SPIRV-Tools` build in Travis because Travis does not support the required Python 3. +- Separate `SPIRVToMSLConverterContext` into input config and output results. +- Fix pipeline cache lookups. +- Doument that the functions in `vk_mvk_moltenvk.h` cannot be used with objects + retrieved through the *Vulkan SDK Loader and Layers* framework. +- Update `VK_MVK_MOLTENVK_SPEC_VERSION` to 21. +- Update to latest SPIRV-Cross version: + - MSL: Add support for `SubgroupSize` / `SubgroupInvocationID` in fragment. + - MSL: Support `OpImageQueryLod`. + - MSL: Support `MinLod` operand. + - MSL: Support `PrimitiveID` in fragment and barycentrics. + - MSL: Support 64-bit integers. + - MSL: New SDK errors out on cull distance. + - MSL: Conditionally validate MSL 2.2 shaders. + - MSL: Rewrite how resource indices are fallback-assigned. + - MSL: Support custom bindings for argument buffers. + - MSL: Fix sampling with FP16 coordinates. + - MSL: Deal with scalar input values for distance/length/normalize. @@ -570,7 +589,7 @@ MoltenVK 1.0.17 Released 2018/07/31 - Disable rasterization and return void from vertex shaders that write to resources. -- Add SPIRVToMSLConverterOptions::isRasterizationDisabled to allow pipeline and +- Add SPIRVToMSLConversionOptions::isRasterizationDisabled to allow pipeline and vertex shader to communicate rasterization status. - Track layered rendering capability. - Add MVKPhysicalDeviceMetalFeatures::layeredRendering. diff --git a/ExternalRevisions/SPIRV-Cross_repo_revision b/ExternalRevisions/SPIRV-Cross_repo_revision index 30297b1d5..a10b1b1ff 100644 --- a/ExternalRevisions/SPIRV-Cross_repo_revision +++ b/ExternalRevisions/SPIRV-Cross_repo_revision @@ -1 +1 @@ -4104e363005a079acc215f0920743a8affb31278 +8ee8e60f70f937c72379ab1fc404a1c36d660a31 diff --git a/MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h b/MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h index 019811790..cf0f232bc 100644 --- a/MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h +++ b/MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h @@ -666,6 +666,12 @@ typedef void (VKAPI_PTR *PFN_vkGetIOSurfaceMVK)(VkImage image, IOSurfaceRef* pIO * that MoltenVK expects the size of MVKConfiguration to be by setting the value of pConfiguration * to NULL. In that case, this function will set *pConfigurationSize to the size that MoltenVK * expects MVKConfiguration to be. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkInstance object you provide here must have been retrieved directly from MoltenVK, + * and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan objects + * are often changed by layers, and passing them from one layer to another, or from + * a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR VkResult VKAPI_CALL vkGetMoltenVKConfigurationMVK( VkInstance instance, @@ -700,6 +706,12 @@ VKAPI_ATTR VkResult VKAPI_CALL vkGetMoltenVKConfigurationMVK( * that MoltenVK expects the size of MVKConfiguration to be by setting the value of pConfiguration * to NULL. In that case, this function will set *pConfigurationSize to the size that MoltenVK * expects MVKConfiguration to be. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkInstance object you provide here must have been retrieved directly from MoltenVK, + * and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan objects + * are often changed by layers, and passing them from one layer to another, or from + * a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR VkResult VKAPI_CALL vkSetMoltenVKConfigurationMVK( VkInstance instance, @@ -728,6 +740,12 @@ VKAPI_ATTR VkResult VKAPI_CALL vkSetMoltenVKConfigurationMVK( * expects the size of MVKPhysicalDeviceMetalFeatures to be by setting the value of pMetalFeatures to NULL. * In that case, this function will set *pMetalFeaturesSize to the size that MoltenVK expects * MVKPhysicalDeviceMetalFeatures to be. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkPhysicalDevice object you provide here must have been retrieved directly from + * MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR VkResult VKAPI_CALL vkGetPhysicalDeviceMetalFeaturesMVK( VkPhysicalDevice physicalDevice, @@ -755,6 +773,12 @@ VKAPI_ATTR VkResult VKAPI_CALL vkGetPhysicalDeviceMetalFeaturesMVK( * that MoltenVK expects the size of MVKSwapchainPerformance to be by setting the value of * pSwapchainPerf to NULL. In that case, this function will set *pSwapchainPerfSize to the * size that MoltenVK expects MVKSwapchainPerformance to be. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkDevice and VkSwapchainKHR objects you provide here must have been retrieved directly + * from MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR VkResult VKAPI_CALL vkGetSwapchainPerformanceMVK( VkDevice device, @@ -783,6 +807,12 @@ VKAPI_ATTR VkResult VKAPI_CALL vkGetSwapchainPerformanceMVK( * that MoltenVK expects the size of MVKPerformanceStatistics to be by setting the value of * pPerf to NULL. In that case, this function will set *pPerfSize to the size that MoltenVK * expects MVKPerformanceStatistics to be. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkDevice object you provide here must have been retrieved directly from + * MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR VkResult VKAPI_CALL vkGetPerformanceStatisticsMVK( VkDevice device, @@ -808,6 +838,12 @@ VKAPI_ATTR void VKAPI_CALL vkGetVersionStringsMVK( * This needs to be called if you are creating compute shader modules from MSL * source code or MSL compiled code. Workgroup size is determined automatically * if you're using SPIR-V. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkShaderModule object you provide here must have been retrieved directly from + * MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR void VKAPI_CALL vkSetWorkgroupSizeMVK( VkShaderModule shaderModule, @@ -817,7 +853,15 @@ VKAPI_ATTR void VKAPI_CALL vkSetWorkgroupSizeMVK( #ifdef __OBJC__ -/** Returns, in the pMTLDevice pointer, the MTLDevice used by the VkPhysicalDevice. */ +/** + * Returns, in the pMTLDevice pointer, the MTLDevice used by the VkPhysicalDevice. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkPhysicalDevice object you provide here must have been retrieved directly from + * MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. + */ VKAPI_ATTR void VKAPI_CALL vkGetMTLDeviceMVK( VkPhysicalDevice physicalDevice, id* pMTLDevice); @@ -831,12 +875,26 @@ VKAPI_ATTR void VKAPI_CALL vkGetMTLDeviceMVK( * If a MTLTexture has already been created for this image, it will be destroyed. * * Returns VK_SUCCESS. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkImage object you provide here must have been retrieved directly from + * MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR VkResult VKAPI_CALL vkSetMTLTextureMVK( VkImage image, id mtlTexture); -/** Returns, in the pMTLTexture pointer, the MTLTexture currently underlaying the VkImage. */ +/** + * Returns, in the pMTLTexture pointer, the MTLTexture currently underlaying the VkImage. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkImage object you provide here must have been retrieved directly from + * MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. + */ VKAPI_ATTR void VKAPI_CALL vkGetMTLTextureMVK( VkImage image, id* pMTLTexture); @@ -864,6 +922,12 @@ VKAPI_ATTR void VKAPI_CALL vkGetMTLTextureMVK( * - VK_SUCCESS. * - VK_ERROR_FEATURE_NOT_PRESENT if IOSurfaces are not supported on the platform. * - VK_ERROR_INITIALIZATION_FAILED if ioSurface is specified and is not compatible with this VkImage. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkImage object you provide here must have been retrieved directly from + * MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR VkResult VKAPI_CALL vkUseIOSurfaceMVK( VkImage image, @@ -873,6 +937,12 @@ VKAPI_ATTR VkResult VKAPI_CALL vkUseIOSurfaceMVK( * Returns, in the pIOSurface pointer, the IOSurface currently underlaying the VkImage, * as set by the useIOSurfaceMVK() function, or returns null if the VkImage is not using * an IOSurface, or if the platform does not support IOSurfaces. + * + * This function is not supported by the Vulkan SDK Loader and Layers framework. + * The VkImage object you provide here must have been retrieved directly from + * MoltenVK, and not through the Vulkan SDK Loader and Layers framework. Opaque Vulkan + * objects are often changed by layers, and passing them from one layer to another, + * or from a layer directly to MoltenVK, will result in undefined behaviour. */ VKAPI_ATTR void VKAPI_CALL vkGetIOSurfaceMVK( VkImage image, diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKDescriptorSet.h b/MoltenVK/MoltenVK/GPUObjects/MVKDescriptorSet.h index 755199eb6..90cf29212 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKDescriptorSet.h +++ b/MoltenVK/MoltenVK/GPUObjects/MVKDescriptorSet.h @@ -93,7 +93,7 @@ class MVKDescriptorSetLayoutBinding : public MVKBaseDeviceObject { MVKShaderResourceBinding& dslMTLRezIdxOffsets); /** Populates the specified shader converter context, at the specified descriptor set binding. */ - void populateShaderConverterContext(mvk::SPIRVToMSLConverterContext& context, + void populateShaderConverterContext(mvk::SPIRVToMSLConversionConfiguration& context, MVKShaderResourceBinding& dslMTLRezIdxOffsets, uint32_t dslIndex); @@ -160,7 +160,7 @@ class MVKDescriptorSetLayout : public MVKVulkanAPIDeviceObject { /** Populates the specified shader converter context, at the specified DSL index. */ - void populateShaderConverterContext(mvk::SPIRVToMSLConverterContext& context, + void populateShaderConverterContext(mvk::SPIRVToMSLConversionConfiguration& context, MVKShaderResourceBinding& dslMTLRezIdxOffsets, uint32_t dslIndex); @@ -427,7 +427,7 @@ void mvkUpdateDescriptorSetWithTemplate(VkDescriptorSet descriptorSet, * If the shader stage binding has a binding defined for the specified stage, populates * the context at the descriptor set binding from the shader stage resource binding. */ -void mvkPopulateShaderConverterContext(mvk::SPIRVToMSLConverterContext& context, +void mvkPopulateShaderConverterContext(mvk::SPIRVToMSLConversionConfiguration& context, MVKShaderStageResourceBinding& ssRB, spv::ExecutionModel stage, uint32_t descriptorSetIndex, diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKDescriptorSet.mm b/MoltenVK/MoltenVK/GPUObjects/MVKDescriptorSet.mm index 93ab0ae7c..832f460cf 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKDescriptorSet.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKDescriptorSet.mm @@ -375,7 +375,7 @@ return true; } -void MVKDescriptorSetLayoutBinding::populateShaderConverterContext(SPIRVToMSLConverterContext& context, +void MVKDescriptorSetLayoutBinding::populateShaderConverterContext(SPIRVToMSLConversionConfiguration& context, MVKShaderResourceBinding& dslMTLRezIdxOffsets, uint32_t dslIndex) { @@ -632,7 +632,7 @@ } } -void MVKDescriptorSetLayout::populateShaderConverterContext(SPIRVToMSLConverterContext& context, +void MVKDescriptorSetLayout::populateShaderConverterContext(SPIRVToMSLConversionConfiguration& context, MVKShaderResourceBinding& dslMTLRezIdxOffsets, uint32_t dslIndex) { uint32_t bindCnt = (uint32_t)_bindings.size(); @@ -1142,7 +1142,7 @@ void mvkUpdateDescriptorSetWithTemplate(VkDescriptorSet descriptorSet, } } -void mvkPopulateShaderConverterContext(SPIRVToMSLConverterContext& context, +void mvkPopulateShaderConverterContext(SPIRVToMSLConversionConfiguration& context, MVKShaderStageResourceBinding& ssRB, spv::ExecutionModel stage, uint32_t descriptorSetIndex, diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm b/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm index fd0ab2ec0..58647c0a1 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm @@ -1568,7 +1568,7 @@ MVKLogInfo(logMsg.c_str(), _properties.deviceName, devTypeStr.c_str(), _properties.vendorID, _properties.deviceID, [[[NSUUID alloc] initWithUUIDBytes: _properties.pipelineCacheUUID] autorelease].UUIDString.UTF8String, - SPIRVToMSLConverterOptions::printMSLVersion(_metalFeatures.mslVersion).c_str()); + SPIRVToMSLConversionOptions::printMSLVersion(_metalFeatures.mslVersion).c_str()); } MVKPhysicalDevice::~MVKPhysicalDevice() { diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.h b/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.h index ebbaa3507..a6ecc88ad 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.h +++ b/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.h @@ -59,7 +59,7 @@ class MVKPipelineLayout : public MVKVulkanAPIDeviceObject { MVKVector& dynamicOffsets); /** Populates the specified shader converter context. */ - void populateShaderConverterContext(SPIRVToMSLConverterContext& context); + void populateShaderConverterContext(SPIRVToMSLConversionConfiguration& context); /** Updates a descriptor set in a command encoder. */ void pushDescriptorSet(MVKCommandEncoder* cmdEncoder, @@ -232,22 +232,18 @@ class MVKGraphicsPipeline : public MVKPipeline { id getOrCompilePipeline(MTLRenderPipelineDescriptor* plDesc, id& plState); id getOrCompilePipeline(MTLComputePipelineDescriptor* plDesc, id& plState, const char* compilerType); void initMTLRenderPipelineState(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData); - void initMVKShaderConverterContext(SPIRVToMSLConverterContext& _shaderContext, - const VkGraphicsPipelineCreateInfo* pCreateInfo, - const SPIRVTessReflectionData& reflectData); - void addVertexInputToShaderConverterContext(SPIRVToMSLConverterContext& shaderContext, - const VkGraphicsPipelineCreateInfo* pCreateInfo); - void addPrevStageOutputToShaderConverterContext(SPIRVToMSLConverterContext& shaderContext, - std::vector& outputs); + void initMVKShaderConverterContext(SPIRVToMSLConversionConfiguration& _shaderContext, const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData); + void addVertexInputToShaderConverterContext(SPIRVToMSLConversionConfiguration& shaderContext, const VkGraphicsPipelineCreateInfo* pCreateInfo); + void addPrevStageOutputToShaderConverterContext(SPIRVToMSLConversionConfiguration& shaderContext, std::vector& outputs); MTLRenderPipelineDescriptor* getMTLRenderPipelineDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData); - MTLRenderPipelineDescriptor* getMTLTessVertexStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConverterContext& shaderContext); - MTLComputePipelineDescriptor* getMTLTessControlStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConverterContext& shaderContext); - MTLRenderPipelineDescriptor* getMTLTessRasterStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConverterContext& shaderContext); - bool addVertexShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConverterContext& shaderContext); - bool addTessCtlShaderToPipeline(MTLComputePipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConverterContext& shaderContext, std::vector& prevOutput); - bool addTessEvalShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConverterContext& shaderContext, std::vector& prevOutput); - bool addFragmentShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConverterContext& shaderContext); - bool addVertexInputToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkPipelineVertexInputStateCreateInfo* pVI, const SPIRVToMSLConverterContext& shaderContext); + MTLRenderPipelineDescriptor* getMTLTessVertexStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConversionConfiguration& shaderContext); + MTLComputePipelineDescriptor* getMTLTessControlStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConversionConfiguration& shaderContext); + MTLRenderPipelineDescriptor* getMTLTessRasterStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConversionConfiguration& shaderContext); + bool addVertexShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConversionConfiguration& shaderContext); + bool addTessCtlShaderToPipeline(MTLComputePipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConversionConfiguration& shaderContext, std::vector& prevOutput); + bool addTessEvalShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConversionConfiguration& shaderContext, std::vector& prevOutput); + bool addFragmentShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConversionConfiguration& shaderContext); + bool addVertexInputToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkPipelineVertexInputStateCreateInfo* pVI, const SPIRVToMSLConversionConfiguration& shaderContext); void addTessellationToPipeline(MTLRenderPipelineDescriptor* plDesc, const SPIRVTessReflectionData& reflectData, const VkPipelineTessellationStateCreateInfo* pTS); void addFragmentOutputToPipeline(MTLRenderPipelineDescriptor* plDesc, const SPIRVTessReflectionData& reflectData, const VkGraphicsPipelineCreateInfo* pCreateInfo, bool isTessellationVertexPipeline = false); bool isRenderingPoints(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData); @@ -356,7 +352,7 @@ class MVKPipelineCache : public MVKVulkanAPIDeviceObject { VkResult writeData(size_t* pDataSize, void* pData); /** Return a shader library from the specified shader context sourced from the specified shader module. */ - MVKShaderLibrary* getShaderLibrary(SPIRVToMSLConverterContext* pContext, MVKShaderModule* shaderModule); + MVKShaderLibrary* getShaderLibrary(SPIRVToMSLConversionConfiguration* pContext, MVKShaderModule* shaderModule); /** Merges the contents of the specified number of pipeline caches into this cache. */ VkResult mergePipelineCaches(uint32_t srcCacheCount, const VkPipelineCache* pSrcCaches); diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm b/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm index 0d31459ea..237b1b0c4 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm @@ -93,7 +93,7 @@ } } -void MVKPipelineLayout::populateShaderConverterContext(SPIRVToMSLConverterContext& context) { +void MVKPipelineLayout::populateShaderConverterContext(SPIRVToMSLConversionConfiguration& context) { context.resourceBindings.clear(); // Add resource bindings defined in the descriptor set layouts @@ -432,8 +432,9 @@ // In this case, we need to create three render pipelines. But, the way Metal handles // index buffers for compute stage-in means we have to defer creation of stage 2 until // draw time. In the meantime, we'll create and retain a descriptor for it. - SPIRVToMSLConverterContext shaderContext; + SPIRVToMSLConversionConfiguration shaderContext; initMVKShaderConverterContext(shaderContext, pCreateInfo, reflectData); + MTLRenderPipelineDescriptor* vtxPLDesc = getMTLTessVertexStageDescriptor(pCreateInfo, reflectData, shaderContext); _mtlTessControlStageDesc = getMTLTessControlStageDescriptor(pCreateInfo, reflectData, shaderContext); // retained MTLRenderPipelineDescriptor* rastPLDesc = getMTLTessRasterStageDescriptor(pCreateInfo, reflectData, shaderContext); @@ -446,8 +447,9 @@ } // Returns a MTLRenderPipelineDescriptor constructed from this instance, or nil if an error occurs. -MTLRenderPipelineDescriptor* MVKGraphicsPipeline::getMTLRenderPipelineDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData) { - SPIRVToMSLConverterContext shaderContext; +MTLRenderPipelineDescriptor* MVKGraphicsPipeline::getMTLRenderPipelineDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, + const SPIRVTessReflectionData& reflectData) { + SPIRVToMSLConversionConfiguration shaderContext; initMVKShaderConverterContext(shaderContext, pCreateInfo, reflectData); MTLRenderPipelineDescriptor* plDesc = [[MTLRenderPipelineDescriptor new] autorelease]; @@ -473,7 +475,9 @@ } // Returns a MTLRenderPipelineDescriptor for the vertex stage of a tessellated draw constructed from this instance, or nil if an error occurs. -MTLRenderPipelineDescriptor* MVKGraphicsPipeline::getMTLTessVertexStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConverterContext& shaderContext) { +MTLRenderPipelineDescriptor* MVKGraphicsPipeline::getMTLTessVertexStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, + const SPIRVTessReflectionData& reflectData, + SPIRVToMSLConversionConfiguration& shaderContext) { MTLRenderPipelineDescriptor* plDesc = [[MTLRenderPipelineDescriptor new] autorelease]; // Add shader stages. @@ -582,7 +586,9 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { } // Returns a MTLComputePipelineDescriptor for the tess. control stage of a tessellated draw constructed from this instance, or nil if an error occurs. -MTLComputePipelineDescriptor* MVKGraphicsPipeline::getMTLTessControlStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConverterContext& shaderContext) { +MTLComputePipelineDescriptor* MVKGraphicsPipeline::getMTLTessControlStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, + const SPIRVTessReflectionData& reflectData, + SPIRVToMSLConversionConfiguration& shaderContext) { MTLComputePipelineDescriptor* plDesc = [MTLComputePipelineDescriptor new]; std::vector vtxOutputs; @@ -627,7 +633,9 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { } // Returns a MTLRenderPipelineDescriptor for the last stage of a tessellated draw constructed from this instance, or nil if an error occurs. -MTLRenderPipelineDescriptor* MVKGraphicsPipeline::getMTLTessRasterStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData, SPIRVToMSLConverterContext& shaderContext) { +MTLRenderPipelineDescriptor* MVKGraphicsPipeline::getMTLTessRasterStageDescriptor(const VkGraphicsPipelineCreateInfo* pCreateInfo, + const SPIRVTessReflectionData& reflectData, + SPIRVToMSLConversionConfiguration& shaderContext) { MTLRenderPipelineDescriptor* plDesc = [[MTLRenderPipelineDescriptor new] autorelease]; std::vector tcOutputs; @@ -750,7 +758,9 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { } // Adds a vertex shader to the pipeline description. -bool MVKGraphicsPipeline::addVertexShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, SPIRVToMSLConverterContext& shaderContext) { +bool MVKGraphicsPipeline::addVertexShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, + const VkGraphicsPipelineCreateInfo* pCreateInfo, + SPIRVToMSLConversionConfiguration& shaderContext) { uint32_t vbCnt = pCreateInfo->pVertexInputState->vertexBindingDescriptionCount; shaderContext.options.entryPointStage = spv::ExecutionModelVertex; shaderContext.options.entryPointName = _pVertexSS->pName; @@ -761,16 +771,20 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { shaderContext.options.mslOptions.capture_output_to_buffer = isTessellationPipeline(); shaderContext.options.mslOptions.disable_rasterization = isTessellationPipeline() || (pCreateInfo->pRasterizationState && (pCreateInfo->pRasterizationState->rasterizerDiscardEnable)); addVertexInputToShaderConverterContext(shaderContext, pCreateInfo); - id mtlFunction = ((MVKShaderModule*)_pVertexSS->module)->getMTLFunction(&shaderContext, _pVertexSS->pSpecializationInfo, _pipelineCache).mtlFunction; - if ( !mtlFunction ) { + + MVKMTLFunction func = ((MVKShaderModule*)_pVertexSS->module)->getMTLFunction(&shaderContext, _pVertexSS->pSpecializationInfo, _pipelineCache); + if ( !func.mtlFunction ) { setConfigurationResult(reportError(VK_ERROR_INVALID_SHADER_NV, "Vertex shader function could not be compiled into pipeline. See previous logged error.")); return false; } - plDesc.vertexFunction = mtlFunction; - plDesc.rasterizationEnabled = !shaderContext.options.mslOptions.disable_rasterization; - _needsVertexSwizzleBuffer = shaderContext.options.needsSwizzleBuffer; - _needsVertexBufferSizeBuffer = shaderContext.options.needsBufferSizeBuffer; - _needsVertexOutputBuffer = shaderContext.options.needsOutputBuffer; + plDesc.vertexFunction = func.mtlFunction; + + auto& funcRslts = func.shaderConversionResults; + plDesc.rasterizationEnabled = !funcRslts.isRasterizationDisabled; + _needsVertexSwizzleBuffer = funcRslts.needsSwizzleBuffer; + _needsVertexBufferSizeBuffer = funcRslts.needsBufferSizeBuffer; + _needsVertexOutputBuffer = funcRslts.needsOutputBuffer; + // If we need the swizzle buffer and there's no place to put it, we're in serious trouble. if (!verifyImplicitBuffer(_needsVertexSwizzleBuffer, _swizzleBufferIndex, kMVKShaderStageVertex, "swizzle", vbCnt)) { return false; @@ -791,7 +805,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { bool MVKGraphicsPipeline::addTessCtlShaderToPipeline(MTLComputePipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, - SPIRVToMSLConverterContext& shaderContext, + SPIRVToMSLConversionConfiguration& shaderContext, std::vector& vtxOutputs) { shaderContext.options.entryPointStage = spv::ExecutionModelTessellationControl; shaderContext.options.entryPointName = _pTessCtlSS->pName; @@ -803,17 +817,21 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { shaderContext.options.mslOptions.buffer_size_buffer_index = _bufferSizeBufferIndex.stages[kMVKShaderStageTessCtl]; shaderContext.options.mslOptions.capture_output_to_buffer = true; addPrevStageOutputToShaderConverterContext(shaderContext, vtxOutputs); - id mtlFunction = ((MVKShaderModule*)_pTessCtlSS->module)->getMTLFunction(&shaderContext, _pTessCtlSS->pSpecializationInfo, _pipelineCache).mtlFunction; - if ( !mtlFunction ) { + + MVKMTLFunction func = ((MVKShaderModule*)_pTessCtlSS->module)->getMTLFunction(&shaderContext, _pTessCtlSS->pSpecializationInfo, _pipelineCache); + if ( !func.mtlFunction ) { setConfigurationResult(reportError(VK_ERROR_INVALID_SHADER_NV, "Tessellation control shader function could not be compiled into pipeline. See previous logged error.")); return false; } - plDesc.computeFunction = mtlFunction; - _needsTessCtlSwizzleBuffer = shaderContext.options.needsSwizzleBuffer; - _needsTessCtlBufferSizeBuffer = shaderContext.options.needsBufferSizeBuffer; - _needsTessCtlOutputBuffer = shaderContext.options.needsOutputBuffer; - _needsTessCtlPatchOutputBuffer = shaderContext.options.needsPatchOutputBuffer; - _needsTessCtlInput = shaderContext.options.needsInputThreadgroupMem; + plDesc.computeFunction = func.mtlFunction; + + auto& funcRslts = func.shaderConversionResults; + _needsTessCtlSwizzleBuffer = funcRslts.needsSwizzleBuffer; + _needsTessCtlBufferSizeBuffer = funcRslts.needsBufferSizeBuffer; + _needsTessCtlOutputBuffer = funcRslts.needsOutputBuffer; + _needsTessCtlPatchOutputBuffer = funcRslts.needsPatchOutputBuffer; + _needsTessCtlInput = funcRslts.needsInputThreadgroupMem; + if (!verifyImplicitBuffer(_needsTessCtlSwizzleBuffer, _swizzleBufferIndex, kMVKShaderStageTessCtl, "swizzle", kMVKTessCtlNumReservedBuffers)) { return false; } @@ -839,7 +857,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { bool MVKGraphicsPipeline::addTessEvalShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, - SPIRVToMSLConverterContext& shaderContext, + SPIRVToMSLConversionConfiguration& shaderContext, std::vector& tcOutputs) { shaderContext.options.entryPointStage = spv::ExecutionModelTessellationEvaluation; shaderContext.options.entryPointName = _pTessEvalSS->pName; @@ -848,16 +866,20 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { shaderContext.options.mslOptions.capture_output_to_buffer = false; shaderContext.options.mslOptions.disable_rasterization = (pCreateInfo->pRasterizationState && (pCreateInfo->pRasterizationState->rasterizerDiscardEnable)); addPrevStageOutputToShaderConverterContext(shaderContext, tcOutputs); - id mtlFunction = ((MVKShaderModule*)_pTessEvalSS->module)->getMTLFunction(&shaderContext, _pTessEvalSS->pSpecializationInfo, _pipelineCache).mtlFunction; - if ( !mtlFunction ) { + + MVKMTLFunction func = ((MVKShaderModule*)_pTessEvalSS->module)->getMTLFunction(&shaderContext, _pTessEvalSS->pSpecializationInfo, _pipelineCache); + if ( !func.mtlFunction ) { setConfigurationResult(reportError(VK_ERROR_INVALID_SHADER_NV, "Tessellation evaluation shader function could not be compiled into pipeline. See previous logged error.")); return false; } // Yeah, you read that right. Tess. eval functions are a kind of vertex function in Metal. - plDesc.vertexFunction = mtlFunction; - plDesc.rasterizationEnabled = !shaderContext.options.mslOptions.disable_rasterization; - _needsTessEvalSwizzleBuffer = shaderContext.options.needsSwizzleBuffer; - _needsTessEvalBufferSizeBuffer = shaderContext.options.needsBufferSizeBuffer; + plDesc.vertexFunction = func.mtlFunction; + + auto& funcRslts = func.shaderConversionResults; + plDesc.rasterizationEnabled = !funcRslts.isRasterizationDisabled; + _needsTessEvalSwizzleBuffer = funcRslts.needsSwizzleBuffer; + _needsTessEvalBufferSizeBuffer = funcRslts.needsBufferSizeBuffer; + if (!verifyImplicitBuffer(_needsTessEvalSwizzleBuffer, _swizzleBufferIndex, kMVKShaderStageTessEval, "swizzle", kMVKTessEvalNumReservedBuffers)) { return false; } @@ -869,21 +891,24 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { bool MVKGraphicsPipeline::addFragmentShaderToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkGraphicsPipelineCreateInfo* pCreateInfo, - SPIRVToMSLConverterContext& shaderContext) { + SPIRVToMSLConversionConfiguration& shaderContext) { if (_pFragmentSS) { shaderContext.options.entryPointStage = spv::ExecutionModelFragment; shaderContext.options.mslOptions.swizzle_buffer_index = _swizzleBufferIndex.stages[kMVKShaderStageFragment]; shaderContext.options.mslOptions.buffer_size_buffer_index = _bufferSizeBufferIndex.stages[kMVKShaderStageFragment]; shaderContext.options.entryPointName = _pFragmentSS->pName; shaderContext.options.mslOptions.capture_output_to_buffer = false; - id mtlFunction = ((MVKShaderModule*)_pFragmentSS->module)->getMTLFunction(&shaderContext, _pFragmentSS->pSpecializationInfo, _pipelineCache).mtlFunction; - if ( !mtlFunction ) { + + MVKMTLFunction func = ((MVKShaderModule*)_pFragmentSS->module)->getMTLFunction(&shaderContext, _pFragmentSS->pSpecializationInfo, _pipelineCache); + if ( !func.mtlFunction ) { setConfigurationResult(reportError(VK_ERROR_INVALID_SHADER_NV, "Fragment shader function could not be compiled into pipeline. See previous logged error.")); return false; } - plDesc.fragmentFunction = mtlFunction; - _needsFragmentSwizzleBuffer = shaderContext.options.needsSwizzleBuffer; - _needsFragmentBufferSizeBuffer = shaderContext.options.needsBufferSizeBuffer; + plDesc.fragmentFunction = func.mtlFunction; + + auto& funcRslts = func.shaderConversionResults; + _needsFragmentSwizzleBuffer = funcRslts.needsSwizzleBuffer; + _needsFragmentBufferSizeBuffer = funcRslts.needsBufferSizeBuffer; if (!verifyImplicitBuffer(_needsFragmentSwizzleBuffer, _swizzleBufferIndex, kMVKShaderStageFragment, "swizzle", 0)) { return false; } @@ -896,7 +921,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { bool MVKGraphicsPipeline::addVertexInputToPipeline(MTLRenderPipelineDescriptor* plDesc, const VkPipelineVertexInputStateCreateInfo* pVI, - const SPIRVToMSLConverterContext& shaderContext) { + const SPIRVToMSLConversionConfiguration& shaderContext) { // Collect extension structures VkPipelineVertexInputDivisorStateCreateInfoEXT* pVertexInputDivisorState = nullptr; auto* next = (MVKVkAPIStructHeader*)pVI->pNext; @@ -1080,7 +1105,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { } // Initializes the context used to prepare the MSL library used by this pipeline. -void MVKGraphicsPipeline::initMVKShaderConverterContext(SPIRVToMSLConverterContext& shaderContext, +void MVKGraphicsPipeline::initMVKShaderConverterContext(SPIRVToMSLConversionConfiguration& shaderContext, const VkGraphicsPipelineCreateInfo* pCreateInfo, const SPIRVTessReflectionData& reflectData) { @@ -1122,7 +1147,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { } // Initializes the vertex attributes in a shader converter context. -void MVKGraphicsPipeline::addVertexInputToShaderConverterContext(SPIRVToMSLConverterContext& shaderContext, +void MVKGraphicsPipeline::addVertexInputToShaderConverterContext(SPIRVToMSLConversionConfiguration& shaderContext, const VkGraphicsPipelineCreateInfo* pCreateInfo) { // Set the shader context vertex attribute information shaderContext.vertexAttributes.clear(); @@ -1186,7 +1211,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { } // Initializes the vertex attributes in a shader converter context from the previous stage output. -void MVKGraphicsPipeline::addPrevStageOutputToShaderConverterContext(SPIRVToMSLConverterContext& shaderContext, +void MVKGraphicsPipeline::addPrevStageOutputToShaderConverterContext(SPIRVToMSLConversionConfiguration& shaderContext, std::vector& shaderOutputs) { // Set the shader context vertex attribute information shaderContext.vertexAttributes.clear(); @@ -1286,7 +1311,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { const VkPipelineShaderStageCreateInfo* pSS = &pCreateInfo->stage; if ( !mvkAreFlagsEnabled(pSS->stage, VK_SHADER_STAGE_COMPUTE_BIT) ) { return MVKMTLFunctionNull; } - SPIRVToMSLConverterContext shaderContext; + SPIRVToMSLConversionConfiguration shaderContext; shaderContext.options.entryPointName = pCreateInfo->stage.pName; shaderContext.options.entryPointStage = spv::ExecutionModelGLCompute; shaderContext.options.mslOptions.msl_version = _device->_pMetalFeatures->mslVersion; @@ -1300,11 +1325,13 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { shaderContext.options.mslOptions.swizzle_buffer_index = _swizzleBufferIndex.stages[kMVKShaderStageCompute]; shaderContext.options.mslOptions.buffer_size_buffer_index = _bufferSizeBufferIndex.stages[kMVKShaderStageCompute]; - MVKShaderModule* mvkShdrMod = (MVKShaderModule*)pSS->module; - MVKMTLFunction func = mvkShdrMod->getMTLFunction(&shaderContext, pSS->pSpecializationInfo, _pipelineCache); - _needsSwizzleBuffer = shaderContext.options.needsSwizzleBuffer; - _needsBufferSizeBuffer = shaderContext.options.needsBufferSizeBuffer; - return func; + MVKMTLFunction func = ((MVKShaderModule*)pSS->module)->getMTLFunction(&shaderContext, pSS->pSpecializationInfo, _pipelineCache); + + auto& funcRslts = func.shaderConversionResults; + _needsSwizzleBuffer = funcRslts.needsSwizzleBuffer; + _needsBufferSizeBuffer = funcRslts.needsBufferSizeBuffer; + + return func; } MVKComputePipeline::~MVKComputePipeline() { @@ -1316,7 +1343,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { #pragma mark MVKPipelineCache // Return a shader library from the specified shader context sourced from the specified shader module. -MVKShaderLibrary* MVKPipelineCache::getShaderLibrary(SPIRVToMSLConverterContext* pContext, MVKShaderModule* shaderModule) { +MVKShaderLibrary* MVKPipelineCache::getShaderLibrary(SPIRVToMSLConversionConfiguration* pContext, MVKShaderModule* shaderModule) { lock_guard lock(_shaderCacheLock); bool wasAdded = false; @@ -1358,9 +1385,9 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { friend MVKPipelineCache; bool next() { return (++_index < (_pSLCache ? _pSLCache->_shaderLibraries.size() : 0)); } - SPIRVToMSLConverterContext& getShaderContext() { return _pSLCache->_shaderLibraries[_index].first; } + SPIRVToMSLConversionConfiguration& getShaderConversionConfig() { return _pSLCache->_shaderLibraries[_index].first; } std::string& getMSL() { return _pSLCache->_shaderLibraries[_index].second->_msl; } - SPIRVEntryPoint& getEntryPoint() { return _pSLCache->_shaderLibraries[_index].second->_entryPoint; } + SPIRVToMSLConversionResults& getShaderConversionResults() { return _pSLCache->_shaderLibraries[_index].second->_shaderConversionResults; } MVKShaderCacheIterator(MVKShaderLibraryCache* pSLCache) : _pSLCache(pSLCache) {} MVKShaderLibraryCache* _pSLCache; @@ -1435,8 +1462,8 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { uint64_t startTime = _device->getPerformanceTimestamp(); writer(cacheEntryType); writer(smKey); - writer(cacheIter.getShaderContext()); - writer(cacheIter.getEntryPoint()); + writer(cacheIter.getShaderConversionConfig()); + writer(cacheIter.getShaderConversionResults()); writer(cacheIter.getMSL()); _device->addActivityPerformance(activityTracker, startTime); } @@ -1492,11 +1519,11 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { MVKShaderModuleKey smKey; reader(smKey); - SPIRVToMSLConverterContext shaderContext; - reader(shaderContext); + SPIRVToMSLConversionConfiguration shaderConversionConfig; + reader(shaderConversionConfig); - SPIRVEntryPoint entryPoint; - reader(entryPoint); + SPIRVToMSLConversionResults shaderConversionResults; + reader(shaderConversionResults); string msl; reader(msl); @@ -1504,7 +1531,7 @@ static VkFormat mvkFormatFromOutput(const SPIRVShaderOutput& output) { // Add the shader library to the staging cache. MVKShaderLibraryCache* slCache = getShaderLibraryCache(smKey); _device->addActivityPerformance(_device->_performanceStatistics.pipelineCache.readPipelineCache, startTime); - slCache->addShaderLibrary(&shaderContext, msl, entryPoint); + slCache->addShaderLibrary(&shaderConversionConfig, msl, shaderConversionResults); break; } @@ -1625,18 +1652,13 @@ void serialize(Archive & archive, SPIRVEntryPoint& ep) { } template - void serialize(Archive & archive, SPIRVToMSLConverterOptions& opt) { + void serialize(Archive & archive, SPIRVToMSLConversionOptions& opt) { archive(opt.mslOptions, opt.entryPointName, opt.entryPointStage, opt.tessPatchKind, opt.numTessControlPoints, - opt.shouldFlipVertexY, - opt.needsSwizzleBuffer, - opt.needsOutputBuffer, - opt.needsPatchOutputBuffer, - opt.needsBufferSizeBuffer, - opt.needsInputThreadgroupMem); + opt.shouldFlipVertexY); } template @@ -1654,10 +1676,21 @@ void serialize(Archive & archive, MSLResourceBinding& rb) { } template - void serialize(Archive & archive, SPIRVToMSLConverterContext& ctx) { + void serialize(Archive & archive, SPIRVToMSLConversionConfiguration& ctx) { archive(ctx.options, ctx.vertexAttributes, ctx.resourceBindings); } + template + void serialize(Archive & archive, SPIRVToMSLConversionResults& scr) { + archive(scr.entryPoint, + scr.isRasterizationDisabled, + scr.needsSwizzleBuffer, + scr.needsOutputBuffer, + scr.needsPatchOutputBuffer, + scr.needsBufferSizeBuffer, + scr.needsInputThreadgroupMem); + } + } template diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKShaderModule.h b/MoltenVK/MoltenVK/GPUObjects/MVKShaderModule.h index d53a4ecc8..ff54bfbcc 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKShaderModule.h +++ b/MoltenVK/MoltenVK/GPUObjects/MVKShaderModule.h @@ -38,10 +38,11 @@ using namespace mvk; #pragma mark - #pragma mark MVKShaderLibrary -/** Specifies the SPIRV LocalSize, which is the number of threads in a compute shader workgroup. */ +/** A MTLFunction and corresponding result information resulting from a shader conversion. */ typedef struct { - id mtlFunction; - MTLSize threadGroupSize; + id mtlFunction; + const SPIRVToMSLConversionResults shaderConversionResults; + MTLSize threadGroupSize; } MVKMTLFunction; /** A MVKMTLFunction indicating an invalid MTLFunction. The mtlFunction member is nil. */ @@ -55,13 +56,28 @@ class MVKShaderLibrary : public MVKBaseObject { /** Returns the Vulkan API opaque object controlling this object. */ MVKVulkanAPIObject* getVulkanAPIObject() override { return _owner->getVulkanAPIObject(); }; - /** Sets the number of threads in a single compute kernel workgroup, per dimension. */ + /** + * Sets the entry point function name. + * + * This is usually set automatically during shader conversion from SPIR-V to MSL. + * For a library that was created directly from MSL, this function can be used to + * set the name of the function if it has a different name than the default main0(). + */ + void setEntryPointName(std::string& funcName); + + /** + * Sets the number of threads in a single compute kernel workgroup, per dimension. + * + * This is usually set automatically during shader conversion from SPIR-V to MSL. + * For a library that was created directly from MSL, this function can be used to + * set the workgroup size.. + */ void setWorkgroupSize(uint32_t x, uint32_t y, uint32_t z); /** Constructs an instance from the specified MSL source code. */ MVKShaderLibrary(MVKVulkanAPIDeviceObject* owner, const std::string& mslSourceCode, - const SPIRVEntryPoint& entryPoint); + const SPIRVToMSLConversionResults& shaderConversionResults); /** Constructs an instance from the specified compiled MSL code data. */ MVKShaderLibrary(MVKVulkanAPIDeviceObject* owner, @@ -84,7 +100,7 @@ class MVKShaderLibrary : public MVKBaseObject { MVKVulkanAPIDeviceObject* _owner; id _mtlLibrary; - SPIRVEntryPoint _entryPoint; + SPIRVToMSLConversionResults _shaderConversionResults; std::string _msl; }; @@ -107,7 +123,7 @@ class MVKShaderLibraryCache : public MVKBaseObject { * If pWasAdded is not nil, this function will set it to true if a new shader library was created, * and to false if an existing shader library was found and returned. */ - MVKShaderLibrary* getShaderLibrary(SPIRVToMSLConverterContext* pContext, + MVKShaderLibrary* getShaderLibrary(SPIRVToMSLConversionConfiguration* pContext, MVKShaderModule* shaderModule, bool* pWasAdded = nullptr); @@ -120,14 +136,14 @@ class MVKShaderLibraryCache : public MVKBaseObject { friend MVKPipelineCache; friend MVKShaderModule; - MVKShaderLibrary* findShaderLibrary(SPIRVToMSLConverterContext* pContext); - MVKShaderLibrary* addShaderLibrary(SPIRVToMSLConverterContext* pContext, + MVKShaderLibrary* findShaderLibrary(SPIRVToMSLConversionConfiguration* pContext); + MVKShaderLibrary* addShaderLibrary(SPIRVToMSLConversionConfiguration* pContext, const std::string& mslSourceCode, - const SPIRVEntryPoint& entryPoint); + const SPIRVToMSLConversionResults& shaderConversionResults); void merge(MVKShaderLibraryCache* other); MVKVulkanAPIDeviceObject* _owner; - std::vector> _shaderLibraries; + std::vector> _shaderLibraries; }; @@ -168,12 +184,12 @@ class MVKShaderModule : public MVKVulkanAPIDeviceObject { VkDebugReportObjectTypeEXT getVkDebugReportObjectType() override { return VK_DEBUG_REPORT_OBJECT_TYPE_SHADER_MODULE_EXT; } /** Returns the Metal shader function, possibly specialized. */ - MVKMTLFunction getMTLFunction(SPIRVToMSLConverterContext* pContext, + MVKMTLFunction getMTLFunction(SPIRVToMSLConversionConfiguration* pContext, const VkSpecializationInfo* pSpecializationInfo, MVKPipelineCache* pipelineCache); /** Convert the SPIR-V to MSL, using the specified shader conversion context. */ - bool convert(SPIRVToMSLConverterContext* pContext); + bool convert(SPIRVToMSLConversionConfiguration* pContext); /** Returns the original SPIR-V code that was specified when this object was created. */ const std::vector& getSPIRV() { return _spvConverter.getSPIRV(); } @@ -184,12 +200,9 @@ class MVKShaderModule : public MVKVulkanAPIDeviceObject { */ const std::string& getMSL() { return _spvConverter.getMSL(); } - /** - * Returns information about the shader entry point as converted by the most recent - * call to convert() function, or set directly using the setMSL() function. - */ - const SPIRVEntryPoint& getEntryPoint() { return _spvConverter.getEntryPoint(); } - + /** Returns information about the shader conversion results. */ + const SPIRVToMSLConversionResults& getConversionResults() { return _spvConverter.getConversionResults(); } + /** Sets the number of threads in a single compute kernel workgroup, per dimension. */ void setWorkgroupSize(uint32_t x, uint32_t y, uint32_t z); @@ -204,12 +217,12 @@ class MVKShaderModule : public MVKVulkanAPIDeviceObject { friend MVKShaderCacheIterator; void propogateDebugName() override {} - MVKGLSLConversionShaderStage getMVKGLSLConversionShaderStage(SPIRVToMSLConverterContext* pContext); + MVKGLSLConversionShaderStage getMVKGLSLConversionShaderStage(SPIRVToMSLConversionConfiguration* pContext); MVKShaderLibraryCache _shaderLibraryCache; SPIRVToMSLConverter _spvConverter; GLSLToSPIRVConverter _glslConverter; - MVKShaderLibrary* _defaultLibrary; + MVKShaderLibrary* _directMSLLibrary; MVKShaderModuleKey _key; std::mutex _accessLock; }; diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKShaderModule.mm b/MoltenVK/MoltenVK/GPUObjects/MVKShaderModule.mm index 426670a40..b80362bc8 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKShaderModule.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKShaderModule.mm @@ -26,7 +26,7 @@ using namespace std; -const MVKMTLFunction MVKMTLFunctionNull = { nil, MTLSizeMake(1, 1, 1) }; +const MVKMTLFunction MVKMTLFunctionNull = { nil, SPIRVToMSLConversionResults(), MTLSizeMake(1, 1, 1) }; #pragma mark - #pragma mark MVKShaderLibrary @@ -42,7 +42,6 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD } } } - return wgDim.size; } @@ -50,7 +49,7 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD if ( !_mtlLibrary ) { return MVKMTLFunctionNull; } - NSString* mtlFuncName = @(_entryPoint.mtlFunctionName.c_str()); + NSString* mtlFuncName = @(_shaderConversionResults.entryPoint.mtlFunctionName.c_str()); MVKDevice* mvkDev = _owner->getDevice(); uint64_t startTime = mvkDev->getPerformanceTimestamp(); id mtlFunc = [[_mtlLibrary newFunctionWithName: mtlFuncName] autorelease]; @@ -96,10 +95,11 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD if ( !dbName ) { dbName = _owner-> getDebugName(); } setLabelIfNotNil(mtlFunc, dbName); - return { mtlFunc, MTLSizeMake(getWorkgroupDimensionSize(_entryPoint.workgroupSize.width, pSpecializationInfo), - getWorkgroupDimensionSize(_entryPoint.workgroupSize.height, pSpecializationInfo), - getWorkgroupDimensionSize(_entryPoint.workgroupSize.depth, pSpecializationInfo)) }; + auto& wgSize = _shaderConversionResults.entryPoint.workgroupSize; + return { mtlFunc, _shaderConversionResults, MTLSizeMake(getWorkgroupDimensionSize(wgSize.width, pSpecializationInfo), + getWorkgroupDimensionSize(wgSize.height, pSpecializationInfo), + getWorkgroupDimensionSize(wgSize.depth, pSpecializationInfo))}; } // Returns the MTLFunctionConstant with the specified ID from the specified array of function constants. @@ -109,12 +109,25 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD return nil; } -MVKShaderLibrary::MVKShaderLibrary(MVKVulkanAPIDeviceObject* owner, const string& mslSourceCode, const SPIRVEntryPoint& entryPoint) : _owner(owner) { +void MVKShaderLibrary::setEntryPointName(string& funcName) { + _shaderConversionResults.entryPoint.mtlFunctionName = funcName; +} + +void MVKShaderLibrary::setWorkgroupSize(uint32_t x, uint32_t y, uint32_t z) { + auto& wgSize = _shaderConversionResults.entryPoint.workgroupSize; + wgSize.width.size = x; + wgSize.height.size = y; + wgSize.depth.size = z; +} + +MVKShaderLibrary::MVKShaderLibrary(MVKVulkanAPIDeviceObject* owner, + const string& mslSourceCode, + const SPIRVToMSLConversionResults& shaderConversionResults) : _owner(owner) { MVKShaderLibraryCompiler* slc = new MVKShaderLibraryCompiler(_owner); _mtlLibrary = slc->newMTLLibrary(@(mslSourceCode.c_str())); // retained slc->destroy(); - _entryPoint = entryPoint; + _shaderConversionResults = shaderConversionResults; _msl = mslSourceCode; } @@ -138,7 +151,7 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD MVKShaderLibrary::MVKShaderLibrary(MVKShaderLibrary& other) : _owner(other._owner) { _mtlLibrary = [other._mtlLibrary retain]; - _entryPoint = other._entryPoint; + _shaderConversionResults = other._shaderConversionResults; _msl = other._msl; } @@ -158,12 +171,6 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD } } -void MVKShaderLibrary::setWorkgroupSize(uint32_t x, uint32_t y, uint32_t z) { - _entryPoint.workgroupSize.width.size = x; - _entryPoint.workgroupSize.height.size = y; - _entryPoint.workgroupSize.depth.size = z; -} - MVKShaderLibrary::~MVKShaderLibrary() { [_mtlLibrary release]; } @@ -172,14 +179,14 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD #pragma mark - #pragma mark MVKShaderLibraryCache -MVKShaderLibrary* MVKShaderLibraryCache::getShaderLibrary(SPIRVToMSLConverterContext* pContext, +MVKShaderLibrary* MVKShaderLibraryCache::getShaderLibrary(SPIRVToMSLConversionConfiguration* pContext, MVKShaderModule* shaderModule, bool* pWasAdded) { bool wasAdded = false; MVKShaderLibrary* shLib = findShaderLibrary(pContext); if ( !shLib ) { if (shaderModule->convert(pContext)) { - shLib = addShaderLibrary(pContext, shaderModule->getMSL(), shaderModule->getEntryPoint()); + shLib = addShaderLibrary(pContext, shaderModule->getMSL(), shaderModule->getConversionResults()); wasAdded = true; } } @@ -191,7 +198,7 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD // Finds and returns a shader library matching the specified context, or returns nullptr if it doesn't exist. // If a match is found, the specified context is aligned with the context of the matching library. -MVKShaderLibrary* MVKShaderLibraryCache::findShaderLibrary(SPIRVToMSLConverterContext* pContext) { +MVKShaderLibrary* MVKShaderLibraryCache::findShaderLibrary(SPIRVToMSLConversionConfiguration* pContext) { for (auto& slPair : _shaderLibraries) { if (slPair.first.matches(*pContext)) { pContext->alignWith(slPair.first); @@ -202,10 +209,10 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD } // Adds and returns a new shader library configured from the specified context. -MVKShaderLibrary* MVKShaderLibraryCache::addShaderLibrary(SPIRVToMSLConverterContext* pContext, +MVKShaderLibrary* MVKShaderLibraryCache::addShaderLibrary(SPIRVToMSLConversionConfiguration* pContext, const string& mslSourceCode, - const SPIRVEntryPoint& entryPoint) { - MVKShaderLibrary* shLib = new MVKShaderLibrary(_owner, mslSourceCode, entryPoint); + const SPIRVToMSLConversionResults& shaderConversionResults) { + MVKShaderLibrary* shLib = new MVKShaderLibrary(_owner, mslSourceCode, shaderConversionResults); _shaderLibraries.emplace_back(*pContext, shLib); return shLib; } @@ -228,12 +235,12 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD #pragma mark - #pragma mark MVKShaderModule -MVKMTLFunction MVKShaderModule::getMTLFunction(SPIRVToMSLConverterContext* pContext, +MVKMTLFunction MVKShaderModule::getMTLFunction(SPIRVToMSLConversionConfiguration* pContext, const VkSpecializationInfo* pSpecializationInfo, MVKPipelineCache* pipelineCache) { lock_guard lock(_accessLock); - MVKShaderLibrary* mvkLib = _defaultLibrary; + MVKShaderLibrary* mvkLib = _directMSLLibrary; if ( !mvkLib ) { uint64_t startTime = _device->getPerformanceTimestamp(); if (pipelineCache) { @@ -243,13 +250,14 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD } _device->addActivityPerformance(_device->_performanceStatistics.shaderCompilation.shaderLibraryFromCache, startTime); } else { + mvkLib->setEntryPointName(pContext->options.entryPointName); pContext->markAllAttributesAndResourcesUsed(); } return mvkLib ? mvkLib->getMTLFunction(pSpecializationInfo, this) : MVKMTLFunctionNull; } -bool MVKShaderModule::convert(SPIRVToMSLConverterContext* pContext) { +bool MVKShaderModule::convert(SPIRVToMSLConversionConfiguration* pContext) { bool shouldLogCode = _device->_pMVKConfig->debugMode; bool shouldLogEstimatedGLSL = shouldLogCode; @@ -283,7 +291,7 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD } // Returns the MVKGLSLConversionShaderStage corresponding to the shader stage in the SPIR-V converter context. -MVKGLSLConversionShaderStage MVKShaderModule::getMVKGLSLConversionShaderStage(SPIRVToMSLConverterContext* pContext) { +MVKGLSLConversionShaderStage MVKShaderModule::getMVKGLSLConversionShaderStage(SPIRVToMSLConversionConfiguration* pContext) { switch (pContext->options.entryPointStage) { case spv::ExecutionModelVertex: return kMVKGLSLConversionShaderStageVertex; case spv::ExecutionModelTessellationControl: return kMVKGLSLConversionShaderStageTessControl; @@ -299,12 +307,18 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD } } +void MVKShaderModule::setWorkgroupSize(uint32_t x, uint32_t y, uint32_t z) { + _spvConverter.setWorkgroupSize(x, y, z); + if(_directMSLLibrary) { _directMSLLibrary->setWorkgroupSize(x, y, z); } +} + + #pragma mark Construction MVKShaderModule::MVKShaderModule(MVKDevice* device, const VkShaderModuleCreateInfo* pCreateInfo) : MVKVulkanAPIDeviceObject(device), _shaderLibraryCache(this) { - _defaultLibrary = nullptr; + _directMSLLibrary = nullptr; size_t codeSize = pCreateInfo->codeSize; @@ -341,7 +355,7 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD _device->addActivityPerformance(_device->_performanceStatistics.shaderCompilation.hashShaderCode, startTime); _spvConverter.setMSL(pMSLCode, nullptr); - _defaultLibrary = new MVKShaderLibrary(this, _spvConverter.getMSL().c_str(), _spvConverter.getEntryPoint()); + _directMSLLibrary = new MVKShaderLibrary(this, _spvConverter.getMSL().c_str(), _spvConverter.getConversionResults()); break; } @@ -355,7 +369,7 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD codeHash = mvkHash(pMSLCode, mslCodeLen, codeHash); _device->addActivityPerformance(_device->_performanceStatistics.shaderCompilation.hashShaderCode, startTime); - _defaultLibrary = new MVKShaderLibrary(this, (void*)(pMSLCode), mslCodeLen); + _directMSLLibrary = new MVKShaderLibrary(this, (void*)(pMSLCode), mslCodeLen); break; } @@ -379,12 +393,7 @@ static uint32_t getWorkgroupDimensionSize(const SPIRVWorkgroupSizeDimension& wgD } MVKShaderModule::~MVKShaderModule() { - if (_defaultLibrary) { _defaultLibrary->destroy(); } -} - -void MVKShaderModule::setWorkgroupSize(uint32_t x, uint32_t y, uint32_t z) { - _spvConverter.setWorkgroupSize(x, y, z); - if(_defaultLibrary) { _defaultLibrary->setWorkgroupSize(x, y, z); } + if (_directMSLLibrary) { _directMSLLibrary->destroy(); } } diff --git a/MoltenVK/MoltenVK/Utility/MVKFoundation.h b/MoltenVK/MoltenVK/Utility/MVKFoundation.h index 00b0b853d..f6c123b31 100644 --- a/MoltenVK/MoltenVK/Utility/MVKFoundation.h +++ b/MoltenVK/MoltenVK/Utility/MVKFoundation.h @@ -391,6 +391,13 @@ void mvkReleaseContainerContents(C& container) { } #endif +/** Returns whether the container contains an item equal to the value. */ +template +bool contains(const C& container, const T& val) { + for (const T& cVal : container) { if (cVal == val) { return true; } } + return false; +} + /** Removes the first occurance of the specified value from the specified container. */ template void mvkRemoveFirstOccurance(C& container, T val) { diff --git a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVConversion.mm b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVConversion.mm index 1b8af9b43..e295c4a8d 100644 --- a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVConversion.mm +++ b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVConversion.mm @@ -31,7 +31,7 @@ MVK_PUBLIC_SYMBOL bool mvkConvertSPIRVToMSL(uint32_t* spvCode, char** pResultLog, bool shouldLogSPIRV, bool shouldLogMSL) { - SPIRVToMSLConverterContext spvCtx; + SPIRVToMSLConversionConfiguration spvCtx; SPIRVToMSLConverter spvConverter; spvConverter.setSPIRV(spvCode, spvLength); bool wasConverted = spvConverter.convert(spvCtx, shouldLogSPIRV, shouldLogMSL); diff --git a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.cpp b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.cpp index b8409c0d2..885bb7945 100644 --- a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.cpp +++ b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.cpp @@ -29,28 +29,23 @@ using namespace SPIRV_CROSS_NAMESPACE; #pragma mark - -#pragma mark SPIRVToMSLConverterContext +#pragma mark SPIRVToMSLConversionConfiguration // Returns whether the vector contains the value (using a matches(T&) comparison member function). */ template -bool contains(const vector& vec, const T& val) { +bool containsMatching(const vector& vec, const T& val) { for (const T& vecVal : vec) { if (vecVal.matches(val)) { return true; } } return false; } -MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverterOptions::matches(const SPIRVToMSLConverterOptions& other) const { +MVK_PUBLIC_SYMBOL bool SPIRVToMSLConversionOptions::matches(const SPIRVToMSLConversionOptions& other) const { if (entryPointStage != other.entryPointStage) { return false; } if (entryPointName != other.entryPointName) { return false; } if (tessPatchKind != other.tessPatchKind) { return false; } if (numTessControlPoints != other.numTessControlPoints) { return false; } if (!!shouldFlipVertexY != !!other.shouldFlipVertexY) { return false; } -// if (!!needsSwizzleBuffer != !!other.needsSwizzleBuffer) { return false; } -// if (!!needsOutputBuffer != !!other.needsOutputBuffer) { return false; } -// if (!!needsPatchOutputBuffer != !!other.needsPatchOutputBuffer) { return false; } -// if (!!needsBufferSizeBuffer != !!other.needsBufferSizeBuffer) { return false; } -// if (!!needsInputThreadgroupMem != !!other.needsInputThreadgroupMem) { return false; } -// if (mslOptions.platform != other.mslOptions.platform) { return false; } + if (mslOptions.platform != other.mslOptions.platform) { return false; } if (mslOptions.msl_version != other.mslOptions.msl_version) { return false; } if (mslOptions.texel_buffer_texture_width != other.mslOptions.texel_buffer_texture_width) { return false; } if (mslOptions.swizzle_buffer_index != other.mslOptions.swizzle_buffer_index) { return false; } @@ -61,18 +56,18 @@ MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverterOptions::matches(const SPIRVToMSLConve if (mslOptions.buffer_size_buffer_index != other.mslOptions.buffer_size_buffer_index) { return false; } if (mslOptions.shader_input_wg_index != other.mslOptions.shader_input_wg_index) { return false; } if (!!mslOptions.enable_point_size_builtin != !!other.mslOptions.enable_point_size_builtin) { return false; } -// if (!!mslOptions.disable_rasterization != !!other.mslOptions.disable_rasterization) { return false; } + if (!!mslOptions.disable_rasterization != !!other.mslOptions.disable_rasterization) { return false; } if (!!mslOptions.capture_output_to_buffer != !!other.mslOptions.capture_output_to_buffer) { return false; } if (!!mslOptions.swizzle_texture_samples != !!other.mslOptions.swizzle_texture_samples) { return false; } if (!!mslOptions.tess_domain_origin_lower_left != !!other.mslOptions.tess_domain_origin_lower_left) { return false; } -// if (mslOptions.argument_buffers != other.mslOptions.argument_buffers) { return false; } -// if (mslOptions.pad_fragment_output_components != other.mslOptions.pad_fragment_output_components) { return false; } -// if (mslOptions.texture_buffer_native != other.mslOptions.texture_buffer_native) { return false; } + if (mslOptions.argument_buffers != other.mslOptions.argument_buffers) { return false; } + if (mslOptions.pad_fragment_output_components != other.mslOptions.pad_fragment_output_components) { return false; } + if (mslOptions.texture_buffer_native != other.mslOptions.texture_buffer_native) { return false; } return true; } -MVK_PUBLIC_SYMBOL std::string SPIRVToMSLConverterOptions::printMSLVersion(uint32_t mslVersion, bool includePatch) { +MVK_PUBLIC_SYMBOL std::string SPIRVToMSLConversionOptions::printMSLVersion(uint32_t mslVersion, bool includePatch) { string verStr; uint32_t major = mslVersion / 10000; @@ -91,7 +86,9 @@ MVK_PUBLIC_SYMBOL std::string SPIRVToMSLConverterOptions::printMSLVersion(uint32 return verStr; } -MVK_PUBLIC_SYMBOL SPIRVToMSLConverterOptions::SPIRVToMSLConverterOptions() { +MVK_PUBLIC_SYMBOL SPIRVToMSLConversionOptions::SPIRVToMSLConversionOptions() { + mslOptions.pad_fragment_output_components = true; + #if MVK_MACOS mslOptions.platform = CompilerMSL::Options::macOS; #endif @@ -143,14 +140,14 @@ MVK_PUBLIC_SYMBOL bool mvk::MSLResourceBinding::matches(const MSLResourceBinding return true; } -MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverterContext::stageSupportsVertexAttributes() const { +MVK_PUBLIC_SYMBOL bool SPIRVToMSLConversionConfiguration::stageSupportsVertexAttributes() const { return (options.entryPointStage == spv::ExecutionModelVertex || options.entryPointStage == spv::ExecutionModelTessellationControl || options.entryPointStage == spv::ExecutionModelTessellationEvaluation); } // Check them all in case inactive VA's duplicate locations used by active VA's. -MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverterContext::isVertexAttributeLocationUsed(uint32_t location) const { +MVK_PUBLIC_SYMBOL bool SPIRVToMSLConversionConfiguration::isVertexAttributeLocationUsed(uint32_t location) const { for (auto& va : vertexAttributes) { if ((va.vertexAttribute.location == location) && va.isUsedByShader) { return true; } } @@ -158,15 +155,14 @@ MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverterContext::isVertexAttributeLocationUsed } // Check them all in case inactive VA's duplicate buffers used by active VA's. -MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverterContext::isVertexBufferUsed(uint32_t mslBuffer) const { +MVK_PUBLIC_SYMBOL bool SPIRVToMSLConversionConfiguration::isVertexBufferUsed(uint32_t mslBuffer) const { for (auto& va : vertexAttributes) { if ((va.vertexAttribute.msl_buffer == mslBuffer) && va.isUsedByShader) { return true; } } return false; } -MVK_PUBLIC_SYMBOL void SPIRVToMSLConverterContext::markAllAttributesAndResourcesUsed() { - +MVK_PUBLIC_SYMBOL void SPIRVToMSLConversionConfiguration::markAllAttributesAndResourcesUsed() { if (stageSupportsVertexAttributes()) { for (auto& va : vertexAttributes) { va.isUsedByShader = true; } } @@ -174,31 +170,25 @@ MVK_PUBLIC_SYMBOL void SPIRVToMSLConverterContext::markAllAttributesAndResources for (auto& rb : resourceBindings) { rb.isUsedByShader = true; } } -MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverterContext::matches(const SPIRVToMSLConverterContext& other) const { +MVK_PUBLIC_SYMBOL bool SPIRVToMSLConversionConfiguration::matches(const SPIRVToMSLConversionConfiguration& other) const { if ( !options.matches(other.options) ) { return false; } if (stageSupportsVertexAttributes()) { for (const auto& va : vertexAttributes) { - if (va.isUsedByShader && !contains(other.vertexAttributes, va)) { return false; } + if (va.isUsedByShader && !containsMatching(other.vertexAttributes, va)) { return false; } } } for (const auto& rb : resourceBindings) { - if (rb.isUsedByShader && !contains(other.resourceBindings, rb)) { return false; } + if (rb.isUsedByShader && !containsMatching(other.resourceBindings, rb)) { return false; } } return true; } -MVK_PUBLIC_SYMBOL void SPIRVToMSLConverterContext::alignWith(const SPIRVToMSLConverterContext& srcContext) { - options.mslOptions.disable_rasterization = srcContext.options.mslOptions.disable_rasterization; - options.needsSwizzleBuffer = srcContext.options.needsSwizzleBuffer; - options.needsOutputBuffer = srcContext.options.needsOutputBuffer; - options.needsPatchOutputBuffer = srcContext.options.needsPatchOutputBuffer; - options.needsBufferSizeBuffer = srcContext.options.needsBufferSizeBuffer; - options.needsInputThreadgroupMem = srcContext.options.needsInputThreadgroupMem; +MVK_PUBLIC_SYMBOL void SPIRVToMSLConversionConfiguration::alignWith(const SPIRVToMSLConversionConfiguration& srcContext) { if (stageSupportsVertexAttributes()) { for (auto& va : vertexAttributes) { @@ -221,9 +211,6 @@ MVK_PUBLIC_SYMBOL void SPIRVToMSLConverterContext::alignWith(const SPIRVToMSLCon #pragma mark - #pragma mark SPIRVToMSLConverter -// Populates the entry point with info extracted from the SPRI-V compiler. -void populateEntryPoint(SPIRVEntryPoint& entryPoint, SPIRV_CROSS_NAMESPACE::Compiler* pCompiler, SPIRVToMSLConverterOptions& options); - MVK_PUBLIC_SYMBOL void SPIRVToMSLConverter::setSPIRV(const uint32_t* spirvCode, size_t length) { _spirv.clear(); // Clear for reuse _spirv.reserve(length); @@ -232,7 +219,7 @@ MVK_PUBLIC_SYMBOL void SPIRVToMSLConverter::setSPIRV(const uint32_t* spirvCode, } } -MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverter::convert(SPIRVToMSLConverterContext& context, +MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverter::convert(SPIRVToMSLConversionConfiguration& context, bool shouldLogSPIRV, bool shouldLogMSL, bool shouldLogGLSL) { @@ -245,6 +232,7 @@ MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverter::convert(SPIRVToMSLConverterContext& _wasConverted = true; _resultLog.clear(); _msl.clear(); + _shaderConversionResults.reset(); if (shouldLogSPIRV) { logSPIRV("Converting"); } @@ -272,7 +260,6 @@ MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverter::convert(SPIRVToMSLConverterContext& // Establish the MSL options for the compiler // This needs to be done in two steps...for CompilerMSL and its superclass. - context.options.mslOptions.pad_fragment_output_components = true; pMSLCompiler->set_msl_options(context.options.mslOptions); auto scOpts = pMSLCompiler->get_common_options(); @@ -312,15 +299,15 @@ MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverter::convert(SPIRVToMSLConverterContext& } #endif - // Populate the shader context with info from the compilation run, including - // which vertex attributes and resource bindings are used by the shader - populateEntryPoint(_entryPoint, pMSLCompiler, context.options); - context.options.mslOptions.disable_rasterization = pMSLCompiler && pMSLCompiler->get_is_rasterization_disabled(); - context.options.needsSwizzleBuffer = pMSLCompiler && pMSLCompiler->needs_swizzle_buffer(); - context.options.needsOutputBuffer = pMSLCompiler && pMSLCompiler->needs_output_buffer(); - context.options.needsPatchOutputBuffer = pMSLCompiler && pMSLCompiler->needs_patch_output_buffer(); - context.options.needsBufferSizeBuffer = pMSLCompiler && pMSLCompiler->needs_buffer_size_buffer(); - context.options.needsInputThreadgroupMem = pMSLCompiler && pMSLCompiler->needs_input_threadgroup_mem(); + // Populate the shader conversion results with info from the compilation run, + // and mark which vertex attributes and resource bindings are used by the shader + populateEntryPoint(pMSLCompiler, context.options); + _shaderConversionResults.isRasterizationDisabled = pMSLCompiler && pMSLCompiler->get_is_rasterization_disabled(); + _shaderConversionResults.needsSwizzleBuffer = pMSLCompiler && pMSLCompiler->needs_swizzle_buffer(); + _shaderConversionResults.needsOutputBuffer = pMSLCompiler && pMSLCompiler->needs_output_buffer(); + _shaderConversionResults.needsPatchOutputBuffer = pMSLCompiler && pMSLCompiler->needs_patch_output_buffer(); + _shaderConversionResults.needsBufferSizeBuffer = pMSLCompiler && pMSLCompiler->needs_buffer_size_buffer(); + _shaderConversionResults.needsInputThreadgroupMem = pMSLCompiler && pMSLCompiler->needs_input_threadgroup_mem(); if (context.stageSupportsVertexAttributes()) { for (auto& ctxVA : context.vertexAttributes) { @@ -433,17 +420,17 @@ void SPIRVToMSLConverter::logSource(string& src, const char* srcLang, const char _resultLog += "\n\n"; } - -#pragma mark Support functions - -// Populate a workgroup size dimension. -void populateWorkgroupDimension(SPIRVWorkgroupSizeDimension& wgDim, uint32_t size, SPIRV_CROSS_NAMESPACE::SpecializationConstant& spvSpecConst) { +void SPIRVToMSLConverter::populateWorkgroupDimension(SPIRVWorkgroupSizeDimension& wgDim, + uint32_t size, + SPIRV_CROSS_NAMESPACE::SpecializationConstant& spvSpecConst) { wgDim.size = max(size, 1u); wgDim.isSpecialized = (spvSpecConst.id != 0); wgDim.specializationID = spvSpecConst.constant_id; } -void populateEntryPoint(SPIRVEntryPoint& entryPoint, SPIRV_CROSS_NAMESPACE::Compiler* pCompiler, SPIRVToMSLConverterOptions& options) { +// Populates the entry point with info extracted from the SPRI-V compiler. +void SPIRVToMSLConverter::populateEntryPoint(SPIRV_CROSS_NAMESPACE::Compiler* pCompiler, + SPIRVToMSLConversionOptions& options) { if ( !pCompiler ) { return; } @@ -458,11 +445,14 @@ void populateEntryPoint(SPIRVEntryPoint& entryPoint, SPIRV_CROSS_NAMESPACE::Comp } } + auto& ep = _shaderConversionResults.entryPoint; + ep.mtlFunctionName = spvEP.name; + SPIRV_CROSS_NAMESPACE::SpecializationConstant widthSC, heightSC, depthSC; pCompiler->get_work_group_size_specialization_constants(widthSC, heightSC, depthSC); - entryPoint.mtlFunctionName = spvEP.name; - populateWorkgroupDimension(entryPoint.workgroupSize.width, spvEP.workgroup_size.x, widthSC); - populateWorkgroupDimension(entryPoint.workgroupSize.height, spvEP.workgroup_size.y, heightSC); - populateWorkgroupDimension(entryPoint.workgroupSize.depth, spvEP.workgroup_size.z, depthSC); + auto& wgSize = ep.workgroupSize; + populateWorkgroupDimension(wgSize.width, spvEP.workgroup_size.x, widthSC); + populateWorkgroupDimension(wgSize.height, spvEP.workgroup_size.y, heightSC); + populateWorkgroupDimension(wgSize.depth, spvEP.workgroup_size.z, depthSC); } diff --git a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.h b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.h index ecaa595e3..82024cd0e 100644 --- a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.h +++ b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.h @@ -27,9 +27,8 @@ namespace mvk { - #pragma mark - -#pragma mark SPIRVToMSLConverterContext +#pragma mark SPIRVToMSLConversionConfiguration /** * Options for converting SPIR-V to Metal Shading Language @@ -37,24 +36,19 @@ namespace mvk { * THIS STRUCT IS STREAMED OUT AS PART OF THE PIEPLINE CACHE. * CHANGES TO THIS STRUCT SHOULD BE CAPTURED IN THE STREAMING LOGIC OF THE PIPELINE CACHE. */ - typedef struct SPIRVToMSLConverterOptions { + typedef struct SPIRVToMSLConversionOptions { SPIRV_CROSS_NAMESPACE::CompilerMSL::Options mslOptions; std::string entryPointName; spv::ExecutionModel entryPointStage = spv::ExecutionModelMax; spv::ExecutionMode tessPatchKind = spv::ExecutionModeMax; uint32_t numTessControlPoints = 0; bool shouldFlipVertexY = true; - bool needsSwizzleBuffer = false; - bool needsOutputBuffer = false; - bool needsPatchOutputBuffer = false; - bool needsBufferSizeBuffer = false; - bool needsInputThreadgroupMem = false; /** * Returns whether the specified options match this one. * It does if all corresponding elements are equal. */ - bool matches(const SPIRVToMSLConverterOptions& other) const; + bool matches(const SPIRVToMSLConversionOptions& other) const; bool hasEntryPoint() const { return !entryPointName.empty() && entryPointStage != spv::ExecutionModelMax; @@ -62,14 +56,17 @@ namespace mvk { static std::string printMSLVersion(uint32_t mslVersion, bool includePatch = false); - SPIRVToMSLConverterOptions(); + SPIRVToMSLConversionOptions(); - } SPIRVToMSLConverterOptions; + } SPIRVToMSLConversionOptions; /** * Defines MSL characteristics of a vertex attribute at a particular location. - * The isUsedByShader flag is set to true during conversion of SPIR-V to MSL - * if the shader makes use of this vertex attribute. + * + * The isUsedByShader flag is set to true during conversion of SPIR-V to MSL if the shader + * makes use of this vertex attribute. This allows a pipeline to be optimized, and for two + * shader conversion configurations to be compared only against the attributes that are + * actually used by the shader. * * THIS STRUCT IS STREAMED OUT AS PART OF THE PIEPLINE CACHE. * CHANGES TO THIS STRUCT SHOULD BE CAPTURED IN THE STREAMING LOGIC OF THE PIPELINE CACHE. @@ -98,6 +95,11 @@ namespace mvk { * hardcoded into the MSL as a constexpr type, instead of passed in as a runtime-bound variable. * The content of that constexpr sampler is defined in the constExprSampler parameter. * + * The isUsedByShader flag is set to true during conversion of SPIR-V to MSL if the shader + * makes use of this resource binding. This allows a pipeline to be optimized, and for two + * shader conversion configurations to be compared only against the resource bindings that + * are actually used by the shader. + * * THIS STRUCT IS STREAMED OUT AS PART OF THE PIEPLINE CACHE. * CHANGES TO THIS STRUCT SHOULD BE CAPTURED IN THE STREAMING LOGIC OF THE PIPELINE CACHE. */ @@ -117,13 +119,13 @@ namespace mvk { } MSLResourceBinding; /** - * Context passed to the SPIRVToMSLConverter to map SPIR-V descriptors to Metal resource indices. + * Configuration passed to the SPIRVToMSLConverter. * * THIS STRUCT IS STREAMED OUT AS PART OF THE PIEPLINE CACHE. * CHANGES TO THIS STRUCT SHOULD BE CAPTURED IN THE STREAMING LOGIC OF THE PIPELINE CACHE. */ - typedef struct SPIRVToMSLConverterContext { - SPIRVToMSLConverterOptions options; + typedef struct SPIRVToMSLConversionConfiguration { + SPIRVToMSLConversionOptions options; std::vector vertexAttributes; std::vector resourceBindings; @@ -140,17 +142,22 @@ namespace mvk { void markAllAttributesAndResourcesUsed(); /** - * Returns whether this context matches the other context. It does if the respective - * options match and any vertex attributes and resource bindings used by this context - * can be found in the other context. Vertex attributes and resource bindings that are - * in the other context but are not used by the shader that created this context, are ignored. + * Returns whether this configuration matches the other context. It does if the + * respective options match and any vertex attributes and resource bindings used + * by this configuration can be found in the other configuration. Vertex attributes + * and resource bindings that are in the other configuration but are not used by + * the shader that created this configuration, are ignored. */ - bool matches(const SPIRVToMSLConverterContext& other) const; + bool matches(const SPIRVToMSLConversionConfiguration& other) const; + + /** Aligns certain aspects of this configuration with the source context. */ + void alignWith(const SPIRVToMSLConversionConfiguration& srcContext); + + } SPIRVToMSLConversionConfiguration; - /** Aligns certain aspects of this context with the source context. */ - void alignWith(const SPIRVToMSLConverterContext& srcContext); - } SPIRVToMSLConverterContext; +#pragma mark - +#pragma mark SPIRVToMSLConversionResults /** * Describes one dimension of the workgroup size of a SPIR-V entry point, including whether @@ -183,6 +190,25 @@ namespace mvk { } workgroupSize; } SPIRVEntryPoint; + /** + * Contains the results of the shader conversion that can be used to populate a pipeline. + * + * THIS STRUCT IS STREAMED OUT AS PART OF THE PIEPLINE CACHE. + * CHANGES TO THIS STRUCT SHOULD BE CAPTURED IN THE STREAMING LOGIC OF THE PIPELINE CACHE. + */ + typedef struct SPIRVToMSLConversionResults { + SPIRVEntryPoint entryPoint; + bool isRasterizationDisabled = false; + bool needsSwizzleBuffer = false; + bool needsOutputBuffer = false; + bool needsPatchOutputBuffer = false; + bool needsBufferSizeBuffer = false; + bool needsInputThreadgroupMem = false; + + void reset() { *this = SPIRVToMSLConversionResults(); } + + } SPIRVToMSLConversionResults; + #pragma mark - #pragma mark SPIRVToMSLConverter @@ -214,7 +240,7 @@ namespace mvk { * and optionally, the original GLSL (as converted from the SPIR_V), should be logged * to the result log of this converter. This can be useful during shader debugging. */ - bool convert(SPIRVToMSLConverterContext& context, + bool convert(SPIRVToMSLConversionConfiguration& context, bool shouldLogSPIRV = false, bool shouldLogMSL = false, bool shouldLogGLSL = false); @@ -232,14 +258,15 @@ namespace mvk { */ const std::string& getMSL() { return _msl; } - /** Returns information about the shader entry point. */ - const SPIRVEntryPoint& getEntryPoint() { return _entryPoint; } + /** Returns information about the shader conversion. */ + const SPIRVToMSLConversionResults& getConversionResults() { return _shaderConversionResults; } /** Sets the number of threads in a single compute kernel workgroup, per dimension. */ void setWorkgroupSize(uint32_t x, uint32_t y, uint32_t z) { - _entryPoint.workgroupSize.width.size = x; - _entryPoint.workgroupSize.height.size = y; - _entryPoint.workgroupSize.depth.size = z; + auto& wgSize = _shaderConversionResults.entryPoint.workgroupSize; + wgSize.width.size = x; + wgSize.height.size = y; + wgSize.depth.size = z; } /** @@ -249,10 +276,10 @@ namespace mvk { const std::string& getResultLog() { return _resultLog; } /** Sets MSL source code. This can be used when MSL is supplied directly. */ - void setMSL(const std::string& msl, const SPIRVEntryPoint* pEntryPoint) { - _msl = msl; - if (pEntryPoint) { _entryPoint = *pEntryPoint; } - } + void setMSL(const std::string& msl, const SPIRVToMSLConversionResults* pShaderConversionResults) { + _msl = msl; + if (pShaderConversionResults) { _shaderConversionResults = *pShaderConversionResults; } + } protected: void logMsg(const char* logMsg); @@ -261,11 +288,13 @@ namespace mvk { bool validateSPIRV(); void writeSPIRVToFile(std::string spvFilepath); void logSource(std::string& src, const char* srcLang, const char* opDesc); + void populateWorkgroupDimension(SPIRVWorkgroupSizeDimension& wgDim, uint32_t size, SPIRV_CROSS_NAMESPACE::SpecializationConstant& spvSpecConst); + void populateEntryPoint(SPIRV_CROSS_NAMESPACE::Compiler* pCompiler, SPIRVToMSLConversionOptions& options); std::vector _spirv; std::string _msl; std::string _resultLog; - SPIRVEntryPoint _entryPoint; + SPIRVToMSLConversionResults _shaderConversionResults; bool _wasConverted = false; }; diff --git a/MoltenVKShaderConverter/MoltenVKShaderConverterTool/MoltenVKShaderConverterTool.cpp b/MoltenVKShaderConverter/MoltenVKShaderConverterTool/MoltenVKShaderConverterTool.cpp index b95ab6bcb..b7c55522b 100644 --- a/MoltenVKShaderConverter/MoltenVKShaderConverterTool/MoltenVKShaderConverterTool.cpp +++ b/MoltenVKShaderConverter/MoltenVKShaderConverterTool/MoltenVKShaderConverterTool.cpp @@ -204,7 +204,7 @@ bool MoltenVKShaderConverterTool::convertSPIRV(const vector& spv, if ( !_shouldWriteMSL ) { return true; } // Derive the context under which conversion will occur - SPIRVToMSLConverterContext mslContext; + SPIRVToMSLConversionConfiguration mslContext; mslContext.options.mslOptions.platform = _mslPlatform; mslContext.options.mslOptions.set_msl_version(_mslVersionMajor, _mslVersionMinor, _mslVersionPatch); mslContext.options.shouldFlipVertexY = _shouldFlipVertexY; @@ -387,7 +387,7 @@ MoltenVKShaderConverterTool::MoltenVKShaderConverterTool(int argc, const char* a _mslVersionMajor = 2; _mslVersionMinor = 1; _mslVersionPatch = 0; - _mslPlatform = SPIRVToMSLConverterOptions().mslOptions.platform; + _mslPlatform = SPIRVToMSLConversionOptions().mslOptions.platform; _isActive = parseArgs(argc, argv); if ( !_isActive ) { showUsage(); }