From f857011ec16fc6cf8d7f88f8c2c7526cc0fd150a Mon Sep 17 00:00:00 2001 From: Bill Hollings Date: Fri, 4 May 2018 12:11:19 -0400 Subject: [PATCH] Add features to support Vulkan CTS. Dynamically create frag shaders for clearning attachments and images. Dynamically create frag shaders for blitting scaled images. MVKGraphicsPipeline don't create MTLRenderPipelineState if vertex function conversion fails. MVKComputePipeline don't create MTLComputePipelineState if compute function conversion fails. Handle SPIRV-Cross errors thrown during SPIR-V parsing in compiler construction. Set undefined property limits to large, but not max, values to avoid casting issues in app. Mark multiDrawIndirect features as available. Update to latest SPIRV-Cross version. Update to MoltenVK version 1.0.5. --- ExternalRevisions/README.md | 9 +- ExternalRevisions/SPIRV-Cross_repo_revision | 2 +- MoltenVK/MoltenVK.xcodeproj/project.pbxproj | 12 ++ MoltenVK/MoltenVK/API/mvk_datatypes.h | 11 +- MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h | 2 +- MoltenVK/MoltenVK/Commands/MVKCmdDraw.h | 4 +- MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm | 16 +- ...KCommandPipelineStateFactoryShaderSource.h | 141 +--------------- .../Commands/MVKCommandResourceFactory.h | 13 +- .../Commands/MVKCommandResourceFactory.mm | 152 +++++++++++++++--- MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm | 46 +++--- MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm | 26 +-- MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm | 2 +- MoltenVK/MoltenVK/GPUObjects/MVKSwapchain.mm | 4 +- MoltenVK/MoltenVK/GPUObjects/MVKSync.mm | 2 +- MoltenVK/MoltenVK/Utility/MVKFoundation.h | 13 +- MoltenVK/MoltenVK/Utility/NSString+MoltenVK.h | 40 +++++ .../MoltenVK/Utility/NSString+MoltenVK.mm | 37 +++++ MoltenVK/MoltenVK/Vulkan/mvk_datatypes.mm | 87 +++++----- MoltenVK/MoltenVK/Vulkan/vulkan.mm | 8 +- .../SPIRVToMSLConverter.cpp | 88 +++++----- .../SPIRVToMSLConverter.h | 3 - .../MoltenVKShaderConverter.xcscheme | 10 +- 23 files changed, 412 insertions(+), 316 deletions(-) create mode 100644 MoltenVK/MoltenVK/Utility/NSString+MoltenVK.h create mode 100644 MoltenVK/MoltenVK/Utility/NSString+MoltenVK.mm diff --git a/ExternalRevisions/README.md b/ExternalRevisions/README.md index 90c20624..f91a43af 100644 --- a/ExternalRevisions/README.md +++ b/ExternalRevisions/README.md @@ -152,12 +152,17 @@ If you make changes to the `SPIRV-Cross` repository, you can regression test you cd External/SPIRV-Cross ./checkout_glslang_spirv_tools.sh + ./build_glslang_spirv_tools.sh -2. Run the regression tests: +2. Build `SPIRV-Cross`: + + make + +3. Run the regression tests: ./test_shaders.sh -3. If your changes result in different expected output for a reference shader, and the new results +4. If your changes result in different expected output for a reference shader, and the new results are correct, you can update the reference shader for a particular regression test by deleting that reference shader, in either `External/SPIRV-Cross/reference/shaders-msl` or `External/SPIRV-Cross/reference/opt/shaders-msl`, and running the test again. The test will diff --git a/ExternalRevisions/SPIRV-Cross_repo_revision b/ExternalRevisions/SPIRV-Cross_repo_revision index 240a9a04..b0d24da2 100644 --- a/ExternalRevisions/SPIRV-Cross_repo_revision +++ b/ExternalRevisions/SPIRV-Cross_repo_revision @@ -1 +1 @@ -7796a9f3ec94733830ad8c648157a3b1c5693e34 +2792f8f3f2ce5cf970ef1cb7ab144445a4f1f6f8 diff --git a/MoltenVK/MoltenVK.xcodeproj/project.pbxproj b/MoltenVK/MoltenVK.xcodeproj/project.pbxproj index 120f26b4..3efeee74 100644 --- a/MoltenVK/MoltenVK.xcodeproj/project.pbxproj +++ b/MoltenVK/MoltenVK.xcodeproj/project.pbxproj @@ -13,6 +13,10 @@ A90C8DEB1F45354D009CB32C /* MVKCommandEncodingPool.h in Headers */ = {isa = PBXBuildFile; fileRef = A90C8DE81F45354D009CB32C /* MVKCommandEncodingPool.h */; }; A90C8DEC1F45354D009CB32C /* MVKCommandEncodingPool.mm in Sources */ = {isa = PBXBuildFile; fileRef = A90C8DE91F45354D009CB32C /* MVKCommandEncodingPool.mm */; }; A90C8DED1F45354D009CB32C /* MVKCommandEncodingPool.mm in Sources */ = {isa = PBXBuildFile; fileRef = A90C8DE91F45354D009CB32C /* MVKCommandEncodingPool.mm */; }; + A92CAF392098166B009DA08E /* NSString+MoltenVK.h in Headers */ = {isa = PBXBuildFile; fileRef = A92CAF342098166A009DA08E /* NSString+MoltenVK.h */; }; + A92CAF3A2098166B009DA08E /* NSString+MoltenVK.h in Headers */ = {isa = PBXBuildFile; fileRef = A92CAF342098166A009DA08E /* NSString+MoltenVK.h */; }; + A92CAF3B2098166B009DA08E /* NSString+MoltenVK.mm in Sources */ = {isa = PBXBuildFile; fileRef = A92CAF382098166B009DA08E /* NSString+MoltenVK.mm */; }; + A92CAF3C2098166B009DA08E /* NSString+MoltenVK.mm in Sources */ = {isa = PBXBuildFile; fileRef = A92CAF382098166B009DA08E /* NSString+MoltenVK.mm */; }; A948BB7F1E51642700DE59F2 /* mvk_vulkan.h in Headers */ = {isa = PBXBuildFile; fileRef = A948BB7E1E51642700DE59F2 /* mvk_vulkan.h */; settings = {ATTRIBUTES = (Public, ); }; }; A948BB801E51642700DE59F2 /* mvk_vulkan.h in Headers */ = {isa = PBXBuildFile; fileRef = A948BB7E1E51642700DE59F2 /* mvk_vulkan.h */; settings = {ATTRIBUTES = (Public, ); }; }; A94FB7B01C7DFB4800632CA3 /* mvk_datatypes.h in Headers */ = {isa = PBXBuildFile; fileRef = A94FB7671C7DFB4800632CA3 /* mvk_datatypes.h */; settings = {ATTRIBUTES = (Public, ); }; }; @@ -232,6 +236,8 @@ A9096E5D1F81E16300DFBEA6 /* MVKCmdDispatch.mm */ = {isa = PBXFileReference; lastKnownFileType = sourcecode.cpp.objcpp; path = MVKCmdDispatch.mm; sourceTree = ""; }; A90C8DE81F45354D009CB32C /* MVKCommandEncodingPool.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = MVKCommandEncodingPool.h; sourceTree = ""; }; A90C8DE91F45354D009CB32C /* MVKCommandEncodingPool.mm */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.cpp.objcpp; path = MVKCommandEncodingPool.mm; sourceTree = ""; }; + A92CAF342098166A009DA08E /* NSString+MoltenVK.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = "NSString+MoltenVK.h"; sourceTree = ""; }; + A92CAF382098166B009DA08E /* NSString+MoltenVK.mm */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.cpp.objcpp; path = "NSString+MoltenVK.mm"; sourceTree = ""; }; A948BB7E1E51642700DE59F2 /* mvk_vulkan.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = mvk_vulkan.h; sourceTree = ""; }; A94FB7671C7DFB4800632CA3 /* mvk_datatypes.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = mvk_datatypes.h; sourceTree = ""; }; A94FB7691C7DFB4800632CA3 /* vk_mvk_moltenvk.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = vk_mvk_moltenvk.h; sourceTree = ""; }; @@ -469,6 +475,8 @@ A981494A1FB6A3F7005F00B4 /* MVKWatermark.mm */, A981494B1FB6A3F7005F00B4 /* MVKWatermarkShaderSource.h */, A981494C1FB6A3F7005F00B4 /* MVKWatermarkTextureContent.h */, + A92CAF342098166A009DA08E /* NSString+MoltenVK.h */, + A92CAF382098166B009DA08E /* NSString+MoltenVK.mm */, ); path = Utility; sourceTree = ""; @@ -564,6 +572,7 @@ A94FB7C41C7DFB4800632CA3 /* MVKCmdRenderPass.h in Headers */, A94FB7BC1C7DFB4800632CA3 /* MVKCmdPipeline.h in Headers */, A94FB7F81C7DFB4800632CA3 /* MVKPipeline.h in Headers */, + A92CAF392098166B009DA08E /* NSString+MoltenVK.h in Headers */, A94FB7F01C7DFB4800632CA3 /* MVKImage.h in Headers */, A94FB7B81C7DFB4800632CA3 /* MVKCmdTransfer.h in Headers */, A94FB7C81C7DFB4800632CA3 /* MVKCmdDraw.h in Headers */, @@ -616,6 +625,7 @@ A94FB7C51C7DFB4800632CA3 /* MVKCmdRenderPass.h in Headers */, A94FB7BD1C7DFB4800632CA3 /* MVKCmdPipeline.h in Headers */, A94FB7F91C7DFB4800632CA3 /* MVKPipeline.h in Headers */, + A92CAF3A2098166B009DA08E /* NSString+MoltenVK.h in Headers */, A94FB7F11C7DFB4800632CA3 /* MVKImage.h in Headers */, A94FB7B91C7DFB4800632CA3 /* MVKCmdTransfer.h in Headers */, A94FB7C91C7DFB4800632CA3 /* MVKCmdDraw.h in Headers */, @@ -798,6 +808,7 @@ A94FB8321C7DFB4800632CA3 /* vulkan.mm in Sources */, A94FB8121C7DFB4800632CA3 /* MVKSurface.mm in Sources */, A94FB7FE1C7DFB4800632CA3 /* MVKQueryPool.mm in Sources */, + A92CAF3B2098166B009DA08E /* NSString+MoltenVK.mm in Sources */, A94FB7F61C7DFB4800632CA3 /* MVKInstance.mm in Sources */, A94FB7EA1C7DFB4800632CA3 /* MVKDeviceMemory.mm in Sources */, A94FB7F21C7DFB4800632CA3 /* MVKImage.mm in Sources */, @@ -842,6 +853,7 @@ A94FB8331C7DFB4800632CA3 /* vulkan.mm in Sources */, A94FB8131C7DFB4800632CA3 /* MVKSurface.mm in Sources */, A94FB7FF1C7DFB4800632CA3 /* MVKQueryPool.mm in Sources */, + A92CAF3C2098166B009DA08E /* NSString+MoltenVK.mm in Sources */, A94FB7F71C7DFB4800632CA3 /* MVKInstance.mm in Sources */, A94FB7EB1C7DFB4800632CA3 /* MVKDeviceMemory.mm in Sources */, A94FB7F31C7DFB4800632CA3 /* MVKImage.mm in Sources */, diff --git a/MoltenVK/MoltenVK/API/mvk_datatypes.h b/MoltenVK/MoltenVK/API/mvk_datatypes.h index 78f35817..c7b331ae 100644 --- a/MoltenVK/MoltenVK/API/mvk_datatypes.h +++ b/MoltenVK/MoltenVK/API/mvk_datatypes.h @@ -45,9 +45,14 @@ extern "C" { /** Enumerates the data type of a format. */ typedef enum { kMVKFormatNone, /**< Format type is unknown. */ - kMVKFormatColorFloat, /**< A floating point color. */ - kMVKFormatColorInt, /**< A signed integer color. */ - kMVKFormatColorUInt, /**< An unsigned integer color. */ + kMVKFormatColorHalf, /**< A 16-bit floating point color. */ + kMVKFormatColorFloat, /**< A 32-bit floating point color. */ + kMVKFormatColorInt8, /**< A signed 8-bit integer color. */ + kMVKFormatColorUInt8, /**< An unsigned 8-bit integer color. */ + kMVKFormatColorInt16, /**< A signed 16-bit integer color. */ + kMVKFormatColorUInt16, /**< An unsigned 16-bit integer color. */ + kMVKFormatColorInt32, /**< A signed 32-bit integer color. */ + kMVKFormatColorUInt32, /**< An unsigned 32-bit integer color. */ kMVKFormatDepthStencil, /**< A depth and stencil value. */ } MVKFormatType; diff --git a/MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h b/MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h index b15e7c55..a5952162 100644 --- a/MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h +++ b/MoltenVK/MoltenVK/API/vk_mvk_moltenvk.h @@ -48,7 +48,7 @@ extern "C" { */ #define MVK_VERSION_MAJOR 1 #define MVK_VERSION_MINOR 0 -#define MVK_VERSION_PATCH 4 +#define MVK_VERSION_PATCH 5 #define MVK_MAKE_VERSION(major, minor, patch) (((major) * 10000) + ((minor) * 100) + (patch)) #define MVK_VERSION MVK_MAKE_VERSION(MVK_VERSION_MAJOR, MVK_VERSION_MINOR, MVK_VERSION_PATCH) diff --git a/MoltenVK/MoltenVK/Commands/MVKCmdDraw.h b/MoltenVK/MoltenVK/Commands/MVKCmdDraw.h index 2580fdcf..91bd8b6f 100644 --- a/MoltenVK/MoltenVK/Commands/MVKCmdDraw.h +++ b/MoltenVK/MoltenVK/Commands/MVKCmdDraw.h @@ -199,13 +199,13 @@ void mvkCmdDrawIndexed(MVKCommandBuffer* cmdBuff, void mvkCmdDrawIndirect(MVKCommandBuffer* cmdBuff, VkBuffer buffer, VkDeviceSize offset, - uint32_t count, + uint32_t drawCount, uint32_t stride); /** Adds an indirect indexed draw command to the specified command buffer. */ void mvkCmdDrawIndexedIndirect(MVKCommandBuffer* cmdBuff, VkBuffer buffer, VkDeviceSize offset, - uint32_t count, + uint32_t drawCount, uint32_t stride); diff --git a/MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm b/MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm index a4a9bbd8..5b2d6304 100644 --- a/MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm +++ b/MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm @@ -174,13 +174,13 @@ MVKCmdDrawIndexed::MVKCmdDrawIndexed(MVKCommandTypePool* pool void MVKCmdDrawIndirect::setContent(VkBuffer buffer, VkDeviceSize offset, - uint32_t count, + uint32_t drawCount, uint32_t stride) { MVKBuffer* mvkBuffer = (MVKBuffer*)buffer; _mtlIndirectBuffer = mvkBuffer->getMTLBuffer(); _mtlIndirectBufferOffset = mvkBuffer->getMTLBufferOffset() + offset; _mtlIndirectBufferStride = stride; - _drawCount = count; + _drawCount = drawCount; // Validate clearConfigurationResult(); @@ -211,13 +211,13 @@ MVKCmdDrawIndirect::MVKCmdDrawIndirect(MVKCommandTypePool* p void MVKCmdDrawIndexedIndirect::setContent(VkBuffer buffer, VkDeviceSize offset, - uint32_t count, + uint32_t drawCount, uint32_t stride) { MVKBuffer* mvkBuffer = (MVKBuffer*)buffer; _mtlIndirectBuffer = mvkBuffer->getMTLBuffer(); _mtlIndirectBufferOffset = mvkBuffer->getMTLBufferOffset() + offset; _mtlIndirectBufferStride = stride; - _drawCount = count; + _drawCount = drawCount; // Validate clearConfigurationResult(); @@ -294,20 +294,20 @@ void mvkCmdBindIndexBuffer(MVKCommandBuffer* cmdBuff, void mvkCmdDrawIndirect(MVKCommandBuffer* cmdBuff, VkBuffer buffer, VkDeviceSize offset, - uint32_t count, + uint32_t drawCount, uint32_t stride) { MVKCmdDrawIndirect* cmd = cmdBuff->_commandPool->_cmdDrawIndirectPool.acquireObject(); - cmd->setContent(buffer, offset, count, stride); + cmd->setContent(buffer, offset, drawCount, stride); cmdBuff->addCommand(cmd); } void mvkCmdDrawIndexedIndirect(MVKCommandBuffer* cmdBuff, VkBuffer buffer, VkDeviceSize offset, - uint32_t count, + uint32_t drawCount, uint32_t stride) { MVKCmdDrawIndexedIndirect* cmd = cmdBuff->_commandPool->_cmdDrawIndexedIndirectPool.acquireObject(); - cmd->setContent(buffer, offset, count, stride); + cmd->setContent(buffer, offset, drawCount, stride); cmdBuff->addCommand(cmd); } diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandPipelineStateFactoryShaderSource.h b/MoltenVK/MoltenVK/Commands/MVKCommandPipelineStateFactoryShaderSource.h index 84b4e0b2..c7233f3a 100644 --- a/MoltenVK/MoltenVK/Commands/MVKCommandPipelineStateFactoryShaderSource.h +++ b/MoltenVK/MoltenVK/Commands/MVKCommandPipelineStateFactoryShaderSource.h @@ -56,156 +56,29 @@ vertex VaryingsPos vtxCmdBlitImageD(AttributesPosTex attributes [[stage_in]], varyings.gl_Position = float4(attributes.a_position, depth, 1.0); \n\ return varyings; \n\ } \n\ - \n\ -fragment float4 fragCmdBlitImageF(VaryingsPosTex varyings [[stage_in]], \n\ - texture2d texture [[texture(0)]], \n\ - sampler sampler [[ sampler(0) ]]) { \n\ - return texture.sample(sampler, varyings.v_texCoord); \n\ -}; \n\ - \n\ -fragment int4 fragCmdBlitImageI(VaryingsPosTex varyings [[stage_in]], \n\ - texture2d texture [[texture(0)]], \n\ - sampler sampler [[ sampler(0) ]]) { \n\ - return texture.sample(sampler, varyings.v_texCoord); \n\ -}; \n\ - \n\ -fragment uint4 fragCmdBlitImageU(VaryingsPosTex varyings [[stage_in]], \n\ - texture2d texture [[texture(0)]], \n\ - sampler sampler [[ sampler(0) ]]) { \n\ - return texture.sample(sampler, varyings.v_texCoord); \n\ -}; \n\ - \n\ -fragment float4 fragCmdBlitImageDF(VaryingsPosTex varyings [[stage_in]], \n\ - depth2d texture [[texture(0)]], \n\ - sampler sampler [[ sampler(0) ]]) { \n\ - return texture.sample(sampler, varyings.v_texCoord); \n\ -}; \n\ - \n\ -fragment int4 fragCmdBlitImageDI(VaryingsPosTex varyings [[stage_in]], \n\ - depth2d texture [[texture(0)]], \n\ - sampler sampler [[ sampler(0) ]]) { \n\ - return int4(texture.sample(sampler, varyings.v_texCoord)); \n\ -}; \n\ - \n\ -fragment uint4 fragCmdBlitImageDU(VaryingsPosTex varyings [[stage_in]], \n\ - depth2d texture [[texture(0)]], \n\ - sampler sampler [[ sampler(0) ]]) { \n\ - return uint4(texture.sample(sampler, varyings.v_texCoord)); \n\ -}; \n\ \n\ typedef struct { \n\ float4 colors[9]; \n\ } ClearColorsIn; \n\ \n\ -typedef struct { \n\ - float4 color0 [[color(0)]]; \n\ - float4 color1 [[color(1)]]; \n\ - float4 color2 [[color(2)]]; \n\ - float4 color3 [[color(3)]]; \n\ - float4 color4 [[color(4)]]; \n\ - float4 color5 [[color(5)]]; \n\ - float4 color6 [[color(6)]]; \n\ - float4 color7 [[color(7)]]; \n\ -} ClearColorsOutF; \n\ - \n\ -typedef struct { \n\ - int4 color0 [[color(0)]]; \n\ - int4 color1 [[color(1)]]; \n\ - int4 color2 [[color(2)]]; \n\ - int4 color3 [[color(3)]]; \n\ - int4 color4 [[color(4)]]; \n\ - int4 color5 [[color(5)]]; \n\ - int4 color6 [[color(6)]]; \n\ - int4 color7 [[color(7)]]; \n\ -} ClearColorsOutI; \n\ - \n\ -typedef struct { \n\ - uint4 color0 [[color(0)]]; \n\ - uint4 color1 [[color(1)]]; \n\ - uint4 color2 [[color(2)]]; \n\ - uint4 color3 [[color(3)]]; \n\ - uint4 color4 [[color(4)]]; \n\ - uint4 color5 [[color(5)]]; \n\ - uint4 color6 [[color(6)]]; \n\ - uint4 color7 [[color(7)]]; \n\ -} ClearColorsOutU; \n\ - \n\ vertex VaryingsPos vtxCmdClearAttachments(AttributesPos attributes [[stage_in]], \n\ constant ClearColorsIn& ccIn [[buffer(0)]]) { \n\ VaryingsPos varyings; \n\ varyings.gl_Position = float4(attributes.a_position.x, -attributes.a_position.y, ccIn.colors[8].r, 1.0); \n\ return varyings; \n\ } \n\ - \n\ -fragment ClearColorsOutF fragCmdClearAttachmentsF(VaryingsPos varyings [[stage_in]], \n\ - constant ClearColorsIn& ccIn [[buffer(0)]]) { \n\ - ClearColorsOutF ccOut; \n\ - ccOut.color0 = ccIn.colors[0]; \n\ - ccOut.color1 = ccIn.colors[1]; \n\ - ccOut.color2 = ccIn.colors[2]; \n\ - ccOut.color3 = ccIn.colors[3]; \n\ - ccOut.color4 = ccIn.colors[4]; \n\ - ccOut.color5 = ccIn.colors[5]; \n\ - ccOut.color6 = ccIn.colors[6]; \n\ - ccOut.color7 = ccIn.colors[7]; \n\ - return ccOut; \n\ -}; \n\ - \n\ -fragment float4 fragCmdClearAttachments0F(VaryingsPos varyings [[stage_in]], \n\ - constant ClearColorsIn& ccIn [[buffer(0)]]) { \n\ - return ccIn.colors[0]; \n\ -}; \n\ \n\ -fragment ClearColorsOutI fragCmdClearAttachmentsI(VaryingsPos varyings [[stage_in]], \n\ - constant ClearColorsIn& ccIn [[buffer(0)]]) { \n\ - ClearColorsOutI ccOut; \n\ - ccOut.color0 = int4(ccIn.colors[0]); \n\ - ccOut.color1 = int4(ccIn.colors[1]); \n\ - ccOut.color2 = int4(ccIn.colors[2]); \n\ - ccOut.color3 = int4(ccIn.colors[3]); \n\ - ccOut.color4 = int4(ccIn.colors[4]); \n\ - ccOut.color5 = int4(ccIn.colors[5]); \n\ - ccOut.color6 = int4(ccIn.colors[6]); \n\ - ccOut.color7 = int4(ccIn.colors[7]); \n\ - return ccOut; \n\ -}; \n\ - \n\ -fragment int4 fragCmdClearAttachments0I(VaryingsPos varyings [[stage_in]], \n\ - constant ClearColorsIn& ccIn [[buffer(0)]]) { \n\ - return int4(ccIn.colors[0]); \n\ -}; \n\ - \n\ -fragment ClearColorsOutU fragCmdClearAttachmentsU(VaryingsPos varyings [[stage_in]], \n\ - constant ClearColorsIn& ccIn [[buffer(0)]]) { \n\ - ClearColorsOutU ccOut; \n\ - ccOut.color0 = uint4(ccIn.colors[0]); \n\ - ccOut.color1 = uint4(ccIn.colors[1]); \n\ - ccOut.color2 = uint4(ccIn.colors[2]); \n\ - ccOut.color3 = uint4(ccIn.colors[3]); \n\ - ccOut.color4 = uint4(ccIn.colors[4]); \n\ - ccOut.color5 = uint4(ccIn.colors[5]); \n\ - ccOut.color6 = uint4(ccIn.colors[6]); \n\ - ccOut.color7 = uint4(ccIn.colors[7]); \n\ - return ccOut; \n\ -}; \n\ - \n\ -fragment uint4 fragCmdClearAttachments0U(VaryingsPos varyings [[stage_in]], \n\ - constant ClearColorsIn& ccIn [[buffer(0)]]) { \n\ - return uint4(ccIn.colors[0]); \n\ -}; \n\ - \n\ -struct CopyInfo \n\ -{ \n\ - uint32_t SrcOffset; \n\ - uint32_t DstOffset; \n\ - uint32_t CopySize; \n\ -}; \n\ +typedef struct { \n\ + uint32_t srcOffset; \n\ + uint32_t dstOffset; \n\ + uint32_t copySize; \n\ +} CopyInfo; \n\ \n\ kernel void compCopyBufferBytes(device uint8_t* src [[ buffer(0) ]], \n\ device uint8_t* dst [[ buffer(1) ]], \n\ constant CopyInfo& info [[ buffer(2) ]]) { \n\ - for (size_t i = 0; i < info.CopySize; i++) { \n\ - dst[i + info.DstOffset] = src[i + info.SrcOffset]; \n\ + for (size_t i = 0; i < info.copySize; i++) { \n\ + dst[i + info.dstOffset] = src[i + info.srcOffset]; \n\ } \n\ }; \n\ "; diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.h b/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.h index 66163b95..d18bb13f 100644 --- a/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.h +++ b/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.h @@ -50,13 +50,6 @@ typedef struct MVKRPSKeyClearAtt_t { bool isEnabled(uint32_t attIdx) { return mvkIsAnyFlagEnabled(enabledFlags, bitFlag << attIdx); } - bool isEnabledOnly(uint32_t attIdx) { - // Ignore depth stencil bit - uint32_t colorFlags = enabledFlags; - mvkDisableFlag(colorFlags, bitFlag << kMVKAttachmentFormatDepthStencilIndex); - return mvkAreOnlyFlagsEnabled(colorFlags, bitFlag << attIdx); - } - bool operator==(const MVKRPSKeyClearAtt_t& rhs) const { return ((enabledFlags == rhs.enabledFlags) && (mtlSampleCount == rhs.mtlSampleCount) && @@ -288,9 +281,11 @@ public: protected: void initMTLLibrary(); - std::string getFragFunctionSuffix(MTLPixelFormat mtlPixFmt); - std::string getFragFunctionSuffix(MVKRPSKeyClearAtt& attKey); + id getBlitFragFunction(MTLPixelFormat mtlPixFmt); + id getClearFragFunction(MVKRPSKeyClearAtt& attKey); + NSString* getMTLFormatTypeString(MTLPixelFormat mtlPixFmt); id getFunctionNamed(const char* funcName); + id newMTLFunction(NSString* mslSrcCode, NSString* funcName); id newMTLRenderPipelineState(MTLRenderPipelineDescriptor* plDesc); id _mtlLibrary; diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.mm b/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.mm index a09a6ab7..6a9e0097 100644 --- a/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.mm +++ b/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.mm @@ -20,6 +20,7 @@ #include "MVKCommandPipelineStateFactoryShaderSource.h" #include "MVKFoundation.h" #include "mvk_datatypes.h" +#include "NSString+MoltenVK.h" #include "MVKLogging.h" using namespace std; @@ -30,19 +31,17 @@ using namespace std; id MVKCommandResourceFactory::newCmdBlitImageMTLRenderPipelineState(MTLPixelFormat mtlPixFmt) { bool isDepthFormat = mvkMTLPixelFormatIsDepthFormat(mtlPixFmt); - string fragFuncSfx = getFragFunctionSuffix(mtlPixFmt); MTLRenderPipelineDescriptor* plDesc = [[[MTLRenderPipelineDescriptor alloc] init] autorelease]; - plDesc.label = [NSString stringWithFormat: @"CmdBlitImage%s-%lu", fragFuncSfx.data(), (unsigned long)mtlPixFmt]; + plDesc.label = [NSString stringWithFormat: @"CmdBlitImage"]; plDesc.vertexFunction = getFunctionNamed(isDepthFormat ? "vtxCmdBlitImageD" : "vtxCmdBlitImage"); - plDesc.fragmentFunction = getFunctionNamed((string("fragCmdBlitImage") + fragFuncSfx).data()); + plDesc.fragmentFunction = getBlitFragFunction(mtlPixFmt); - if ( isDepthFormat ) { + if (isDepthFormat) { plDesc.depthAttachmentPixelFormat = mtlPixFmt; } else { - MTLRenderPipelineColorAttachmentDescriptor* colorDesc = plDesc.colorAttachments[0]; - colorDesc.pixelFormat = mtlPixFmt; + plDesc.colorAttachments[0].pixelFormat = mtlPixFmt; } MTLVertexDescriptor* vtxDesc = plDesc.vertexDescriptor; @@ -91,12 +90,10 @@ id MVKCommandResourceFactory::newCmdBlitImageMTLSamplerState(MT } id MVKCommandResourceFactory::newCmdClearMTLRenderPipelineState(MVKRPSKeyClearAtt& attKey) { - string fragFuncSfx = getFragFunctionSuffix(attKey); - MTLRenderPipelineDescriptor* plDesc = [[[MTLRenderPipelineDescriptor alloc] init] autorelease]; - plDesc.label = [NSString stringWithFormat: @"CmdClearAttachments%s", fragFuncSfx.data()]; + plDesc.label = [NSString stringWithFormat: @"CmdClearAttachments"]; plDesc.vertexFunction = getFunctionNamed("vtxCmdClearAttachments"); - plDesc.fragmentFunction = getFunctionNamed((string("fragCmdClearAttachments") + fragFuncSfx).data()); + plDesc.fragmentFunction = getClearFragFunction(attKey); plDesc.sampleCount = attKey.mtlSampleCount; for (uint32_t caIdx = 0; caIdx < kMVKAttachmentFormatDepthStencilIndex; caIdx++) { @@ -133,6 +130,106 @@ id MVKCommandResourceFactory::newCmdClearMTLRenderPipeli return newMTLRenderPipelineState(plDesc); } +id MVKCommandResourceFactory::getBlitFragFunction(MTLPixelFormat mtlPixFmt) { + id mtlFunc = nil; + bool isDepthFormat = mvkMTLPixelFormatIsDepthFormat(mtlPixFmt); + NSString* typeStr = getMTLFormatTypeString(mtlPixFmt); + + @autoreleasepool { + NSMutableString* msl = [NSMutableString stringWithCapacity: (2 * KIBI) ]; + [msl appendLineMVK: @"#include "]; + [msl appendLineMVK: @"using namespace metal;"]; + [msl appendLineMVK]; + [msl appendLineMVK: @"typedef struct {"]; + [msl appendLineMVK: @" float4 gl_Position [[position]];"]; + [msl appendLineMVK: @" float2 v_texCoord;"]; + [msl appendLineMVK: @"} VaryingsPosTex;"]; + [msl appendLineMVK]; + + NSString* funcName = @"fragBlit"; + [msl appendFormat: @"fragment %@4 %@(VaryingsPosTex varyings [[stage_in]],", typeStr, funcName]; + [msl appendLineMVK]; + if (isDepthFormat) { + [msl appendLineMVK: @" depth2d texture [[texture(0)]],"]; + } else { + [msl appendFormat: @" texture2d<%@> texture [[texture(0)]],", typeStr]; + [msl appendLineMVK]; + } + [msl appendLineMVK: @" sampler sampler [[ sampler(0) ]]) {"]; + if (isDepthFormat) { + [msl appendFormat: @" return %@4(texture.sample(sampler, varyings.v_texCoord));", typeStr]; + [msl appendLineMVK]; + } else { + [msl appendLineMVK: @" return texture.sample(sampler, varyings.v_texCoord);"]; + } + [msl appendLineMVK: @"}"]; + + mtlFunc = newMTLFunction(msl, funcName); +// MVKLogDebug("\n%s", msl.UTF8String); + } + return [mtlFunc autorelease]; +} + +id MVKCommandResourceFactory::getClearFragFunction(MVKRPSKeyClearAtt& attKey) { + id mtlFunc = nil; + @autoreleasepool { + NSMutableString* msl = [NSMutableString stringWithCapacity: (2 * KIBI) ]; + [msl appendLineMVK: @"#include "]; + [msl appendLineMVK: @"using namespace metal;"]; + [msl appendLineMVK]; + [msl appendLineMVK: @"typedef struct {"]; + [msl appendLineMVK: @" float4 gl_Position [[position]];"]; + [msl appendLineMVK: @"} VaryingsPos;"]; + [msl appendLineMVK]; + [msl appendLineMVK: @"typedef struct {"]; + [msl appendLineMVK: @" float4 colors[9];"]; + [msl appendLineMVK: @"} ClearColorsIn;"]; + [msl appendLineMVK]; + [msl appendLineMVK: @"typedef struct {"]; + for (uint32_t caIdx = 0; caIdx < kMVKAttachmentFormatDepthStencilIndex; caIdx++) { + if (attKey.isEnabled(caIdx)) { + NSString* typeStr = getMTLFormatTypeString((MTLPixelFormat)attKey.attachmentMTLPixelFormats[caIdx]); + [msl appendFormat: @" %@4 color%u [[color(%u)]];", typeStr, caIdx, caIdx]; + [msl appendLineMVK]; + } + } + [msl appendLineMVK: @"} ClearColorsOut;"]; + [msl appendLineMVK]; + + NSString* funcName = @"fragClear"; + [msl appendFormat: @"fragment ClearColorsOut %@(VaryingsPos varyings [[stage_in]], constant ClearColorsIn& ccIn [[buffer(0)]]) {", funcName]; + [msl appendLineMVK]; + [msl appendLineMVK: @" ClearColorsOut ccOut;"]; + for (uint32_t caIdx = 0; caIdx < kMVKAttachmentFormatDepthStencilIndex; caIdx++) { + if (attKey.isEnabled(caIdx)) { + NSString* typeStr = getMTLFormatTypeString((MTLPixelFormat)attKey.attachmentMTLPixelFormats[caIdx]); + [msl appendFormat: @" ccOut.color%u = %@4(ccIn.colors[%u]);", caIdx, typeStr, caIdx]; + [msl appendLineMVK]; + } + } + [msl appendLineMVK: @" return ccOut;"]; + [msl appendLineMVK: @"}"]; + + mtlFunc = newMTLFunction(msl, funcName); +// MVKLogDebug("\n%s", msl.UTF8String); + } + return [mtlFunc autorelease]; +} + +NSString* MVKCommandResourceFactory::getMTLFormatTypeString(MTLPixelFormat mtlPixFmt) { + switch (mvkFormatTypeFromMTLPixelFormat(mtlPixFmt)) { + case kMVKFormatColorHalf: return @"half"; + case kMVKFormatColorFloat: return @"float"; + case kMVKFormatColorInt8: return @"char"; + case kMVKFormatColorUInt8: return @"uchar"; + case kMVKFormatColorInt16: return @"short"; + case kMVKFormatColorUInt16: return @"ushort"; + case kMVKFormatColorInt32: return @"int"; + case kMVKFormatColorUInt32: return @"uint"; + default: return @"unexpected_type"; + } +} + id MVKCommandResourceFactory::newMTLDepthStencilState(bool useDepth, bool useStencil) { MTLDepthStencilDescriptor* dsDesc = [[[MTLDepthStencilDescriptor alloc] init] autorelease]; @@ -200,22 +297,6 @@ MVKImage* MVKCommandResourceFactory::newMVKImage(MVKImageDescriptorData& imgData return _device->createImage(&createInfo, nullptr); } -string MVKCommandResourceFactory::getFragFunctionSuffix(MTLPixelFormat mtlPixFmt) { - switch (mvkFormatTypeFromMTLPixelFormat(mtlPixFmt)) { - case kMVKFormatDepthStencil: return "DS"; - case kMVKFormatColorUInt: return "U"; - case kMVKFormatColorInt: return "I"; - default: return "F"; - } -} - -string MVKCommandResourceFactory::getFragFunctionSuffix(MVKRPSKeyClearAtt& attKey) { - string suffix; - if (attKey.isEnabledOnly(0)) { suffix += "0"; } - suffix += getFragFunctionSuffix((MTLPixelFormat)attKey.attachmentMTLPixelFormats[0]); - return suffix; -} - id MVKCommandResourceFactory::getFunctionNamed(const char* funcName) { uint64_t startTime = _device->getPerformanceTimestamp(); id mtlFunc = [[_mtlLibrary newFunctionWithName: @(funcName)] autorelease]; @@ -223,6 +304,25 @@ id MVKCommandResourceFactory::getFunctionNamed(const char* funcName return mtlFunc; } +id MVKCommandResourceFactory::newMTLFunction(NSString* mslSrcCode, NSString* funcName) { + uint64_t startTime = _device->getPerformanceTimestamp(); + MTLCompileOptions* shdrOpts = [[MTLCompileOptions new] autorelease]; + NSError* err = nil; + id mtlLib = [[getMTLDevice() newLibraryWithSource: mslSrcCode + options: shdrOpts + error: &err] autorelease]; + _device->addShaderCompilationEventPerformance(_device->_shaderCompilationPerformance.mslCompile, startTime); + if (err) { + mvkNotifyErrorWithText(VK_ERROR_INITIALIZATION_FAILED, "Could not compile support shader from MSL source:\n%s\n %s (code %li) %s", mslSrcCode.UTF8String, err.localizedDescription.UTF8String, (long)err.code, err.localizedFailureReason.UTF8String); + return nil; + } + + startTime = _device->getPerformanceTimestamp(); + id mtlFunc = [mtlLib newFunctionWithName: funcName]; + _device->addShaderCompilationEventPerformance(_device->_shaderCompilationPerformance.functionRetrieval, startTime); + return mtlFunc; +} + id MVKCommandResourceFactory::newMTLRenderPipelineState(MTLRenderPipelineDescriptor* plDesc) { uint64_t startTime = _device->getPerformanceTimestamp(); NSError* err = nil; diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm b/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm index 11452400..a7649b96 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm @@ -136,7 +136,7 @@ VkResult MVKPhysicalDevice::getImageFormatProperties(VkFormat format, pImageFormatProperties->maxMipLevels = mvkMipmapLevels3D(maxExt); pImageFormatProperties->maxArrayLayers = maxLayers; pImageFormatProperties->sampleCounts = _metalFeatures.supportedSampleCounts; - pImageFormatProperties->maxResourceSize = numeric_limits::max(); + pImageFormatProperties->maxResourceSize = kMVKUndefinedLargeUInt64; return VK_SUCCESS; } @@ -379,6 +379,7 @@ void MVKPhysicalDevice::initFeatures() { _features.shaderStorageImageExtendedFormats = true; _features.shaderClipDistance = true; _features.shaderInt16 = true; + _features.multiDrawIndirect = true; #if MVK_IOS _features.textureCompressionETC2 = true; @@ -420,7 +421,7 @@ void MVKPhysicalDevice::initFeatures() { // VkBool32 sampleRateShading; // VkBool32 dualSrcBlend; // done // VkBool32 logicOp; -// VkBool32 multiDrawIndirect; +// VkBool32 multiDrawIndirect; // done // VkBool32 drawIndirectFirstInstance; // VkBool32 depthClamp; // done // VkBool32 depthBiasClamp; // done @@ -620,31 +621,33 @@ void MVKPhysicalDevice::initProperties() { _properties.limits.standardSampleLocations = VK_FALSE; _properties.limits.strictLines = VK_FALSE; + _properties.limits.maxComputeWorkGroupSize[0] = _properties.limits.maxComputeWorkGroupInvocations; + _properties.limits.maxComputeWorkGroupSize[1] = _properties.limits.maxComputeWorkGroupInvocations; + _properties.limits.maxComputeWorkGroupSize[2] = _properties.limits.maxComputeWorkGroupInvocations; + // Features with no specific limits - default to unlimited int values - _properties.limits.maxMemoryAllocationCount = numeric_limits::max(); - _properties.limits.maxSamplerAllocationCount = numeric_limits::max(); - _properties.limits.maxBoundDescriptorSets = numeric_limits::max(); + _properties.limits.maxMemoryAllocationCount = kMVKUndefinedLargeUInt32; + _properties.limits.maxSamplerAllocationCount = kMVKUndefinedLargeUInt32; + _properties.limits.maxBoundDescriptorSets = kMVKUndefinedLargeUInt32; - _properties.limits.maxComputeWorkGroupCount[0] = numeric_limits::max(); - _properties.limits.maxComputeWorkGroupCount[1] = numeric_limits::max(); - _properties.limits.maxComputeWorkGroupCount[2] = numeric_limits::max(); + _properties.limits.maxComputeWorkGroupCount[0] = kMVKUndefinedLargeUInt32; + _properties.limits.maxComputeWorkGroupCount[1] = kMVKUndefinedLargeUInt32; + _properties.limits.maxComputeWorkGroupCount[2] = kMVKUndefinedLargeUInt32; - _properties.limits.maxComputeWorkGroupSize[0] = numeric_limits::max(); - _properties.limits.maxComputeWorkGroupSize[1] = numeric_limits::max(); - _properties.limits.maxComputeWorkGroupSize[2] = numeric_limits::max(); + _properties.limits.maxDrawIndexedIndexValue = numeric_limits::max() - 1; + _properties.limits.maxDrawIndirectCount = kMVKUndefinedLargeUInt32; - _properties.limits.maxDrawIndexedIndexValue = numeric_limits::max(); - _properties.limits.maxDrawIndirectCount = numeric_limits::max(); + _properties.limits.minTexelOffset = -8; + _properties.limits.maxTexelOffset = 7; + _properties.limits.minTexelGatherOffset = _properties.limits.minTexelOffset; + _properties.limits.maxTexelGatherOffset = _properties.limits.maxTexelOffset; - _properties.limits.minTexelOffset = numeric_limits::min(); - _properties.limits.maxTexelOffset = numeric_limits::max(); - _properties.limits.minTexelGatherOffset = numeric_limits::min(); - _properties.limits.maxTexelGatherOffset = numeric_limits::max(); - - _properties.limits.maxClipDistances = numeric_limits::max(); - _properties.limits.maxCombinedClipAndCullDistances = numeric_limits::max(); + _properties.limits.maxClipDistances = kMVKUndefinedLargeUInt32; + _properties.limits.maxCullDistances = 0; // unsupported + _properties.limits.maxCombinedClipAndCullDistances = _properties.limits.maxClipDistances + + _properties.limits.maxCullDistances; // Features with unknown limits - default to Vulkan required limits @@ -683,9 +686,6 @@ void MVKPhysicalDevice::initProperties() { _properties.limits.minInterpolationOffset = 0.0; _properties.limits.maxInterpolationOffset = 0.0; _properties.limits.subPixelInterpolationOffsetBits = 0; - - _properties.limits.maxCullDistances = 0; - } diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm b/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm index a31e55ef..d2f62fb1 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm @@ -278,7 +278,12 @@ MTLRenderPipelineDescriptor* MVKGraphicsPipeline::getMTLRenderPipelineDescriptor // Vertex shader if (mvkAreFlagsEnabled(pSS->stage, VK_SHADER_STAGE_VERTEX_BIT)) { shaderContext.options.entryPointStage = spv::ExecutionModelVertex; - plDesc.vertexFunction = mvkShdrMod->getMTLFunction(&shaderContext, pSS->pSpecializationInfo, _pipelineCache).mtlFunction; + id mtlFunction = mvkShdrMod->getMTLFunction(&shaderContext, pSS->pSpecializationInfo, _pipelineCache).mtlFunction; + if ( !mtlFunction ) { + setConfigurationResult(mvkNotifyErrorWithText(VK_ERROR_INITIALIZATION_FAILED, "Vertex shader function could not be compiled into pipeline. See previous error.")); + return nil; + } + plDesc.vertexFunction = mtlFunction; } // Fragment shader @@ -418,14 +423,17 @@ MVKComputePipeline::MVKComputePipeline(MVKDevice* device, _mtlThreadgroupSize = shaderFunc.threadGroupSize; _mtlPipelineState = nil; - NSError* psError = nil; - uint64_t startTime = _device->getPerformanceTimestamp(); - _mtlPipelineState = [getMTLDevice() newComputePipelineStateWithFunction: shaderFunc.mtlFunction error: &psError]; // retained - if (psError) { - setConfigurationResult(mvkNotifyErrorWithText(VK_ERROR_INITIALIZATION_FAILED, "Could not create compute pipeline:\n%s.", psError.description.UTF8String)); - } - _device->addShaderCompilationEventPerformance(_device->_shaderCompilationPerformance.pipelineCompile, startTime); - + if (shaderFunc.mtlFunction) { + NSError* psError = nil; + uint64_t startTime = _device->getPerformanceTimestamp(); + _mtlPipelineState = [getMTLDevice() newComputePipelineStateWithFunction: shaderFunc.mtlFunction error: &psError]; // retained + if (psError) { + setConfigurationResult(mvkNotifyErrorWithText(VK_ERROR_INITIALIZATION_FAILED, "Could not create compute pipeline:\n%s.", psError.description.UTF8String)); + } + _device->addShaderCompilationEventPerformance(_device->_shaderCompilationPerformance.pipelineCompile, startTime); + } else { + setConfigurationResult(mvkNotifyErrorWithText(VK_ERROR_INITIALIZATION_FAILED, "Compute shader function could not be compiled into pipeline. See previous error.")); + } } } diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm b/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm index 39e6b1c6..0986215a 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm @@ -255,7 +255,7 @@ MVKRenderPassAttachment::MVKRenderPassAttachment(MVKRenderPass* renderPass, _attachmentIndex = uint32_t(_renderPass->_attachments.size()); // Determine the indices of the first and last render subpasses to use that attachment. - _firstUseSubpassIdx = kMVKMaxUnsigned; + _firstUseSubpassIdx = kMVKUndefinedLargeUInt32; _lastUseSubpassIdx = 0; for (auto& subPass : _renderPass->_subpasses) { if (subPass.isUsingAttachmentAt(_attachmentIndex)) { diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKSwapchain.mm b/MoltenVK/MoltenVK/GPUObjects/MVKSwapchain.mm index 8f4380b3..bbc9ab8e 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKSwapchain.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKSwapchain.mm @@ -65,7 +65,9 @@ VkResult MVKSwapchain::acquireNextImageKHR(uint64_t timeout, uint32_t* pImageIndex) { // Find the image that has the smallest availability measure uint32_t minWaitIndex = 0; - MVKSwapchainImageAvailability minAvailability = { .acquisitionID = numeric_limits::max(), .waitCount = numeric_limits::max(), .isAvailable = false }; + MVKSwapchainImageAvailability minAvailability = { .acquisitionID = kMVKUndefinedLargeUInt64, + .waitCount = kMVKUndefinedLargeUInt32, + .isAvailable = false }; for (MVKSwapchainImage* mvkSCImg : _surfaceImages) { const MVKSwapchainImageAvailability* currAvailability = mvkSCImg->getAvailability(); if (*currAvailability < minAvailability) { diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKSync.mm b/MoltenVK/MoltenVK/GPUObjects/MVKSync.mm index e74b8255..c2e35439 100644 --- a/MoltenVK/MoltenVK/GPUObjects/MVKSync.mm +++ b/MoltenVK/MoltenVK/GPUObjects/MVKSync.mm @@ -55,7 +55,7 @@ bool MVKSemaphoreImpl::wait(uint64_t timeout, bool reserveAgain) { isDone = true; } else { // Limit timeout to avoid overflow since wait_for() uses wait_until() - uint64_t nanoTimeout = min(timeout, numeric_limits::max() >> 4); + uint64_t nanoTimeout = min(timeout, kMVKUndefinedLargeUInt64); chrono::nanoseconds nanos(nanoTimeout); isDone = _blocker.wait_for(lock, nanos, [this]{ return isClear(); }); } diff --git a/MoltenVK/MoltenVK/Utility/MVKFoundation.h b/MoltenVK/MoltenVK/Utility/MVKFoundation.h index fd0a40e0..631be2e6 100644 --- a/MoltenVK/MoltenVK/Utility/MVKFoundation.h +++ b/MoltenVK/MoltenVK/Utility/MVKFoundation.h @@ -28,8 +28,17 @@ #pragma mark Math -/** Maximum value of any variable of unsigned integral type. */ -#define kMVKMaxUnsigned (~0U) +/** + * The following constants are used to indicate values that have no defined limit. + * They are ridiculously large numbers, but low enough to be safely used as both + * uint and int values without risking overflowing between positive and negative values. + */ +static int32_t kMVKUndefinedLargeNegativeInt32 = std::numeric_limits::min() / 2; +static int32_t kMVKUndefinedLargePositiveInt32 = std::numeric_limits::max() / 2; +static uint32_t kMVKUndefinedLargeUInt32 = kMVKUndefinedLargePositiveInt32; +static int64_t kMVKUndefinedLargeNegativeInt64 = std::numeric_limits::min() / 2; +static int64_t kMVKUndefinedLargePositiveInt64 = std::numeric_limits::max() / 2; +static uint64_t kMVKUndefinedLargeUInt64 = kMVKUndefinedLargePositiveInt64; // Common scaling multipliers #define KIBI (1024) diff --git a/MoltenVK/MoltenVK/Utility/NSString+MoltenVK.h b/MoltenVK/MoltenVK/Utility/NSString+MoltenVK.h new file mode 100644 index 00000000..361dc031 --- /dev/null +++ b/MoltenVK/MoltenVK/Utility/NSString+MoltenVK.h @@ -0,0 +1,40 @@ +/* + * NSString+MoltenVK.h + * + * Copyright (c) 2014-2018 The Brenwill Workshop Ltd. (http://www.brenwill.com) + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + + +/* This file contains convenient functions for accessing Metal components during execution. */ + +#pragma once + +#include "MVKCommonEnvironment.h" +#import + + +#pragma mark - +#pragma mark NSMutableString extension + +/** Extensions to NSMutableString to support MoltenVK. */ +@interface NSMutableString (MoltenVK) + +/** Appends the string and a new line. */ +-(void) appendLineMVK:(NSString*) aString; + +/** Appends an empty new line. */ +-(void) appendLineMVK; + +@end diff --git a/MoltenVK/MoltenVK/Utility/NSString+MoltenVK.mm b/MoltenVK/MoltenVK/Utility/NSString+MoltenVK.mm new file mode 100644 index 00000000..791c18f0 --- /dev/null +++ b/MoltenVK/MoltenVK/Utility/NSString+MoltenVK.mm @@ -0,0 +1,37 @@ +/* + * NSString+MoltenVK.mm + * + * Copyright (c) 2014-2018 The Brenwill Workshop Ltd. (http://www.brenwill.com) + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + + +#include "NSString+MoltenVK.h" + + +#pragma mark - +#pragma mark NSMutableString extension + +@implementation NSMutableString (MoltenVK) + +-(void) appendLineMVK:(NSString*) aString { + [self appendString: aString]; + [self appendLineMVK]; +} + +-(void) appendLineMVK { + [self appendString: @"\n"]; +} + +@end diff --git a/MoltenVK/MoltenVK/Vulkan/mvk_datatypes.mm b/MoltenVK/MoltenVK/Vulkan/mvk_datatypes.mm index df29bb04..66a57634 100644 --- a/MoltenVK/MoltenVK/Vulkan/mvk_datatypes.mm +++ b/MoltenVK/MoltenVK/Vulkan/mvk_datatypes.mm @@ -194,118 +194,118 @@ static const MVKFormatDesc _formatDescriptions[] { MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8_SNORM, MTLPixelFormatR8Snorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 1, MTLVertexFormatChar2Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 1, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 1, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8_UINT, MTLPixelFormatR8Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 1, MTLVertexFormatUChar2, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8_SINT, MTLPixelFormatR8Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 1, MTLVertexFormatChar2, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8_UINT, MTLPixelFormatR8Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 1, MTLVertexFormatUChar2, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8_SINT, MTLPixelFormatR8Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 1, MTLVertexFormatChar2, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8_SRGB, MTLPixelFormatR8Unorm_sRGB, MTLPixelFormatInvalid, 8.0, kMTLFmtNA, 1, 1, 1, MTLVertexFormatUChar2, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_UNORM, MTLPixelFormatRG8Unorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatUChar2Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_SNORM, MTLPixelFormatRG8Snorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatChar2Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 2, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 2, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_UINT, MTLPixelFormatRG8Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatUChar2, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_SINT, MTLPixelFormatRG8Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatChar2, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_UINT, MTLPixelFormatRG8Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatUChar2, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_SINT, MTLPixelFormatRG8Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatChar2, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8_SRGB, MTLPixelFormatRG8Unorm_sRGB, MTLPixelFormatInvalid, 8.0, kMTLFmtNA, 1, 1, 2, MTLVertexFormatUChar2, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_UNORM, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatUChar3Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_SNORM, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatChar3Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatUChar3, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatChar3, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatUChar3, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatChar3, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8_SRGB, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatUChar3, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_UNORM, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_SNORM, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorInt, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8_SRGB, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 3, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_UNORM, MTLPixelFormatRGBA8Unorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUChar4Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_SNORM, MTLPixelFormatRGBA8Snorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatChar4Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_UINT, MTLPixelFormatRGBA8Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUChar4, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_SINT, MTLPixelFormatRGBA8Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatChar4, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_UINT, MTLPixelFormatRGBA8Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUChar4, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_SINT, MTLPixelFormatRGBA8Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatChar4, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R8G8B8A8_SRGB, MTLPixelFormatRGBA8Unorm_sRGB, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUChar4, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_UNORM, MTLPixelFormatBGRA8Unorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_SNORM, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_B8G8R8A8_SRGB, MTLPixelFormatBGRA8Unorm_sRGB, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_UNORM_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_SNORM_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_USCALED_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_SSCALED_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_UINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_SINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_SRGB_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_UNORM_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_SNORM_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_USCALED_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_SSCALED_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_UINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_SINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A8B8G8R8_SRGB_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2R10G10B10_UNORM_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2R10G10B10_SNORM_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2R10G10B10_USCALED_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2R10G10B10_SSCALED_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2R10G10B10_UINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2R10G10B10_SINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2R10G10B10_UINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2R10G10B10_SINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2B10G10R10_UNORM_PACK32, MTLPixelFormatRGB10A2Unorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), // Vulkan packed is reversed MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2B10G10R10_SNORM_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2B10G10R10_USCALED_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2B10G10R10_SSCALED_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2B10G10R10_UINT_PACK32, MTLPixelFormatRGB10A2Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_FEATS ), // Vulkan packed is reversed - MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2B10G10R10_SINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt, MVK_FMT_ALL_CLR_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2B10G10R10_UINT_PACK32, MTLPixelFormatRGB10A2Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_FEATS ), // Vulkan packed is reversed + MVK_MAKE_FMT_STRUCT( VK_FORMAT_A2B10G10R10_SINT_PACK32, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_UNORM, MTLPixelFormatR16Unorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatUShort2Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_SNORM, MTLPixelFormatR16Snorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatShort2Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 2, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 2, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_UINT, MTLPixelFormatR16Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatUShort2, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_SINT, MTLPixelFormatR16Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatShort2, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_UINT, MTLPixelFormatR16Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatUShort2, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_SINT, MTLPixelFormatR16Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatShort2, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16_SFLOAT, MTLPixelFormatR16Float, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 2, MTLVertexFormatHalf2, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_UNORM, MTLPixelFormatRG16Unorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUShort2Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_SNORM, MTLPixelFormatRG16Snorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatShort2Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 4, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_UINT, MTLPixelFormatRG16Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUShort2, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_SINT, MTLPixelFormatRG16Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatShort2, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_UINT, MTLPixelFormatRG16Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUShort2, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_SINT, MTLPixelFormatRG16Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatShort2, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16_SFLOAT, MTLPixelFormatRG16Float, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatHalf2, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_UNORM, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatUShort3Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_SNORM, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatShort3Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatUShort3, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatShort3, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatUShort3, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatShort3, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16_SFLOAT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 6, MTLVertexFormatHalf3, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_UNORM, MTLPixelFormatRGBA16Unorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatUShort4Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_SNORM, MTLPixelFormatRGBA16Snorm, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatShort4Normalized, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_USCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 8, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_SSCALED, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 8, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_UINT, MTLPixelFormatRGBA16Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatUShort4, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_SINT, MTLPixelFormatRGBA16Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatShort4, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_UINT, MTLPixelFormatRGBA16Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatUShort4, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_SINT, MTLPixelFormatRGBA16Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatShort4, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R16G16B16A16_SFLOAT, MTLPixelFormatRGBA16Float, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatHalf4, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32_UINT, MTLPixelFormatR32Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUInt, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32_SINT, MTLPixelFormatR32Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatInt, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32_UINT, MTLPixelFormatR32Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatUInt, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32_SINT, MTLPixelFormatR32Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatInt, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32_SFLOAT, MTLPixelFormatR32Float, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 4, MTLVertexFormatFloat, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32_UINT, MTLPixelFormatRG32Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatUInt2, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32_SINT, MTLPixelFormatRG32Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatInt2, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32_UINT, MTLPixelFormatRG32Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatUInt2, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32_SINT, MTLPixelFormatRG32Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatInt2, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32_SFLOAT, MTLPixelFormatRG32Float, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 8, MTLVertexFormatFloat2, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 12, MTLVertexFormatUInt3, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 12, MTLVertexFormatInt3, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 12, MTLVertexFormatUInt3, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32_SINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 12, MTLVertexFormatInt3, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32_SFLOAT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 12, MTLVertexFormatFloat3, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32A32_UINT, MTLPixelFormatRGBA32Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 16, MTLVertexFormatUInt4, kMVKFormatColorUInt, MVK_FMT_ALL_CLR_VTX_FEATS ), - MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32A32_SINT, MTLPixelFormatRGBA32Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 16, MTLVertexFormatInt4, kMVKFormatColorInt, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32A32_UINT, MTLPixelFormatRGBA32Uint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 16, MTLVertexFormatUInt4, kMVKFormatColorUInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), + MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32A32_SINT, MTLPixelFormatRGBA32Sint, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 16, MTLVertexFormatInt4, kMVKFormatColorInt32, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R32G32B32A32_SFLOAT, MTLPixelFormatRGBA32Float, MTLPixelFormatInvalid, 8.0, 10.11, 1, 1, 16, MTLVertexFormatFloat4, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_VTX_FEATS ), MVK_MAKE_FMT_STRUCT( VK_FORMAT_R64_UINT, MTLPixelFormatInvalid, MTLPixelFormatInvalid, kMTLFmtNA, kMTLFmtNA, 1, 1, 8, MTLVertexFormatInvalid, kMVKFormatColorFloat, MVK_FMT_ALL_CLR_FEATS ), @@ -598,19 +598,24 @@ MVK_PUBLIC_SYMBOL MTLClearColor mvkMTLClearColorFromVkClearValue(VkClearValue vk VkFormat vkFormat) { MTLClearColor mtlClr; switch (mvkFormatTypeFromVkFormat(vkFormat)) { + case kMVKFormatColorHalf: case kMVKFormatColorFloat: mtlClr.red = vkClearValue.color.float32[0]; mtlClr.green = vkClearValue.color.float32[1]; mtlClr.blue = vkClearValue.color.float32[2]; mtlClr.alpha = vkClearValue.color.float32[3]; break; - case kMVKFormatColorUInt: + case kMVKFormatColorUInt8: + case kMVKFormatColorUInt16: + case kMVKFormatColorUInt32: mtlClr.red = vkClearValue.color.uint32[0]; mtlClr.green = vkClearValue.color.uint32[1]; mtlClr.blue = vkClearValue.color.uint32[2]; mtlClr.alpha = vkClearValue.color.uint32[3]; break; - case kMVKFormatColorInt: + case kMVKFormatColorInt8: + case kMVKFormatColorInt16: + case kMVKFormatColorInt32: mtlClr.red = vkClearValue.color.int32[0]; mtlClr.green = vkClearValue.color.int32[1]; mtlClr.blue = vkClearValue.color.int32[2]; diff --git a/MoltenVK/MoltenVK/Vulkan/vulkan.mm b/MoltenVK/MoltenVK/Vulkan/vulkan.mm index 8aba8b38..f95adea3 100644 --- a/MoltenVK/MoltenVK/Vulkan/vulkan.mm +++ b/MoltenVK/MoltenVK/Vulkan/vulkan.mm @@ -1122,22 +1122,22 @@ MVK_PUBLIC_SYMBOL void vkCmdDrawIndirect( VkCommandBuffer commandBuffer, VkBuffer buffer, VkDeviceSize offset, - uint32_t count, + uint32_t drawCount, uint32_t stride) { MVKCommandBuffer* cmdBuff = MVKCommandBuffer::getMVKCommandBuffer(commandBuffer); - mvkCmdDrawIndirect(cmdBuff, buffer, offset, count, stride); + mvkCmdDrawIndirect(cmdBuff, buffer, offset, drawCount, stride); } MVK_PUBLIC_SYMBOL void vkCmdDrawIndexedIndirect( VkCommandBuffer commandBuffer, VkBuffer buffer, VkDeviceSize offset, - uint32_t count, + uint32_t drawCount, uint32_t stride) { MVKCommandBuffer* cmdBuff = MVKCommandBuffer::getMVKCommandBuffer(commandBuffer); - mvkCmdDrawIndexedIndirect(cmdBuff, buffer, offset, count, stride); + mvkCmdDrawIndexedIndirect(cmdBuff, buffer, offset, drawCount, stride); } MVK_PUBLIC_SYMBOL void vkCmdDispatch( diff --git a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.cpp b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.cpp index 4211adb9..e7cdeb92 100644 --- a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.cpp +++ b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.cpp @@ -120,7 +120,7 @@ MVK_PUBLIC_SYMBOL void SPIRVToMSLConverterContext::alignUsageWith(const SPIRVToM #pragma mark SPIRVToMSLConverter /** Populates content extracted from the SPRI-V compiler. */ -void populateFromCompiler(spirv_cross::Compiler& compiler, SPIRVEntryPoint& entryPoint, SPIRVToMSLConverterOptions& options); +void populateFromCompiler(spirv_cross::Compiler* pCompiler, SPIRVEntryPoint& entryPoint, SPIRVToMSLConverterOptions& options); MVK_PUBLIC_SYMBOL void SPIRVToMSLConverter::setSPIRV(const vector& spirv) { _spirv = spirv; } @@ -171,68 +171,74 @@ MVK_PUBLIC_SYMBOL bool SPIRVToMSLConverter::convert(SPIRVToMSLConverterContext& resBindings.push_back(rb); } - spirv_cross::CompilerMSL mslCompiler(_spirv); - if (context.options.hasEntryPoint()) { - mslCompiler.set_entry_point(context.options.entryPointName, context.options.entryPointStage); - } - - // Establish the MSL options for the compiler - // This needs to be done in two steps...for CompilerMSL and its superclass. - auto mslOpts = mslCompiler.get_msl_options(); - -#if MVK_MACOS - mslOpts.platform = spirv_cross::CompilerMSL::Options::macOS; -#endif -#if MVK_IOS - mslOpts.platform = spirv_cross::CompilerMSL::Options::iOS; -#endif - - mslOpts.msl_version = context.options.mslVersion; - mslOpts.enable_point_size_builtin = context.options.isRenderingPoints; - mslOpts.resolve_specialized_array_lengths = true; - mslCompiler.set_msl_options(mslOpts); - - auto scOpts = mslCompiler.get_common_options(); - scOpts.vertex.flip_vert_y = context.options.shouldFlipVertexY; - mslCompiler.set_common_options(scOpts); + spirv_cross::CompilerMSL* pMSLCompiler = nullptr; #ifndef SPIRV_CROSS_EXCEPTIONS_TO_ASSERTIONS try { #endif - _msl = mslCompiler.compile(&vtxAttrs, &resBindings); + pMSLCompiler = new spirv_cross::CompilerMSL(_spirv); + + if (context.options.hasEntryPoint()) { + pMSLCompiler->set_entry_point(context.options.entryPointName, context.options.entryPointStage); + } + + // Establish the MSL options for the compiler + // This needs to be done in two steps...for CompilerMSL and its superclass. + auto mslOpts = pMSLCompiler->get_msl_options(); + +#if MVK_MACOS + mslOpts.platform = spirv_cross::CompilerMSL::Options::macOS; +#endif +#if MVK_IOS + mslOpts.platform = spirv_cross::CompilerMSL::Options::iOS; +#endif + + mslOpts.msl_version = context.options.mslVersion; + mslOpts.enable_point_size_builtin = context.options.isRenderingPoints; + mslOpts.resolve_specialized_array_lengths = true; + pMSLCompiler->set_msl_options(mslOpts); + + auto scOpts = pMSLCompiler->get_common_options(); + scOpts.vertex.flip_vert_y = context.options.shouldFlipVertexY; + pMSLCompiler->set_common_options(scOpts); + + _msl = pMSLCompiler->compile(&vtxAttrs, &resBindings); if (shouldLogMSL) { logSource(_msl, "MSL", "Converted"); } #ifndef SPIRV_CROSS_EXCEPTIONS_TO_ASSERTIONS } catch (spirv_cross::CompilerError& ex) { string errMsg("MSL conversion error: "); errMsg += ex.what(); logError(errMsg.data()); - if (shouldLogMSL) { - _msl = mslCompiler.get_partial_source(); + if (shouldLogMSL && pMSLCompiler) { + _msl = pMSLCompiler->get_partial_source(); logSource(_msl, "MSL", "Partially converted"); } } #endif // Populate content extracted from the SPRI-V compiler. - populateFromCompiler(mslCompiler, _entryPoint, context.options); + populateFromCompiler(pMSLCompiler, _entryPoint, context.options); // To check GLSL conversion if (shouldLogGLSL) { - spirv_cross::CompilerGLSL glslCompiler(_spirv); - string glsl; + spirv_cross::CompilerGLSL* pGLSLCompiler = nullptr; + #ifndef SPIRV_CROSS_EXCEPTIONS_TO_ASSERTIONS - try { + try { #endif - glsl = glslCompiler.compile(); + pGLSLCompiler = new spirv_cross::CompilerGLSL(_spirv); + string glsl = pGLSLCompiler->compile(); logSource(glsl, "GLSL", "Estimated original"); #ifndef SPIRV_CROSS_EXCEPTIONS_TO_ASSERTIONS } catch (spirv_cross::CompilerError& ex) { string errMsg("Original GLSL extraction error: "); errMsg += ex.what(); logMsg(errMsg.data()); - glsl = glslCompiler.get_partial_source(); - logSource(glsl, "GLSL", "Partially converted"); + if (pGLSLCompiler) { + string glsl = pGLSLCompiler->get_partial_source(); + logSource(glsl, "GLSL", "Partially converted"); + } } #endif } @@ -322,16 +328,18 @@ void SPIRVToMSLConverter::logSource(string& src, const char* srcLang, const char #pragma mark Support functions -void populateFromCompiler(spirv_cross::Compiler& compiler, SPIRVEntryPoint& entryPoint, SPIRVToMSLConverterOptions& options) { +void populateFromCompiler(spirv_cross::Compiler* pCompiler, SPIRVEntryPoint& entryPoint, SPIRVToMSLConverterOptions& options) { + + if ( !pCompiler ) { return; } spirv_cross::SPIREntryPoint spvEP; if (options.hasEntryPoint()) { - spvEP = compiler.get_entry_point(options.entryPointName, options.entryPointStage); + spvEP = pCompiler->get_entry_point(options.entryPointName, options.entryPointStage); } else { - const auto& entryPoints = compiler.get_entry_points_and_stages(); + const auto& entryPoints = pCompiler->get_entry_points_and_stages(); if ( !entryPoints.empty() ) { auto& ep = entryPoints[0]; - spvEP = compiler.get_entry_point(ep.name, ep.execution_model); + spvEP = pCompiler->get_entry_point(ep.name, ep.execution_model); } } @@ -344,7 +352,7 @@ void populateFromCompiler(spirv_cross::Compiler& compiler, SPIRVEntryPoint& entr entryPoint.workgroupSize.depth = max(wgSize.z, minDim); spirv_cross::SpecializationConstant width, height, depth; - entryPoint.workgroupSizeId.constant = compiler.get_work_group_size_specialization_constants(width, height, depth); + entryPoint.workgroupSizeId.constant = pCompiler->get_work_group_size_specialization_constants(width, height, depth); entryPoint.workgroupSizeId.width = width.constant_id; entryPoint.workgroupSizeId.height = height.constant_id; entryPoint.workgroupSizeId.depth = depth.constant_id; diff --git a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.h b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.h index b5c9c67d..c7b16e56 100644 --- a/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.h +++ b/MoltenVKShaderConverter/MoltenVKSPIRVToMSLConverter/SPIRVToMSLConverter.h @@ -156,9 +156,6 @@ namespace mvk { } workgroupSizeId; } SPIRVEntryPoint; - /** Holds a map of entry point info, indexed by the SPIRV entry point name. */ -// typedef std::unordered_map SPIRVEntryPointsByName; - /** Special constant used in a MSLResourceBinding descriptorSet element to indicate the bindings for the push constants. */ static const uint32_t kPushConstDescSet = std::numeric_limits::max(); diff --git a/MoltenVKShaderConverter/MoltenVKShaderConverter.xcodeproj/xcshareddata/xcschemes/MoltenVKShaderConverter.xcscheme b/MoltenVKShaderConverter/MoltenVKShaderConverter.xcodeproj/xcshareddata/xcschemes/MoltenVKShaderConverter.xcscheme index 45a5b0ed..104a2c17 100644 --- a/MoltenVKShaderConverter/MoltenVKShaderConverter.xcodeproj/xcshareddata/xcschemes/MoltenVKShaderConverter.xcscheme +++ b/MoltenVKShaderConverter/MoltenVKShaderConverter.xcodeproj/xcshareddata/xcschemes/MoltenVKShaderConverter.xcscheme @@ -72,19 +72,19 @@ + isEnabled = "YES"> + argument = "/Users/bill/Documents/Dev/iOSProjects/Molten/MoltenVK-bh/External/SPIRV-Cross/shaders-msl/frag/gather-offset.frag" + isEnabled = "YES"> + isEnabled = "NO"> + isEnabled = "NO">