diff --git a/Docs/MoltenVK_Runtime_UserGuide.md b/Docs/MoltenVK_Runtime_UserGuide.md
index 28e913f..51a63e3 100644
--- a/Docs/MoltenVK_Runtime_UserGuide.md
+++ b/Docs/MoltenVK_Runtime_UserGuide.md
@@ -286,6 +286,7 @@
 - `VK_KHR_maintenance2`
 - `VK_KHR_maintenance3`
 - `VK_KHR_multiview`
+- `VK_KHR_portability_subset`
 - `VK_KHR_push_descriptor`
 - `VK_KHR_relaxed_block_layout`
 - `VK_KHR_sampler_mirror_clamp_to_edge` *(macOS)*
@@ -315,7 +316,6 @@
 - `VK_EXT_swapchain_colorspace`
 - `VK_EXT_vertex_attribute_divisor`
 - `VK_EXT_texel_buffer_alignment` *(requires Metal 2.0)*
-- `VK_EXTX_portability_subset`
 - `VK_MVK_ios_surface` *(iOS) (Obsolete. Use `VK_EXT_metal_surface` instead.)*
 - `VK_MVK_macos_surface` *(macOS) (Obsolete. Use `VK_EXT_metal_surface` instead.)*
 - `VK_MVK_moltenvk`
diff --git a/Docs/Whats_New.md b/Docs/Whats_New.md
index b34e5a7..9eb11bf 100644
--- a/Docs/Whats_New.md
+++ b/Docs/Whats_New.md
@@ -25,6 +25,7 @@
 	- A feature struct for `VK_KHR_shader_draw_parameters`
 	- All extensions that were promoted to core in Vulkan 1.1
 - Add support for extensions:
+	- `VK_KHR_portability_subset`
 	- `VK_KHR_create_renderpass2`
 	- `VK_KHR_external_fence` (non-functional groundwork for future extensions,
 	  including support for GCD and Mach semaphores)
@@ -35,6 +36,7 @@
 	- `VK_KHR_external_semaphore_capabilities` (non-functional groundwork for
 	  future `MTLSharedEvent` Vulkan extension)
 	- `VK_KHR_multiview`
+- Remove support for obsolete `VK_EXTX_portability_subset` extension.
 - Improve performance of tessellation control pipeline stage by processing multiple 
   patches per workgroup.
 - `vkCmdBindDescriptorSets` order `pDynamicOffsets` by descriptor binding number 
diff --git a/ExternalRevisions/Vulkan-Headers_repo_revision b/ExternalRevisions/Vulkan-Headers_repo_revision
index d347369..77ca791 100644
--- a/ExternalRevisions/Vulkan-Headers_repo_revision
+++ b/ExternalRevisions/Vulkan-Headers_repo_revision
@@ -1 +1 @@
-83825d55c7d522931124696ecb07ed48f2693e5c
+7f9879b1b1fab53f719a9ed5e6e29533b10972b2
diff --git a/ExternalRevisions/Vulkan-Portability_repo_revision b/ExternalRevisions/Vulkan-Portability_repo_revision
deleted file mode 100644
index 3a9f9eb..0000000
--- a/ExternalRevisions/Vulkan-Portability_repo_revision
+++ /dev/null
@@ -1 +0,0 @@
-53be040f04ce55463d0e5b25fd132f45f003e903
diff --git a/MoltenVK/MoltenVK.xcodeproj/project.pbxproj b/MoltenVK/MoltenVK.xcodeproj/project.pbxproj
index b3b771f..3de75fc 100644
--- a/MoltenVK/MoltenVK.xcodeproj/project.pbxproj
+++ b/MoltenVK/MoltenVK.xcodeproj/project.pbxproj
@@ -571,7 +571,6 @@
 		A9E53DFE21064F84002781DD /* MTLRenderPipelineDescriptor+MoltenVK.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = "MTLRenderPipelineDescriptor+MoltenVK.h"; sourceTree = "<group>"; };
 		A9F0429D1FB4CF82009FCCB8 /* MVKCommonEnvironment.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = MVKCommonEnvironment.h; sourceTree = "<group>"; };
 		A9F0429E1FB4CF82009FCCB8 /* MVKLogging.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = MVKLogging.h; sourceTree = "<group>"; };
-		A9F2559121F96814008C7785 /* vulkan-portability */ = {isa = PBXFileReference; lastKnownFileType = folder; path = "vulkan-portability"; sourceTree = "<group>"; };
 		A9F3D9D924732A4C00745190 /* MVKSmallVectorAllocator.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = MVKSmallVectorAllocator.h; sourceTree = "<group>"; };
 		A9F3D9DB24732A4D00745190 /* MVKSmallVector.h */ = {isa = PBXFileReference; fileEncoding = 4; lastKnownFileType = sourcecode.c.h; path = MVKSmallVector.h; sourceTree = "<group>"; };
 /* End PBXFileReference section */
@@ -750,7 +749,6 @@
 			isa = PBXGroup;
 			children = (
 				A9AD67C72054DD6C00ED3C08 /* vulkan */,
-				A9F2559121F96814008C7785 /* vulkan-portability */,
 			);
 			path = include;
 			sourceTree = "<group>";
diff --git a/MoltenVK/MoltenVK/API/mvk_vulkan.h b/MoltenVK/MoltenVK/API/mvk_vulkan.h
index 757bb7e..f234ead 100644
--- a/MoltenVK/MoltenVK/API/mvk_vulkan.h
+++ b/MoltenVK/MoltenVK/API/mvk_vulkan.h
@@ -35,6 +35,8 @@
 
 #define VK_USE_PLATFORM_METAL_EXT				1
 
+#define VK_ENABLE_BETA_EXTENSIONS				1		// VK_KHR_portability_subset
+
 #ifdef __IPHONE_OS_VERSION_MAX_ALLOWED
 #	define VK_USE_PLATFORM_IOS_MVK				1
 #endif
@@ -44,6 +46,5 @@
 #endif
 
 #include <vulkan/vulkan.h>
-#include <vulkan-portability/vk_extx_portability_subset.h>
 
 #endif
diff --git a/MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm b/MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm
index 964af0d..e657b04 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm
+++ b/MoltenVK/MoltenVK/Commands/MVKCmdDraw.mm
@@ -629,6 +629,11 @@
 					[mtlTessCtlEncoder dispatchThreadgroups: MTLSizeMake(mvkCeilingDivide<NSUInteger>(_drawCount, mtlConvertState.threadExecutionWidth), 1, 1)
 									  threadsPerThreadgroup: MTLSizeMake(mtlConvertState.threadExecutionWidth, 1, 1)];
 				}
+                // Mark pipelines, resources, and vertex push constants as dirty
+                // so I apply them during the next stage.
+                cmdEncoder->_graphicsPipelineState.beginMetalRenderPass();
+                cmdEncoder->_graphicsResourcesState.beginMetalRenderPass();
+                cmdEncoder->getPushConstants(VK_SHADER_STAGE_VERTEX_BIT)->beginMetalRenderPass();
             } else if (drawIdx == 0 && needsInstanceAdjustment) {
                 // Similarly, for multiview, we need to adjust the instance count now.
                 // Unfortunately, this requires switching to compute.
@@ -954,6 +959,11 @@
 													 indirectBufferOffset: mtlTempIndBuffOfst
                                                     threadsPerThreadgroup: MTLSizeMake(vtxThreadExecWidth, 1, 1)];
 				mtlIndBuffOfst += sizeof(MTLDrawIndexedPrimitivesIndirectArguments);
+                // Mark pipeline, resources, and vertex push constants as dirty
+                // so I apply them during the next stage.
+                cmdEncoder->_graphicsPipelineState.beginMetalRenderPass();
+                cmdEncoder->_graphicsResourcesState.beginMetalRenderPass();
+                cmdEncoder->getPushConstants(VK_SHADER_STAGE_VERTEX_BIT)->beginMetalRenderPass();
             } else if (drawIdx == 0 && needsInstanceAdjustment) {
                 // Similarly, for multiview, we need to adjust the instance count now.
                 // Unfortunately, this requires switching to compute. Luckily, we don't also
@@ -1089,11 +1099,11 @@
 						}
 
 						mtlTempIndBuffOfst += sizeof(MTLDrawPatchIndirectArguments);
-                        // Mark pipeline, resources, and tess control push constants as dirty
+                        // Mark pipeline, resources, and vertex push constants as dirty
                         // so I apply them during the next stage.
                         cmdEncoder->_graphicsPipelineState.beginMetalRenderPass();
                         cmdEncoder->_graphicsResourcesState.beginMetalRenderPass();
-                        cmdEncoder->getPushConstants(VK_SHADER_STAGE_TESSELLATION_CONTROL_BIT)->beginMetalRenderPass();
+                        cmdEncoder->getPushConstants(VK_SHADER_STAGE_VERTEX_BIT)->beginMetalRenderPass();
                     } else {
                         [cmdEncoder->_mtlRenderEncoder drawIndexedPrimitives: cmdEncoder->_mtlPrimitiveType
                                                                    indexType: (MTLIndexType)ibb.mtlIndexType
diff --git a/MoltenVK/MoltenVK/Commands/MVKCmdPipeline.h b/MoltenVK/MoltenVK/Commands/MVKCmdPipeline.h
index b2985ee..98616cb 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCmdPipeline.h
+++ b/MoltenVK/MoltenVK/Commands/MVKCmdPipeline.h
@@ -139,11 +139,13 @@
 
 	void encode(MVKCommandEncoder* cmdEncoder) override;
 
+	~MVKCmdBindDescriptorSetsStatic() override;
+
 protected:
 	MVKCommandTypePool<MVKCommand>* getTypePool(MVKCommandPool* cmdPool) override;
 
 	MVKSmallVector<MVKDescriptorSet*, N> _descriptorSets;
-	MVKPipelineLayout* _pipelineLayout;
+	MVKPipelineLayout* _pipelineLayout = nullptr;
 	VkPipelineBindPoint _pipelineBindPoint;
 	uint32_t _firstSet;
 };
@@ -211,7 +213,6 @@
 	MVKCommandTypePool<MVKCommand>* getTypePool(MVKCommandPool* cmdPool) override;
 
 	MVKSmallVector<char, N> _pushConstants;
-	MVKPipelineLayout* _pipelineLayout;
 	VkShaderStageFlags _stageFlags;
 	uint32_t _offset;
 };
@@ -245,7 +246,7 @@
 	void clearDescriptorWrites();
 
 	MVKSmallVector<VkWriteDescriptorSet, 1> _descriptorWrites;
-	MVKPipelineLayout* _pipelineLayout;
+	MVKPipelineLayout* _pipelineLayout = nullptr;
 	VkPipelineBindPoint _pipelineBindPoint;
 	uint32_t _set;
 };
@@ -272,7 +273,7 @@
 	MVKCommandTypePool<MVKCommand>* getTypePool(MVKCommandPool* cmdPool) override;
 
 	MVKDescriptorUpdateTemplate* _descUpdateTemplate;
-	MVKPipelineLayout* _pipelineLayout;
+	MVKPipelineLayout* _pipelineLayout = nullptr;
 	void* _pData = nullptr;
 	uint32_t _set;
 };
diff --git a/MoltenVK/MoltenVK/Commands/MVKCmdPipeline.mm b/MoltenVK/MoltenVK/Commands/MVKCmdPipeline.mm
index 4e12de5..00029a6 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCmdPipeline.mm
+++ b/MoltenVK/MoltenVK/Commands/MVKCmdPipeline.mm
@@ -193,10 +193,14 @@
 													   uint32_t firstSet,
 													   uint32_t setCount,
 													   const VkDescriptorSet* pDescriptorSets) {
+	if (_pipelineLayout) { _pipelineLayout->release(); }
+
 	_pipelineBindPoint = pipelineBindPoint;
 	_pipelineLayout = (MVKPipelineLayout*)layout;
 	_firstSet = firstSet;
 
+	_pipelineLayout->retain();
+
 	// Add the descriptor sets
 	_descriptorSets.clear();	// Clear for reuse
 	_descriptorSets.reserve(setCount);
@@ -212,6 +216,11 @@
 	_pipelineLayout->bindDescriptorSets(cmdEncoder, _descriptorSets.contents(), _firstSet, MVKArrayRef<uint32_t>());
 }
 
+template <size_t N>
+MVKCmdBindDescriptorSetsStatic<N>::~MVKCmdBindDescriptorSetsStatic() {
+	if (_pipelineLayout) { _pipelineLayout->release(); }
+}
+
 template class MVKCmdBindDescriptorSetsStatic<1>;
 template class MVKCmdBindDescriptorSetsStatic<4>;
 template class MVKCmdBindDescriptorSetsStatic<8>;
@@ -262,7 +271,6 @@
 											uint32_t offset,
 											uint32_t size,
 											const void* pValues) {
-	_pipelineLayout = (MVKPipelineLayout*)layout;
 	_stageFlags = stageFlags;
 	_offset = offset;
 
@@ -302,10 +310,14 @@
 											 uint32_t set,
 											 uint32_t descriptorWriteCount,
 											 const VkWriteDescriptorSet* pDescriptorWrites) {
+	if (_pipelineLayout) { _pipelineLayout->release(); }
+
 	_pipelineBindPoint = pipelineBindPoint;
 	_pipelineLayout = (MVKPipelineLayout*)layout;
 	_set = set;
 
+	_pipelineLayout->retain();
+
 	// Add the descriptor writes
 	MVKDevice* mvkDvc = cmdBuff->getDevice();
 	clearDescriptorWrites();	// Clear for reuse
@@ -360,6 +372,7 @@
 
 MVKCmdPushDescriptorSet::~MVKCmdPushDescriptorSet() {
 	clearDescriptorWrites();
+	if (_pipelineLayout) { _pipelineLayout->release(); }
 }
 
 void MVKCmdPushDescriptorSet::clearDescriptorWrites() {
@@ -393,9 +406,14 @@
 														 VkPipelineLayout layout,
 														 uint32_t set,
 														 const void* pData) {
+	if (_pipelineLayout) { _pipelineLayout->release(); }
+
 	_descUpdateTemplate = (MVKDescriptorUpdateTemplate*)descUpdateTemplate;
 	_pipelineLayout = (MVKPipelineLayout*)layout;
 	_set = set;
+
+	_pipelineLayout->retain();
+
 	if (_pData) delete[] (char*)_pData;
 	// Work out how big the memory block in pData is.
 	const VkDescriptorUpdateTemplateEntryKHR* pEntry =
@@ -443,6 +461,7 @@
 }
 
 MVKCmdPushDescriptorSetWithTemplate::~MVKCmdPushDescriptorSetWithTemplate() {
+	if (_pipelineLayout) { _pipelineLayout->release(); }
 	if (_pData) delete[] (char*)_pData;
 }
 
diff --git a/MoltenVK/MoltenVK/Commands/MVKCmdTransfer.h b/MoltenVK/MoltenVK/Commands/MVKCmdTransfer.h
index 4bc8b11..849d51c 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCmdTransfer.h
+++ b/MoltenVK/MoltenVK/Commands/MVKCmdTransfer.h
@@ -125,9 +125,8 @@
 
 /** Describes Metal texture resolve parameters. */
 typedef struct {
-    VkImageCopy* copyRegion;
-    uint32_t level;
-    uint32_t slice;
+    VkImageSubresource srcSubresource;
+    VkImageSubresource dstSubresource;
 } MVKMetalResolveSlice;
 
 /**
diff --git a/MoltenVK/MoltenVK/Commands/MVKCmdTransfer.mm b/MoltenVK/MoltenVK/Commands/MVKCmdTransfer.mm
index f494748..15e4b1f 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCmdTransfer.mm
+++ b/MoltenVK/MoltenVK/Commands/MVKCmdTransfer.mm
@@ -77,11 +77,6 @@
 		_vkImageCopies.push_back(vkIR);
 	}
     
-    // Validate
-    if ((_srcImage->getMTLTextureType() == MTLTextureType3D) != (_dstImage->getMTLTextureType() == MTLTextureType3D)) {
-        return cmdBuff->reportError(VK_ERROR_FEATURE_NOT_PRESENT, "vkCmdCopyImage(): Metal does not support copying to or from slices of a 3D texture.");
-    }
-
 	return VK_SUCCESS;
 }
 
@@ -160,25 +155,62 @@
             // If copies can be performed using direct texture-texture copying, do so
             uint32_t srcLevel = vkIC.srcSubresource.mipLevel;
             MTLOrigin srcOrigin = mvkMTLOriginFromVkOffset3D(vkIC.srcOffset);
-            MTLSize srcSize = mvkClampMTLSize(mvkMTLSizeFromVkExtent3D(vkIC.extent),
-                                              srcOrigin,
-                                              mvkMTLSizeFromVkExtent3D(_srcImage->getExtent3D(srcPlaneIndex, srcLevel)));
+            MTLSize srcSize;
+            uint32_t layCnt;
+            if ((_srcImage->getMTLTextureType() == MTLTextureType3D) != (_dstImage->getMTLTextureType() == MTLTextureType3D)) {
+                // In the case, the number of layers to copy is in extent.depth. Use that value,
+                // then clamp the depth so we don't try to copy more than Metal will allow.
+                layCnt = vkIC.extent.depth;
+                srcSize = mvkClampMTLSize(mvkMTLSizeFromVkExtent3D(vkIC.extent),
+                                          srcOrigin,
+                                          mvkMTLSizeFromVkExtent3D(_srcImage->getExtent3D(srcPlaneIndex, srcLevel)));
+                srcSize.depth = 1;
+            } else {
+                layCnt = vkIC.srcSubresource.layerCount;
+                srcSize = mvkClampMTLSize(mvkMTLSizeFromVkExtent3D(vkIC.extent),
+                                          srcOrigin,
+                                          mvkMTLSizeFromVkExtent3D(_srcImage->getExtent3D(srcPlaneIndex, srcLevel)));
+            }
             uint32_t dstLevel = vkIC.dstSubresource.mipLevel;
             MTLOrigin dstOrigin = mvkMTLOriginFromVkOffset3D(vkIC.dstOffset);
             uint32_t srcBaseLayer = vkIC.srcSubresource.baseArrayLayer;
             uint32_t dstBaseLayer = vkIC.dstSubresource.baseArrayLayer;
-            uint32_t layCnt = vkIC.srcSubresource.layerCount;
-
+            
             for (uint32_t layIdx = 0; layIdx < layCnt; layIdx++) {
-                [mtlBlitEnc copyFromTexture: srcMTLTex
-                                sourceSlice: srcBaseLayer + layIdx
-                                sourceLevel: srcLevel
-                               sourceOrigin: srcOrigin
-                                 sourceSize: srcSize
-                                  toTexture: dstMTLTex
-                           destinationSlice: dstBaseLayer + layIdx
-                           destinationLevel: dstLevel
-                          destinationOrigin: dstOrigin];
+                // We can copy between a 3D and a 2D image easily. Just copy between
+                // one slice of the 2D image and one plane of the 3D image at a time.
+                if ((_srcImage->getMTLTextureType() == MTLTextureType3D) == (_dstImage->getMTLTextureType() == MTLTextureType3D)) {
+                    [mtlBlitEnc copyFromTexture: srcMTLTex
+                                    sourceSlice: srcBaseLayer + layIdx
+                                    sourceLevel: srcLevel
+                                   sourceOrigin: srcOrigin
+                                     sourceSize: srcSize
+                                      toTexture: dstMTLTex
+                               destinationSlice: dstBaseLayer + layIdx
+                               destinationLevel: dstLevel
+                              destinationOrigin: dstOrigin];
+                } else if (_srcImage->getMTLTextureType() == MTLTextureType3D) {
+                    [mtlBlitEnc copyFromTexture: srcMTLTex
+                                    sourceSlice: srcBaseLayer
+                                    sourceLevel: srcLevel
+                                   sourceOrigin: MTLOriginMake(srcOrigin.x, srcOrigin.y, srcOrigin.z + layIdx)
+                                     sourceSize: srcSize
+                                      toTexture: dstMTLTex
+                               destinationSlice: dstBaseLayer + layIdx
+                               destinationLevel: dstLevel
+                              destinationOrigin: dstOrigin];
+                } else {
+                    assert(_dstImage->getMTLTextureType() == MTLTextureType3D);
+                    [mtlBlitEnc copyFromTexture: srcMTLTex
+                                    sourceSlice: srcBaseLayer + layIdx
+                                    sourceLevel: srcLevel
+                                   sourceOrigin: srcOrigin
+                                     sourceSize: srcSize
+                                      toTexture: dstMTLTex
+                               destinationSlice: dstBaseLayer
+                               destinationLevel: dstLevel
+                              destinationOrigin: MTLOriginMake(dstOrigin.x, dstOrigin.y, dstOrigin.z + layIdx)];
+                }
             }
         }
     }
@@ -396,6 +428,18 @@
             MVKRPSKeyBlitImg blitKey;
             blitKey.srcMTLPixelFormat = _srcImage->getMTLPixelFormat(srcPlaneIndex);
             blitKey.srcMTLTextureType = _srcImage->getMTLTextureType();
+            if (blitKey.srcMTLTextureType == MTLTextureTypeCube || blitKey.srcMTLTextureType == MTLTextureTypeCubeArray) {
+                // In this case, I'll use a temp 2D array view. That way, I don't have to
+                // deal with mapping the blit coordinates to a cube direction vector.
+                blitKey.srcMTLTextureType = MTLTextureType2DArray;
+                srcMTLTex = [srcMTLTex newTextureViewWithPixelFormat: (MTLPixelFormat)blitKey.srcMTLPixelFormat
+                                                         textureType: MTLTextureType2DArray
+                                                              levels: NSMakeRange(0, srcMTLTex.mipmapLevelCount)
+                                                              slices: NSMakeRange(0, srcMTLTex.arrayLength)];
+                [cmdEncoder->_mtlCmdBuffer addCompletedHandler: ^(id<MTLCommandBuffer>) {
+                    [srcMTLTex release];
+                }];
+            }
             blitKey.dstMTLPixelFormat = _dstImage->getMTLPixelFormat(dstPlaneIndex);
             blitKey.srcFilter = mvkMTLSamplerMinMagFilterFromVkFilter(_filter);
             blitKey.dstSampleCount = mvkSampleCountFromVkSampleCountFlagBits(_dstImage->getSampleCount());
@@ -406,12 +450,31 @@
             mtlColorAttDesc.level = mvkIBR.region.dstSubresource.mipLevel;
 
             uint32_t layCnt = mvkIBR.region.srcSubresource.layerCount;
+            if (_dstImage->getMTLTextureType() == MTLTextureType3D) {
+                layCnt = mvkAbsDiff(mvkIBR.region.dstOffsets[1].z, mvkIBR.region.dstOffsets[0].z);
+            }
             for (uint32_t layIdx = 0; layIdx < layCnt; layIdx++) {
                 // Update the render pass descriptor for the texture level and slice, and create a render encoder.
-                mtlColorAttDesc.slice = mvkIBR.region.dstSubresource.baseArrayLayer + layIdx;
+                if (_dstImage->getMTLTextureType() == MTLTextureType3D) {
+                    mtlColorAttDesc.depthPlane = mvkIBR.region.dstOffsets[0].z + (mvkIBR.region.dstOffsets[1].z > mvkIBR.region.dstOffsets[0].z ? layIdx : -(layIdx + 1));
+                } else {
+                    mtlColorAttDesc.slice = mvkIBR.region.dstSubresource.baseArrayLayer + layIdx;
+                }
                 id<MTLRenderCommandEncoder> mtlRendEnc = [cmdEncoder->_mtlCmdBuffer renderCommandEncoderWithDescriptor: mtlRPD];
                 setLabelIfNotNil(mtlRendEnc, mvkMTLRenderCommandEncoderLabel(commandUse));
 
+                if (blitKey.srcMTLTextureType == MTLTextureType3D) {
+                    // In this case, I need to interpolate along the third dimension manually.
+                    VkExtent3D srcExtent = _srcImage->getExtent3D(srcPlaneIndex, mvkIBR.region.dstSubresource.mipLevel);
+                    VkOffset3D so0 = mvkIBR.region.srcOffsets[0], so1 = mvkIBR.region.srcOffsets[1];
+                    VkOffset3D do0 = mvkIBR.region.dstOffsets[0], do1 = mvkIBR.region.dstOffsets[1];
+                    CGFloat startZ = (CGFloat)so0.z / (CGFloat)srcExtent.depth;
+                    CGFloat endZ = (CGFloat)so1.z / (CGFloat)srcExtent.depth;
+                    CGFloat zIncr = (endZ - startZ) / mvkAbsDiff(do1.z, do0.z);
+                    for (uint32_t i = 0; i < kMVKBlitVertexCount; ++i) {
+                        mvkIBR.vertices[i].texCoord.z = startZ + layIdx * zIncr;
+                    }
+                }
                 [mtlRendEnc pushDebugGroup: @"vkCmdBlitImage"];
                 [mtlRendEnc setRenderPipelineState: mtlRPS];
                 cmdEncoder->setVertexBytes(mtlRendEnc, mvkIBR.vertices, sizeof(mvkIBR.vertices), vtxBuffIdx);
@@ -490,68 +553,75 @@
         uint8_t srcPlaneIndex = MVKImage::getPlaneFromVkImageAspectFlags(vkIR.srcSubresource.aspectMask);
         uint8_t dstPlaneIndex = MVKImage::getPlaneFromVkImageAspectFlags(vkIR.dstSubresource.aspectMask);
 
-		uint32_t mipLvl = vkIR.dstSubresource.mipLevel;
-		VkExtent3D srcImgExt = _srcImage->getExtent3D(srcPlaneIndex, mipLvl);
-		VkExtent3D dstImgExt = _dstImage->getExtent3D(dstPlaneIndex, mipLvl);
+		VkExtent3D srcImgExt = _srcImage->getExtent3D(srcPlaneIndex, vkIR.srcSubresource.mipLevel);
+		VkExtent3D dstImgExt = _dstImage->getExtent3D(dstPlaneIndex, vkIR.dstSubresource.mipLevel);
 
-		// If the region does not cover the entire content of the source level, expand the
-		// destination content in the region to the temporary image. The purpose of this
+		// If the region does not cover the entire content of the destination level, expand
+		// the destination content in the region to the temporary image. The purpose of this
 		// expansion is to render the existing content of the destination image to the
 		// temporary transfer multisample image, so that regions of that temporary transfer
 		// image can then be overwritten with content from the source image, prior to
-		// resolving it back to the destination image. The source of this temporary content
-		// move is the full extent of the DESTINATION image of the resolve command, and the
-		// destination of this temporary content move is the full extent of the SOURCE image.
-		if ( !mvkVkExtent3DsAreEqual(srcImgExt, vkIR.extent) ) {
+		// resolving it back to the destination image.
+		if ( !mvkVkExtent3DsAreEqual(dstImgExt, vkIR.extent) ) {
 			VkImageBlit& expRgn = expansionRegions[expCnt++];
 			expRgn.srcSubresource = vkIR.dstSubresource;
 			expRgn.srcOffsets[0] = { 0, 0, 0 };
 			expRgn.srcOffsets[1] = { int32_t(dstImgExt.width), int32_t(dstImgExt.height), int32_t(dstImgExt.depth) };
 			expRgn.dstSubresource = vkIR.dstSubresource;
 			expRgn.dstOffsets[0] = { 0, 0, 0 };
-			expRgn.dstOffsets[1] = { int32_t(srcImgExt.width), int32_t(srcImgExt.height), int32_t(srcImgExt.depth) };
+			expRgn.dstOffsets[1] = { int32_t(dstImgExt.width), int32_t(dstImgExt.height), int32_t(dstImgExt.depth) };
 		}
 
 		// Copy the region from the source image to the temporary multisample image,
 		// prior to the temporary image being resolved back to the destination image.
 		// The source of this copy stage is the source image, and the destination of
 		// this copy stage is the temporary transfer image.
-		VkImageCopy& cpyRgn = copyRegions[copyCnt++];
-		cpyRgn.srcSubresource = vkIR.srcSubresource;
-		cpyRgn.srcOffset = vkIR.srcOffset;
-		cpyRgn.dstSubresource = vkIR.srcSubresource;
-		cpyRgn.dstOffset = vkIR.srcOffset;
-		cpyRgn.extent = vkIR.extent;
+		bool needXfrImage = !mvkVkExtent3DsAreEqual(srcImgExt, vkIR.extent) || !mvkVkExtent3DsAreEqual(dstImgExt, vkIR.extent);
+		if ( needXfrImage ) {
+			VkImageCopy& cpyRgn = copyRegions[copyCnt++];
+			cpyRgn.srcSubresource = vkIR.srcSubresource;
+			cpyRgn.srcOffset = vkIR.srcOffset;
+			cpyRgn.dstSubresource = vkIR.dstSubresource;
+			cpyRgn.dstOffset = vkIR.dstOffset;
+			cpyRgn.extent = vkIR.extent;
+		}
 
 		// Adds a resolve slice struct for each destination layer in the resolve region.
-		uint32_t baseLayer = vkIR.dstSubresource.baseArrayLayer;
+		// Note that the source subresource for this is that of the SOURCE image if we're doing a
+		// direct resolve, but that of the DESTINATION if we need a temporary transfer image.
 		uint32_t layCnt = vkIR.dstSubresource.layerCount;
 		for (uint32_t layIdx = 0; layIdx < layCnt; layIdx++) {
 			MVKMetalResolveSlice& rslvSlice = mtlResolveSlices[sliceCnt++];
-            rslvSlice.copyRegion = &cpyRgn;
-			rslvSlice.level = vkIR.dstSubresource.mipLevel;
-			rslvSlice.slice = baseLayer + layIdx;
+			rslvSlice.dstSubresource.aspectMask = vkIR.dstSubresource.aspectMask;
+			rslvSlice.dstSubresource.mipLevel = vkIR.dstSubresource.mipLevel;
+			rslvSlice.dstSubresource.arrayLayer = vkIR.dstSubresource.baseArrayLayer + layIdx;
+			rslvSlice.srcSubresource.aspectMask = needXfrImage ? vkIR.dstSubresource.aspectMask : vkIR.srcSubresource.aspectMask;
+			rslvSlice.srcSubresource.mipLevel = needXfrImage ? vkIR.dstSubresource.mipLevel : vkIR.srcSubresource.mipLevel;
+			rslvSlice.srcSubresource.arrayLayer = needXfrImage ? vkIR.dstSubresource.baseArrayLayer : vkIR.srcSubresource.baseArrayLayer;
+			rslvSlice.srcSubresource.arrayLayer += layIdx;
 		}
 	}
 
     // Expansion and copying is not required. Each mip level of the source image
     // is being resolved entirely. Resolve directly from the source image.
     MVKImage* xfrImage = _srcImage;
-	if (expCnt) {
-		// Expansion and copying is required. Acquire a temporary transfer image, expand
-		// the destination image into it, copy from the source image to the temporary image,
-		// and then resolve from the temporary image to the destination image.
+	if (copyCnt) {
+		// Expansion and/or copying is required. Acquire a temporary transfer image, expand
+		// the destination image into it if necessary, copy from the source image to the
+		// temporary image, and then resolve from the temporary image to the destination image.
 		MVKImageDescriptorData xferImageData;
 		_dstImage->getTransferDescriptorData(xferImageData);
 		xferImageData.samples = _srcImage->getSampleCount();
 		xfrImage = cmdEncoder->getCommandEncodingPool()->getTransferMVKImage(xferImageData);
 
-		// Expand the current content of the destination image to the temporary transfer image.
-		MVKCmdBlitImage<N> expCmd;
-		expCmd.setContent(cmdEncoder->_cmdBuffer,
-						  (VkImage)_dstImage, _dstLayout, (VkImage)xfrImage, _dstLayout,
-						  expCnt, expansionRegions, VK_FILTER_LINEAR);
-		expCmd.encode(cmdEncoder, kMVKCommandUseResolveExpandImage);
+		if (expCnt) {
+			// Expand the current content of the destination image to the temporary transfer image.
+			MVKCmdBlitImage<N> expCmd;
+			expCmd.setContent(cmdEncoder->_cmdBuffer,
+							  (VkImage)_dstImage, _dstLayout, (VkImage)xfrImage, _dstLayout,
+							  expCnt, expansionRegions, VK_FILTER_LINEAR);
+			expCmd.encode(cmdEncoder, kMVKCommandUseResolveExpandImage);
+		}
 
 		// Copy the resolve regions of the source image to the temporary transfer image.
 		MVKCmdCopyImage<N> copyCmd;
@@ -573,15 +643,15 @@
 	// the texture level and slice and create a render encoder.
 	for (uint32_t sIdx = 0; sIdx < sliceCnt; sIdx++) {
 		MVKMetalResolveSlice& rslvSlice = mtlResolveSlices[sIdx];
-        uint8_t srcPlaneIndex = MVKImage::getPlaneFromVkImageAspectFlags(rslvSlice.copyRegion->srcSubresource.aspectMask);
-        uint8_t dstPlaneIndex = MVKImage::getPlaneFromVkImageAspectFlags(rslvSlice.copyRegion->dstSubresource.aspectMask);
+        uint8_t srcPlaneIndex = MVKImage::getPlaneFromVkImageAspectFlags(rslvSlice.srcSubresource.aspectMask);
+        uint8_t dstPlaneIndex = MVKImage::getPlaneFromVkImageAspectFlags(rslvSlice.dstSubresource.aspectMask);
 
         mtlColorAttDesc.texture = xfrImage->getMTLTexture(srcPlaneIndex);
         mtlColorAttDesc.resolveTexture = _dstImage->getMTLTexture(dstPlaneIndex);
-		mtlColorAttDesc.level = rslvSlice.level;
-		mtlColorAttDesc.slice = rslvSlice.slice;
-		mtlColorAttDesc.resolveLevel = rslvSlice.level;
-		mtlColorAttDesc.resolveSlice = rslvSlice.slice;
+		mtlColorAttDesc.level = rslvSlice.srcSubresource.mipLevel;
+		mtlColorAttDesc.slice = rslvSlice.srcSubresource.arrayLayer;
+		mtlColorAttDesc.resolveLevel = rslvSlice.dstSubresource.mipLevel;
+		mtlColorAttDesc.resolveSlice = rslvSlice.dstSubresource.arrayLayer;
 		id<MTLRenderCommandEncoder> mtlRendEnc = [cmdEncoder->_mtlCmdBuffer renderCommandEncoderWithDescriptor: mtlRPD];
 		setLabelIfNotNil(mtlRendEnc, mvkMTLRenderCommandEncoderLabel(kMVKCommandUseResolveImage));
 
@@ -1159,8 +1229,8 @@
 
         // Validate
         MVKMTLFmtCaps mtlFmtCaps = cmdBuff->getPixelFormats()->getCapabilities(_image->getMTLPixelFormat(planeIndex));
-        if ((isDS && !mvkAreAllFlagsEnabled(mtlFmtCaps, kMVKMTLFmtCapsDSAtt)) ||
-            ( !isDS && !mvkAreAllFlagsEnabled(mtlFmtCaps, kMVKMTLFmtCapsColorAtt))) {
+		uint32_t reqCap = isDS ? kMVKMTLFmtCapsDSAtt : (_image->getIsLinear() ? kMVKMTLFmtCapsWrite : kMVKMTLFmtCapsColorAtt);
+        if (!mvkAreAllFlagsEnabled(mtlFmtCaps, reqCap)) {
             return cmdBuff->reportError(VK_ERROR_FEATURE_NOT_PRESENT, "vkCmdClear%sImage(): Format %s cannot be cleared on this device.", (isDS ? "DepthStencil" : "Color"), cmdBuff->getPixelFormats()->getName(_image->getVkFormat()));
         }
         
@@ -1186,9 +1256,35 @@
 
 	MVKPixelFormats* pixFmts = cmdEncoder->getPixelFormats();
 	for (auto& srRange : _subresourceRanges) {
-        id<MTLTexture> imgMTLTex = _image->getMTLTexture(MVKImage::getPlaneFromVkImageAspectFlags(srRange.aspectMask));
+		uint8_t planeIndex = MVKImage::getPlaneFromVkImageAspectFlags(srRange.aspectMask);
+        id<MTLTexture> imgMTLTex = _image->getMTLTexture(planeIndex);
         if ( !imgMTLTex ) { continue; }
 
+#if MVK_MACOS
+        if ( _image->getIsLinear() ) {
+            // These images cannot be rendered. Instead, use a compute shader.
+            // Luckily for us, linear images only have one mip and one array layer under Metal.
+            assert( !isDS );
+            id<MTLComputePipelineState> mtlClearState = cmdEncoder->getCommandEncodingPool()->getCmdClearColorImageMTLComputePipelineState(pixFmts->getFormatType(_image->getVkFormat()));
+            id<MTLComputeCommandEncoder> mtlComputeEnc = cmdEncoder->getMTLComputeEncoder(kMVKCommandUseClearColorImage);
+            [mtlComputeEnc pushDebugGroup: @"vkCmdClearColorImage"];
+            [mtlComputeEnc setComputePipelineState: mtlClearState];
+            [mtlComputeEnc setTexture: imgMTLTex atIndex: 0];
+            cmdEncoder->setComputeBytes(mtlComputeEnc, &_clearValue, sizeof(_clearValue), 0);
+            MTLSize gridSize = mvkMTLSizeFromVkExtent3D(_image->getExtent3D());
+            MTLSize tgSize = MTLSizeMake(mtlClearState.threadExecutionWidth, 1, 1);
+            if (cmdEncoder->getDevice()->_pMetalFeatures->nonUniformThreadgroups) {
+                [mtlComputeEnc dispatchThreads: gridSize threadsPerThreadgroup: tgSize];
+            } else {
+                MTLSize tgCount = MTLSizeMake(gridSize.width / tgSize.width, gridSize.height, gridSize.depth);
+                if (gridSize.width % tgSize.width) { tgCount.width += 1; }
+                [mtlComputeEnc dispatchThreadgroups: tgCount threadsPerThreadgroup: tgSize];
+            }
+            [mtlComputeEnc popDebugGroup];
+            continue;
+        }
+#endif
+
 		MTLRenderPassDescriptor* mtlRPDesc = [MTLRenderPassDescriptor renderPassDescriptor];
 		MTLRenderPassColorAttachmentDescriptor* mtlRPCADesc = nil;
 		MTLRenderPassDepthAttachmentDescriptor* mtlRPDADesc = nil;
@@ -1230,7 +1326,8 @@
                               : (mipLvlStart + mipLvlCnt));
 
         // Extract the cube or array layers (slices) that are to be updated
-        uint32_t layerStart = srRange.baseArrayLayer;
+		bool is3D = _image->getMTLTextureType() == MTLTextureType3D;
+        uint32_t layerStart = is3D ? 0 : srRange.baseArrayLayer;
         uint32_t layerCnt = srRange.layerCount;
         uint32_t layerEnd = (layerCnt == VK_REMAINING_ARRAY_LAYERS
                              ? _image->getLayerCount()
@@ -1242,10 +1339,22 @@
 			mtlRPDADesc.level = mipLvl;
 			mtlRPSADesc.level = mipLvl;
 
+			// If a 3D image, we need to get the depth for each level.
+			if (is3D) {
+				layerCnt = _image->getExtent3D(planeIndex, mipLvl).depth;
+				layerEnd = layerStart + layerCnt;
+			}
+
             for (uint32_t layer = layerStart; layer < layerEnd; layer++) {
-                mtlRPCADesc.slice = layer;
-				mtlRPDADesc.slice = layer;
-				mtlRPSADesc.slice = layer;
+                if (is3D) {
+                    mtlRPCADesc.depthPlane = layer;
+                    mtlRPDADesc.depthPlane = layer;
+                    mtlRPSADesc.depthPlane = layer;
+                } else {
+                    mtlRPCADesc.slice = layer;
+                    mtlRPDADesc.slice = layer;
+                    mtlRPSADesc.slice = layer;
+                }
 
                 id<MTLRenderCommandEncoder> mtlRendEnc = [cmdEncoder->_mtlCmdBuffer renderCommandEncoderWithDescriptor: mtlRPDesc];
 				setLabelIfNotNil(mtlRendEnc, mtlRendEncName);
diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandBuffer.mm b/MoltenVK/MoltenVK/Commands/MVKCommandBuffer.mm
index 24b65a4..0889577 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCommandBuffer.mm
+++ b/MoltenVK/MoltenVK/Commands/MVKCommandBuffer.mm
@@ -752,6 +752,7 @@
         case kMVKCommandUseCopyBufferToImage:   return @"vkCmdCopyBufferToImage ComputeEncoder";
         case kMVKCommandUseCopyImageToBuffer:   return @"vkCmdCopyImageToBuffer ComputeEncoder";
         case kMVKCommandUseFillBuffer:          return @"vkCmdFillBuffer ComputeEncoder";
+        case kMVKCommandUseClearColorImage:     return @"vkCmdClearColorImage ComputeEncoder";
         case kMVKCommandUseTessellationVertexTessCtl: return @"vkCmdDraw (vertex and tess control stages) ComputeEncoder";
         case kMVKCommandUseMultiviewInstanceCountAdjust: return @"vkCmdDraw (multiview instance count adjustment) ComputeEncoder";
         case kMVKCommandUseCopyQueryPoolResults:return @"vkCmdCopyQueryPoolResults ComputeEncoder";
diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandEncodingPool.h b/MoltenVK/MoltenVK/Commands/MVKCommandEncodingPool.h
index 52e4704..eac9f3f 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCommandEncodingPool.h
+++ b/MoltenVK/MoltenVK/Commands/MVKCommandEncodingPool.h
@@ -109,6 +109,11 @@
 	/** Returns a MTLComputePipelineState for filling a buffer. */
 	id<MTLComputePipelineState> getCmdFillBufferMTLComputePipelineState();
 
+#if MVK_MACOS
+	/** Returns a MTLComputePipelineState for clearing an image. Currently only used for 2D linear images on Mac. */
+	id<MTLComputePipelineState> getCmdClearColorImageMTLComputePipelineState(MVKFormatType type);
+#endif
+
 	/** Returns a MTLComputePipelineState for decompressing a buffer into a 3D image. */
 	id<MTLComputePipelineState> getCmdCopyBufferToImage3DDecompressMTLComputePipelineState(bool needsTempBuff);
 
@@ -151,6 +156,9 @@
     id<MTLDepthStencilState> _cmdClearDefaultDepthStencilState = nil;
     id<MTLComputePipelineState> _mtlCopyBufferBytesComputePipelineState = nil;
 	id<MTLComputePipelineState> _mtlFillBufferComputePipelineState = nil;
+#if MVK_MACOS
+	id<MTLComputePipelineState> _mtlClearColorImageComputePipelineState[3] = {nil, nil, nil};
+#endif
 	id<MTLComputePipelineState> _mtlCopyBufferToImage3DDecompressComputePipelineState[2] = {nil, nil};
 	id<MTLComputePipelineState> _mtlDrawIndirectMultiviewConvertBuffersComputePipelineState[2] = {nil, nil};
 	id<MTLComputePipelineState> _mtlDrawIndirectTessConvertBuffersComputePipelineState[2] = {nil, nil};
diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandEncodingPool.mm b/MoltenVK/MoltenVK/Commands/MVKCommandEncodingPool.mm
index da0e661..b1a5a26 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCommandEncodingPool.mm
+++ b/MoltenVK/MoltenVK/Commands/MVKCommandEncodingPool.mm
@@ -102,6 +102,30 @@
 	MVK_ENC_REZ_ACCESS(_mtlFillBufferComputePipelineState, newCmdFillBufferMTLComputePipelineState(_commandPool));
 }
 
+#if MVK_MACOS
+static inline uint32_t getClearStateIndex(MVKFormatType type) {
+	switch (type) {
+		case kMVKFormatColorHalf:
+		case kMVKFormatColorFloat:
+			return 0;
+		case kMVKFormatColorInt8:
+		case kMVKFormatColorInt16:
+		case kMVKFormatColorInt32:
+			return 1;
+		case kMVKFormatColorUInt8:
+		case kMVKFormatColorUInt16:
+		case kMVKFormatColorUInt32:
+			return 2;
+		default:
+			return 0;
+	}
+}
+
+id<MTLComputePipelineState> MVKCommandEncodingPool::getCmdClearColorImageMTLComputePipelineState(MVKFormatType type) {
+	MVK_ENC_REZ_ACCESS(_mtlClearColorImageComputePipelineState[getClearStateIndex(type)], newCmdClearColorImageMTLComputePipelineState(type, _commandPool));
+}
+#endif
+
 id<MTLComputePipelineState> MVKCommandEncodingPool::getCmdCopyBufferToImage3DDecompressMTLComputePipelineState(bool needsTempBuff) {
 	MVK_ENC_REZ_ACCESS(_mtlCopyBufferToImage3DDecompressComputePipelineState[needsTempBuff ? 1 : 0], newCmdCopyBufferToImage3DDecompressMTLComputePipelineState(needsTempBuff, _commandPool));
 }
@@ -178,6 +202,15 @@
     [_mtlFillBufferComputePipelineState release];
     _mtlFillBufferComputePipelineState = nil;
 
+#if MVK_MACOS
+    [_mtlClearColorImageComputePipelineState[0] release];
+    [_mtlClearColorImageComputePipelineState[1] release];
+    [_mtlClearColorImageComputePipelineState[2] release];
+    _mtlClearColorImageComputePipelineState[0] = nil;
+    _mtlClearColorImageComputePipelineState[1] = nil;
+    _mtlClearColorImageComputePipelineState[2] = nil;
+#endif
+
     [_mtlCopyBufferToImage3DDecompressComputePipelineState[0] release];
     [_mtlCopyBufferToImage3DDecompressComputePipelineState[1] release];
     _mtlCopyBufferToImage3DDecompressComputePipelineState[0] = nil;
diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandPipelineStateFactoryShaderSource.h b/MoltenVK/MoltenVK/Commands/MVKCommandPipelineStateFactoryShaderSource.h
index 124f6d9..d562183 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCommandPipelineStateFactoryShaderSource.h
+++ b/MoltenVK/MoltenVK/Commands/MVKCommandPipelineStateFactoryShaderSource.h
@@ -29,12 +29,12 @@
                                                                                                                 \n\
 typedef struct {                                                                                                \n\
     float2 a_position [[attribute(0)]];                                                                         \n\
-    float2 a_texCoord [[attribute(1)]];                                                                         \n\
+    float3 a_texCoord [[attribute(1)]];                                                                         \n\
 } AttributesPosTex;                                                                                             \n\
                                                                                                                 \n\
 typedef struct {                                                                                                \n\
     float4 v_position [[position]];                                                                             \n\
-    float2 v_texCoord;                                                                                          \n\
+    float3 v_texCoord;                                                                                          \n\
 } VaryingsPosTex;                                                                                               \n\
                                                                                                                 \n\
 typedef size_t VkDeviceSize;                                                                                    \n\
@@ -91,13 +91,31 @@
     for (size_t i = 0; i < info.size; i++) {                                                                    \n\
         dst[i + info.dstOffset] = src[i + info.srcOffset];                                                      \n\
     }                                                                                                           \n\
-};                                                                                                              \n\
+}                                                                                                               \n\
                                                                                                                 \n\
 kernel void cmdFillBuffer(device uint32_t* dst [[ buffer(0) ]],                                                 \n\
                           constant uint32_t& fillValue [[ buffer(1) ]],                                         \n\
                           uint pos [[thread_position_in_grid]]) {                                               \n\
     dst[pos] = fillValue;                                                                                       \n\
-};                                                                                                              \n\
+}                                                                                                               \n\
+                                                                                                                \n\
+kernel void cmdClearColorImage2DFloat(texture2d<float, access::write> dst [[ texture(0) ]],                     \n\
+                                      constant float4& clearValue [[ buffer(0) ]],                              \n\
+                                      uint2 pos [[thread_position_in_grid]]) {                                  \n\
+    dst.write(clearValue, pos);                                                                                 \n\
+}                                                                                                               \n\
+                                                                                                                \n\
+kernel void cmdClearColorImage2DUInt(texture2d<uint, access::write> dst [[ texture(0) ]],                       \n\
+                                     constant uint4& clearValue [[ buffer(0) ]],                                \n\
+                                     uint2 pos [[thread_position_in_grid]]) {                                   \n\
+    dst.write(clearValue, pos);                                                                                 \n\
+}                                                                                                               \n\
+                                                                                                                \n\
+kernel void cmdClearColorImage2DInt(texture2d<int, access::write> dst [[ texture(0) ]],                         \n\
+                                    constant int4& clearValue [[ buffer(0) ]],                                  \n\
+                                    uint2 pos [[thread_position_in_grid]]) {                                    \n\
+    dst.write(clearValue, pos);                                                                                 \n\
+}                                                                                                               \n\
                                                                                                                 \n\
 typedef struct {                                                                                                \n\
     uint32_t srcRowStride;                                                                                      \n\
diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandPool.mm b/MoltenVK/MoltenVK/Commands/MVKCommandPool.mm
index fb20ebf..7e8d55a 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCommandPool.mm
+++ b/MoltenVK/MoltenVK/Commands/MVKCommandPool.mm
@@ -59,7 +59,10 @@
 
 		// Command buffers start out in a VK_NOT_READY config result
 		VkResult cbRslt = mvkCmdBuff->getConfigurationResult();
-		if (rslt == VK_SUCCESS && cbRslt != VK_NOT_READY) { rslt = cbRslt; }
+		if (cbRslt != VK_NOT_READY) {
+			if (rslt == VK_SUCCESS) { rslt = cbRslt; }
+			freeCommandBuffers(1, &pCmdBuffer[cbIdx]);
+		}
 	}
 	return rslt;
 }
diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.h b/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.h
index 7672fed..105483f 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.h
+++ b/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.h
@@ -58,6 +58,8 @@
 
 	inline MTLSamplerMinMagFilter getSrcMTLSamplerMinMagFilter() { return (MTLSamplerMinMagFilter)srcFilter; }
 
+	inline MTLTextureType getSrcMTLTextureType() { return (MTLTextureType)srcMTLTextureType; }
+
 	inline bool isSrcArrayType() {
 		return (srcMTLTextureType == MTLTextureType2DArray ||
 #if MVK_MACOS
@@ -417,6 +419,12 @@
 	/** Returns a new MTLComputePipelineState for filling a buffer. */
 	id<MTLComputePipelineState> newCmdFillBufferMTLComputePipelineState(MVKVulkanAPIDeviceObject* owner);
 
+#if MVK_MACOS
+	/** Returns a new MTLComputePipelineState for clearing an image. */
+	id<MTLComputePipelineState> newCmdClearColorImageMTLComputePipelineState(MVKFormatType type,
+																			 MVKVulkanAPIDeviceObject* owner);
+#endif
+
 	/** Returns a new MTLComputePipelineState for copying between a buffer holding compressed data and a 3D image. */
 	id<MTLComputePipelineState> newCmdCopyBufferToImage3DDecompressMTLComputePipelineState(bool needTempBuf,
 																						   MVKVulkanAPIDeviceObject* owner);
diff --git a/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.mm b/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.mm
index a616a64..51640ae 100644
--- a/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.mm
+++ b/MoltenVK/MoltenVK/Commands/MVKCommandResourceFactory.mm
@@ -57,14 +57,14 @@
     vaDesc.format = MTLVertexFormatFloat2;
     vaDesc.bufferIndex = vtxBuffIdx;
     vaDesc.offset = vtxStride;
-    vtxStride += sizeof(simd::float2);
+    vtxStride += sizeof(simd::float4);
 
     // Vertex texture coords
     vaDesc = vaDescArray[1];
-    vaDesc.format = MTLVertexFormatFloat2;
+    vaDesc.format = MTLVertexFormatFloat3;
     vaDesc.bufferIndex = vtxBuffIdx;
     vaDesc.offset = vtxStride;
-    vtxStride += sizeof(simd::float2);
+    vtxStride += sizeof(simd::float4);
 
     // Vertex attribute buffer.
     MTLVertexBufferLayoutDescriptorArray* vbDescArray = vtxDesc.layouts;
@@ -158,7 +158,34 @@
 
 		bool isArrayType = blitKey.isSrcArrayType();
 		bool isLinearFilter = (blitKey.getSrcMTLSamplerMinMagFilter() == MTLSamplerMinMagFilterLinear);
-		NSString* arraySuffix = isArrayType ? @"_array" : @"";
+		NSString* typeSuffix;
+		NSString* coordArg;
+		switch (blitKey.getSrcMTLTextureType()) {
+			case MTLTextureType1D:
+				typeSuffix = @"1d";
+				coordArg = @".x";
+				break;
+			case MTLTextureType1DArray:
+				typeSuffix = @"1d_array";
+				coordArg = @".x";
+				break;
+			case MTLTextureType2D:
+				typeSuffix = @"2d";
+				coordArg = @".xy";
+				break;
+			case MTLTextureType2DArray:
+				typeSuffix = @"2d_array";
+				coordArg = @".xy";
+				break;
+			case MTLTextureType3D:
+				typeSuffix = @"3d";
+				coordArg = @"";
+				break;
+			default:
+				typeSuffix = @"unsupported";
+				coordArg = @"";
+				break;
+		}
 		NSString* sliceArg = isArrayType ? @", subRez.slice" : @"";
 		NSString* srcFilter = isLinearFilter ? @"linear" : @"nearest";
 
@@ -168,7 +195,7 @@
 		[msl appendLineMVK];
 		[msl appendLineMVK: @"typedef struct {"];
 		[msl appendLineMVK: @"    float4 v_position [[position]];"];
-		[msl appendLineMVK: @"    float2 v_texCoord;"];
+		[msl appendLineMVK: @"    float3 v_texCoord;"];
 		[msl appendLineMVK: @"} VaryingsPosTex;"];
 		[msl appendLineMVK];
 		[msl appendLineMVK: @"typedef struct {"];
@@ -183,10 +210,10 @@
 		NSString* funcName = @"fragCmdBlitImage";
 		[msl appendFormat: @"fragment %@4 %@(VaryingsPosTex varyings [[stage_in]],", typeStr, funcName];
 		[msl appendLineMVK];
-		[msl appendFormat: @"                         texture2d%@<%@> tex [[texture(0)]],", arraySuffix, typeStr];
+		[msl appendFormat: @"                         texture%@<%@> tex [[texture(0)]],", typeSuffix, typeStr];
 		[msl appendLineMVK];
 		[msl appendLineMVK: @"                         constant TexSubrez& subRez [[buffer(0)]]) {"];
-		[msl appendFormat: @"    return tex.sample(ce_sampler, varyings.v_texCoord%@, level(subRez.lod));", sliceArg];
+		[msl appendFormat: @"    return tex.sample(ce_sampler, varyings.v_texCoord%@%@, level(subRez.lod));", coordArg, sliceArg];
 		[msl appendLineMVK];
 		[msl appendLineMVK: @"}"];
 
@@ -410,6 +437,34 @@
 	return newMTLComputePipelineState("cmdFillBuffer", owner);
 }
 
+#if MVK_MACOS
+id<MTLComputePipelineState> MVKCommandResourceFactory::newCmdClearColorImageMTLComputePipelineState(MVKFormatType type,
+					MVKVulkanAPIDeviceObject* owner) {
+	const char* funcName;
+	switch (type) {
+		case kMVKFormatColorHalf:
+		case kMVKFormatColorFloat:
+			funcName = "cmdClearColorImage2DFloat";
+			break;
+		case kMVKFormatColorInt8:
+		case kMVKFormatColorInt16:
+		case kMVKFormatColorInt32:
+			funcName = "cmdClearColorImage2DInt";
+			break;
+		case kMVKFormatColorUInt8:
+		case kMVKFormatColorUInt16:
+		case kMVKFormatColorUInt32:
+			funcName = "cmdClearColorImage2DUInt";
+			break;
+		default:
+			owner->reportError(VK_ERROR_FORMAT_NOT_SUPPORTED,
+							   "Format type %u is not supported for clearing with a compute shader.", type);
+			return nil;
+	}
+	return newMTLComputePipelineState(funcName, owner);
+}
+#endif
+
 id<MTLComputePipelineState> MVKCommandResourceFactory::newCmdCopyBufferToImage3DDecompressMTLComputePipelineState(bool needTempBuf,
 																												  MVKVulkanAPIDeviceObject* owner) {
 	return newMTLComputePipelineState(needTempBuf
diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKDevice.h b/MoltenVK/MoltenVK/GPUObjects/MVKDevice.h
index 23d74f5..e3013e1 100644
--- a/MoltenVK/MoltenVK/GPUObjects/MVKDevice.h
+++ b/MoltenVK/MoltenVK/GPUObjects/MVKDevice.h
@@ -362,7 +362,6 @@
 	void initPipelineCacheUUID();
 	uint32_t getHighestMTLFeatureSet();
 	uint64_t getMoltenVKGitRevision();
-	bool getImageViewIsSupported(const VkPhysicalDeviceImageFormatInfo2 *pImageFormatInfo);
 	void populate(VkPhysicalDeviceIDProperties* pDevIdProps);
 	void logGPUInfo();
 
@@ -678,7 +677,7 @@
 	const VkPhysicalDeviceScalarBlockLayoutFeaturesEXT _enabledScalarLayoutFeatures;
 	const VkPhysicalDeviceTexelBufferAlignmentFeaturesEXT _enabledTexelBuffAlignFeatures;
 	const VkPhysicalDeviceVertexAttributeDivisorFeaturesEXT _enabledVtxAttrDivFeatures;
-	const VkPhysicalDevicePortabilitySubsetFeaturesEXTX _enabledPortabilityFeatures;
+	const VkPhysicalDevicePortabilitySubsetFeaturesKHR _enabledPortabilityFeatures;
 
 	/** The list of Vulkan extensions, indicating whether each has been enabled by the app for this device. */
 	const MVKExtensionList _enabledExtensions;
diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm b/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm
index 5aaf418..ac1c96d 100644
--- a/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm
+++ b/MoltenVK/MoltenVK/GPUObjects/MVKDevice.mm
@@ -164,13 +164,24 @@
 				divisorFeatures->vertexAttributeInstanceRateZeroDivisor = true;
 				break;
 			}
-			case VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_PORTABILITY_SUBSET_FEATURES_EXTX: {
-				auto* portabilityFeatures = (VkPhysicalDevicePortabilitySubsetFeaturesEXTX*)next;
-				portabilityFeatures->triangleFans = false;
-				portabilityFeatures->separateStencilMaskRef = true;
+			case VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_PORTABILITY_SUBSET_FEATURES_KHR: {
+				auto* portabilityFeatures = (VkPhysicalDevicePortabilitySubsetFeaturesKHR*)next;
+				portabilityFeatures->constantAlphaColorBlendFactors = true;
 				portabilityFeatures->events = true;
-				portabilityFeatures->standardImageViews = _mvkInstance->getMoltenVKConfiguration()->fullImageViewSwizzle || _metalFeatures.nativeTextureSwizzle;
+				portabilityFeatures->imageViewFormatReinterpretation = true;
+				portabilityFeatures->imageViewFormatSwizzle = (_metalFeatures.nativeTextureSwizzle ||
+															   _mvkInstance->getMoltenVKConfiguration()->fullImageViewSwizzle);
+				portabilityFeatures->imageView2DOn3DImage = false;
+				portabilityFeatures->multisampleArrayImage = _metalFeatures.multisampleArrayTextures;
+				portabilityFeatures->mutableComparisonSamplers = _metalFeatures.depthSampleCompare;
+				portabilityFeatures->pointPolygons = false;
 				portabilityFeatures->samplerMipLodBias = false;
+				portabilityFeatures->separateStencilMaskRef = true;
+				portabilityFeatures->shaderSampleRateInterpolationFunctions = false;
+				portabilityFeatures->tessellationIsolines = false;
+				portabilityFeatures->tessellationPointMode = false;
+				portabilityFeatures->triangleFans = false;
+				portabilityFeatures->vertexAttributeAccessBeyondStride = true;	// Costs additional buffers. Should make configuration switch.
 				break;
 			}
 			case VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_SHADER_INTEGER_FUNCTIONS_2_FEATURES_INTEL: {
@@ -298,8 +309,8 @@
 				divisorProps->maxVertexAttribDivisor = kMVKUndefinedLargeUInt32;
 				break;
 			}
-			case VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_PORTABILITY_SUBSET_PROPERTIES_EXTX: {
-				auto* portabilityProps = (VkPhysicalDevicePortabilitySubsetPropertiesEXTX*)next;
+			case VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_PORTABILITY_SUBSET_PROPERTIES_KHR: {
+				auto* portabilityProps = (VkPhysicalDevicePortabilitySubsetPropertiesKHR*)next;
 				portabilityProps->minVertexInputBindingStrideAlignment = (uint32_t)_metalFeatures.vertexStrideAlignment;
 				break;
 			}
@@ -542,52 +553,12 @@
 
 	if ( !_pixelFormats.isSupported(pImageFormatInfo->format) ) { return VK_ERROR_FORMAT_NOT_SUPPORTED; }
 
-	if ( !getImageViewIsSupported(pImageFormatInfo) ) { return VK_ERROR_FORMAT_NOT_SUPPORTED; }
-
 	return getImageFormatProperties(pImageFormatInfo->format, pImageFormatInfo->type,
 									pImageFormatInfo->tiling, pImageFormatInfo->usage,
 									pImageFormatInfo->flags,
 									&pImageFormatProperties->imageFormatProperties);
 }
 
-// If the image format info links portability image view info, test if an image view of that configuration is supported
-bool MVKPhysicalDevice::getImageViewIsSupported(const VkPhysicalDeviceImageFormatInfo2 *pImageFormatInfo) {
-	for (const auto* next = (VkBaseInStructure*)pImageFormatInfo->pNext; next; next = next->pNext) {
-		switch ((uint32_t)next->sType) {
-			case VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_IMAGE_VIEW_SUPPORT_EXTX: {
-				auto* portImgViewInfo = (VkPhysicalDeviceImageViewSupportEXTX*)next;
-
-				// Create an image view and test whether it could be configured
-				VkImageViewCreateInfo viewInfo = {
-					.sType = VK_STRUCTURE_TYPE_IMAGE_VIEW_CREATE_INFO,
-					.pNext = portImgViewInfo->pNext,
-					.flags = portImgViewInfo->flags,
-					.image = nullptr,
-					.viewType = portImgViewInfo->viewType,
-					.format = portImgViewInfo->format,
-					.components = portImgViewInfo->components,
-					.subresourceRange = {
-						.aspectMask = portImgViewInfo->aspectMask,
-						.baseMipLevel = 0,
-						.levelCount = 1,
-						.baseArrayLayer = 0,
-						.layerCount = 1},
-				};
-                MTLPixelFormat mtlPixFmt = _pixelFormats.getMTLPixelFormat(viewInfo.format);
-				bool useSwizzle;
-				return (MVKImageView::validateSwizzledMTLPixelFormat(&viewInfo, this,
-																	 _metalFeatures.nativeTextureSwizzle,
-																	 _mvkInstance->getMoltenVKConfiguration()->fullImageViewSwizzle,
-																	 mtlPixFmt, useSwizzle) == VK_SUCCESS);
-			}
-			default:
-				break;
-		}
-	}
-
-	return true;
-}
-
 void MVKPhysicalDevice::getExternalBufferProperties(const VkPhysicalDeviceExternalBufferInfo* pExternalBufferInfo,
 													VkExternalBufferProperties* pExternalBufferProperties) {
 	pExternalBufferProperties->externalMemoryProperties = getExternalBufferProperties(pExternalBufferInfo->handleType);
@@ -1268,7 +1239,6 @@
     _features.shaderClipDistance = true;
     _features.shaderInt16 = true;
     _features.multiDrawIndirect = true;
-    _features.variableMultisampleRate = true;
     _features.inheritedQueries = true;
 
 	_features.shaderSampledImageArrayDynamicIndexing = _metalFeatures.arrayOfTextures;
@@ -1410,7 +1380,7 @@
 //    VkBool32    sparseResidency8Samples;
 //    VkBool32    sparseResidency16Samples;
 //    VkBool32    sparseResidencyAliased;
-//    VkBool32    variableMultisampleRate;                      // done
+//    VkBool32    variableMultisampleRate;
 //    VkBool32    inheritedQueries;                             // done
 //} VkPhysicalDeviceFeatures;
 
@@ -3175,8 +3145,8 @@
 	mvkClear(&_enabledPortabilityFeatures);
 
 	// Fetch the available physical device features.
-	VkPhysicalDevicePortabilitySubsetFeaturesEXTX pdPortabilityFeatures;
-	pdPortabilityFeatures.sType = VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_PORTABILITY_SUBSET_FEATURES_EXTX;
+	VkPhysicalDevicePortabilitySubsetFeaturesKHR pdPortabilityFeatures;
+	pdPortabilityFeatures.sType = VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_PORTABILITY_SUBSET_FEATURES_KHR;
 	pdPortabilityFeatures.pNext = NULL;
 
 	VkPhysicalDeviceVertexAttributeDivisorFeaturesEXT pdVtxAttrDivFeatures;
@@ -3322,11 +3292,11 @@
 							   &pdVtxAttrDivFeatures.vertexAttributeInstanceRateDivisor, 2);
 				break;
 			}
-			case VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_PORTABILITY_SUBSET_FEATURES_EXTX: {
-				auto* requestedFeatures = (VkPhysicalDevicePortabilitySubsetFeaturesEXTX*)next;
+			case VK_STRUCTURE_TYPE_PHYSICAL_DEVICE_PORTABILITY_SUBSET_FEATURES_KHR: {
+				auto* requestedFeatures = (VkPhysicalDevicePortabilitySubsetFeaturesKHR*)next;
 				enableFeatures(&_enabledPortabilityFeatures.triangleFans,
 							   &requestedFeatures->triangleFans,
-							   &pdPortabilityFeatures.triangleFans, 5);
+							   &pdPortabilityFeatures.triangleFans, 15);
 				break;
 			}
 			default:
diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKImage.mm b/MoltenVK/MoltenVK/GPUObjects/MVKImage.mm
index 4bcf46d..55c29d1 100644
--- a/MoltenVK/MoltenVK/GPUObjects/MVKImage.mm
+++ b/MoltenVK/MoltenVK/GPUObjects/MVKImage.mm
@@ -960,6 +960,10 @@
 		setConfigurationResult(reportError(VK_ERROR_FEATURE_NOT_PRESENT, "vkCreateImage() : If tiling is VK_IMAGE_TILING_LINEAR, format must not be a depth/stencil format."));
 		isLin = false;
 	}
+	if (getPixelFormats()->getFormatType(pCreateInfo->format) == kMVKFormatCompressed) {
+		setConfigurationResult(reportError(VK_ERROR_FEATURE_NOT_PRESENT, "vkCreateImage() : If tiling is VK_IMAGE_TILING_LINEAR, format must not be a compressed format."));
+		isLin = false;
+	}
 
 	if (pCreateInfo->mipLevels > 1) {
 		setConfigurationResult(reportError(VK_ERROR_FEATURE_NOT_PRESENT, "vkCreateImage() : If tiling is VK_IMAGE_TILING_LINEAR, mipLevels must be 1."));
@@ -1700,7 +1704,9 @@
 	MTLSamplerDescriptor* mtlSampDesc = [MTLSamplerDescriptor new];		// retained
 	mtlSampDesc.sAddressMode = mvkMTLSamplerAddressModeFromVkSamplerAddressMode(pCreateInfo->addressModeU);
 	mtlSampDesc.tAddressMode = mvkMTLSamplerAddressModeFromVkSamplerAddressMode(pCreateInfo->addressModeV);
-    mtlSampDesc.rAddressMode = mvkMTLSamplerAddressModeFromVkSamplerAddressMode(pCreateInfo->addressModeW);
+    if (!pCreateInfo->unnormalizedCoordinates) {
+        mtlSampDesc.rAddressMode = mvkMTLSamplerAddressModeFromVkSamplerAddressMode(pCreateInfo->addressModeW);
+    }
 	mtlSampDesc.minFilter = mvkMTLSamplerMinMagFilterFromVkFilter(pCreateInfo->minFilter);
 	mtlSampDesc.magFilter = mvkMTLSamplerMinMagFilterFromVkFilter(pCreateInfo->magFilter);
     mtlSampDesc.mipFilter = (pCreateInfo->unnormalizedCoordinates
@@ -1754,9 +1760,11 @@
 
 	_requiresConstExprSampler = (pCreateInfo->compareEnable && !_device->_pMetalFeatures->depthSampleCompare) || _ycbcrConversion;
 
-	MTLSamplerDescriptor* mtlSampDesc = newMTLSamplerDescriptor(pCreateInfo);	// temp retain
-    _mtlSamplerState = [getMTLDevice() newSamplerStateWithDescriptor: mtlSampDesc];
-	[mtlSampDesc release];														// temp release
+	@autoreleasepool {
+		MTLSamplerDescriptor* mtlSampDesc = newMTLSamplerDescriptor(pCreateInfo);	// temp retain
+		_mtlSamplerState = [getMTLDevice() newSamplerStateWithDescriptor: mtlSampDesc];
+		[mtlSampDesc release];														// temp release
+	}
 
 	initConstExprSampler(pCreateInfo);
 }
diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm b/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm
index ce836c1..28d7c9c 100644
--- a/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm
+++ b/MoltenVK/MoltenVK/GPUObjects/MVKPipeline.mm
@@ -194,7 +194,7 @@
 
     id<MTLRenderCommandEncoder> mtlCmdEnc = cmdEncoder->_mtlRenderEncoder;
 	id<MTLComputeCommandEncoder> tessCtlEnc;
-    if ( stage != kMVKGraphicsStageTessControl && !mtlCmdEnc ) { return; }   // Pre-renderpass. Come back later.
+    if ( stage == kMVKGraphicsStageRasterization && !mtlCmdEnc ) { return; }   // Pre-renderpass. Come back later.
 
     switch (stage) {
 
@@ -386,6 +386,11 @@
 	_mtlPrimitiveType = MTLPrimitiveTypePoint;
 	if (pCreateInfo->pInputAssemblyState && !isRenderingPoints(pCreateInfo)) {
 		_mtlPrimitiveType = mvkMTLPrimitiveTypeFromVkPrimitiveTopology(pCreateInfo->pInputAssemblyState->topology);
+		// Explicitly fail creation with triangle fan topology.
+		if (pCreateInfo->pInputAssemblyState->topology == VK_PRIMITIVE_TOPOLOGY_TRIANGLE_FAN) {
+			setConfigurationResult(reportError(VK_ERROR_FEATURE_NOT_PRESENT, "Metal does not support triangle fans."));
+			return;
+		}
 	}
 
 	// Tessellation
@@ -1134,8 +1139,8 @@
         vbCnt = pVertexInputDivisorState->vertexBindingDivisorCount;
         for (uint32_t i = 0; i < vbCnt; i++) {
             const VkVertexInputBindingDivisorDescriptionEXT* pVKVB = &pVertexInputDivisorState->pVertexBindingDivisors[i];
-            uint32_t vbIdx = getMetalBufferIndexForVertexAttributeBinding(pVKVB->binding);
-            if (shaderContext.isVertexBufferUsed(vbIdx)) {
+            if (shaderContext.isVertexBufferUsed(pVKVB->binding)) {
+                uint32_t vbIdx = getMetalBufferIndexForVertexAttributeBinding(pVKVB->binding);
                 if ((NSUInteger)inputDesc.layouts[vbIdx].stepFunction == MTLStepFunctionPerInstance ||
 					(NSUInteger)inputDesc.layouts[vbIdx].stepFunction == MTLStepFunctionThreadPositionInGridY) {
                     if (pVKVB->divisor == 0)
@@ -1353,6 +1358,7 @@
     // Multisampling
     if (pCreateInfo->pMultisampleState) {
         plDesc.sampleCount = mvkSampleCountFromVkSampleCountFlagBits(pCreateInfo->pMultisampleState->rasterizationSamples);
+        mvkRenderSubpass->setDefaultSampleCount(pCreateInfo->pMultisampleState->rasterizationSamples);
         plDesc.alphaToCoverageEnabled = pCreateInfo->pMultisampleState->alphaToCoverageEnable;
         plDesc.alphaToOneEnabled = pCreateInfo->pMultisampleState->alphaToOneEnable;
     }
diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKPixelFormats.mm b/MoltenVK/MoltenVK/GPUObjects/MVKPixelFormats.mm
index 65caf00..9fb584e 100644
--- a/MoltenVK/MoltenVK/GPUObjects/MVKPixelFormats.mm
+++ b/MoltenVK/MoltenVK/GPUObjects/MVKPixelFormats.mm
@@ -477,6 +477,14 @@
 
 		mvkEnableFlags(mtlUsage, MTLTextureUsageShaderWrite);
 	}
+#if MVK_MACOS
+    // Clearing a linear image may use shader writes.
+    if (mvkIsAnyFlagEnabled(vkImageUsageFlags, VK_IMAGE_USAGE_TRANSFER_DST_BIT) &&
+        mvkIsAnyFlagEnabled(mtlFmtCaps, kMVKMTLFmtCapsWrite) && isLinear) {
+
+		mvkEnableFlags(mtlUsage, MTLTextureUsageShaderWrite);
+    }
+#endif
 
 	// Render to but only if format supports rendering...
 	if (mvkIsAnyFlagEnabled(vkImageUsageFlags, (VK_IMAGE_USAGE_COLOR_ATTACHMENT_BIT |
diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.h b/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.h
index f8decda..ca51bd0 100644
--- a/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.h
+++ b/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.h
@@ -67,6 +67,9 @@
 	/** Returns the Vulkan sample count of the attachments used in this subpass. */
 	VkSampleCountFlagBits getSampleCount();
 
+	/** Sets the default sample count for when there are no attachments used in this subpass. */
+	void setDefaultSampleCount(VkSampleCountFlagBits count) { _defaultSampleCount = count; }
+
 	/** Returns whether or not this is a multiview subpass. */
 	bool isMultiview() const { return _viewMask != 0; }
 
@@ -141,6 +144,7 @@
 	MVKSmallVector<uint32_t, kMVKDefaultAttachmentCount> _preserveAttachments;
 	VkAttachmentReference2 _depthStencilAttachment;
 	id<MTLTexture> _mtlDummyTex = nil;
+	VkSampleCountFlagBits _defaultSampleCount = VK_SAMPLE_COUNT_1_BIT;
 };
 
 
diff --git a/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm b/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm
index c3eeb6a..464f1d0 100644
--- a/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm
+++ b/MoltenVK/MoltenVK/GPUObjects/MVKRenderPass.mm
@@ -259,10 +259,11 @@
 
 	_mtlDummyTex = nil;
 	if (caUsedCnt == 0 && dsRPAttIdx == VK_ATTACHMENT_UNUSED) {
+		uint32_t sampleCount = mvkSampleCountFromVkSampleCountFlagBits(_defaultSampleCount);
         if (_renderPass->getDevice()->_pMetalFeatures->renderWithoutAttachments) {
             // We support having no attachments.
 #if MVK_MACOS_OR_IOS
-            mtlRPDesc.defaultRasterSampleCount = 1;
+            mtlRPDesc.defaultRasterSampleCount = sampleCount;
 #endif
             return;
         }
@@ -271,11 +272,32 @@
 		VkExtent2D fbExtent = framebuffer->getExtent2D();
 		MTLTextureDescriptor* mtlTexDesc = [MTLTextureDescriptor texture2DDescriptorWithPixelFormat: MTLPixelFormatR8Unorm width: fbExtent.width height: fbExtent.height mipmapped: NO];
 		if (isMultiview()) {
+#if MVK_MACOS
+			if (sampleCount > 1 && _renderPass->getDevice()->_pMetalFeatures->multisampleLayeredRendering) {
+				mtlTexDesc.textureType = MTLTextureType2DMultisampleArray;
+				mtlTexDesc.sampleCount = sampleCount;
+			} else {
+				mtlTexDesc.textureType = MTLTextureType2DArray;
+			}
+#else
 			mtlTexDesc.textureType = MTLTextureType2DArray;
+#endif
 			mtlTexDesc.arrayLength = getViewCountInMetalPass(passIdx);
 		} else if (framebuffer->getLayerCount() > 1) {
+#if MVK_MACOS
+			if (sampleCount > 1 && _renderPass->getDevice()->_pMetalFeatures->multisampleLayeredRendering) {
+				mtlTexDesc.textureType = MTLTextureType2DMultisampleArray;
+				mtlTexDesc.sampleCount = sampleCount;
+			} else {
+				mtlTexDesc.textureType = MTLTextureType2DArray;
+			}
+#else
 			mtlTexDesc.textureType = MTLTextureType2DArray;
+#endif
 			mtlTexDesc.arrayLength = framebuffer->getLayerCount();
+		} else if (sampleCount > 1) {
+			mtlTexDesc.textureType = MTLTextureType2DMultisample;
+			mtlTexDesc.sampleCount = sampleCount;
 		}
 #if MVK_IOS
 		if ([_renderPass->getMTLDevice() supportsFeatureSet: MTLFeatureSet_iOS_GPUFamily1_v3]) {
diff --git a/MoltenVK/MoltenVK/Layers/MVKExtensions.def b/MoltenVK/MoltenVK/Layers/MVKExtensions.def
index cbb1f16..805c7ac 100644
--- a/MoltenVK/MoltenVK/Layers/MVKExtensions.def
+++ b/MoltenVK/MoltenVK/Layers/MVKExtensions.def
@@ -62,6 +62,7 @@
 MVK_EXTENSION(KHR_maintenance2, KHR_MAINTENANCE2, DEVICE)
 MVK_EXTENSION(KHR_maintenance3, KHR_MAINTENANCE3, DEVICE)
 MVK_EXTENSION(KHR_multiview, KHR_MULTIVIEW, DEVICE)
+MVK_EXTENSION(KHR_portability_subset, KHR_PORTABILITY_SUBSET, DEVICE)
 MVK_EXTENSION(KHR_push_descriptor, KHR_PUSH_DESCRIPTOR, DEVICE)
 MVK_EXTENSION(KHR_relaxed_block_layout, KHR_RELAXED_BLOCK_LAYOUT, DEVICE)
 MVK_EXTENSION(KHR_sampler_mirror_clamp_to_edge, KHR_SAMPLER_MIRROR_CLAMP_TO_EDGE, DEVICE)
@@ -92,7 +93,6 @@
 MVK_EXTENSION(EXT_swapchain_colorspace, EXT_SWAPCHAIN_COLOR_SPACE, INSTANCE)
 MVK_EXTENSION(EXT_texel_buffer_alignment, EXT_TEXEL_BUFFER_ALIGNMENT, DEVICE)
 MVK_EXTENSION(EXT_vertex_attribute_divisor, EXT_VERTEX_ATTRIBUTE_DIVISOR, DEVICE)
-MVK_EXTENSION(EXTX_portability_subset, EXTX_PORTABILITY_SUBSET, DEVICE)
 MVK_EXTENSION(MVK_ios_surface, MVK_IOS_SURFACE, INSTANCE)
 MVK_EXTENSION(MVK_macos_surface, MVK_MACOS_SURFACE, INSTANCE)
 MVK_EXTENSION(MVK_moltenvk, MVK_MOLTENVK, INSTANCE)
diff --git a/MoltenVK/MoltenVK/Utility/MVKFoundation.h b/MoltenVK/MoltenVK/Utility/MVKFoundation.h
index 9748abc..d8d1ed2 100644
--- a/MoltenVK/MoltenVK/Utility/MVKFoundation.h
+++ b/MoltenVK/MoltenVK/Utility/MVKFoundation.h
@@ -54,7 +54,7 @@
 /** 2D vertex position and texcoord content. */
 typedef struct {
 	simd::float2 position;
-	simd::float2 texCoord;
+	simd::float3 texCoord;
 } MVKVertexPosTex;
 
 
@@ -376,6 +376,12 @@
 	}
 };
 
+/** Returns the absolute value of the difference of two numbers. */
+template<typename T, typename U>
+constexpr typename std::common_type<T, U>::type mvkAbsDiff(T x, U y) {
+	return x >= y ? x - y : y - x;
+}
+
 /** Returns the greatest common divisor of two numbers. */
 template<typename T>
 constexpr T mvkGreatestCommonDivisorImpl(T a, T b) {
diff --git a/MoltenVK/MoltenVK/Vulkan/vulkan.mm b/MoltenVK/MoltenVK/Vulkan/vulkan.mm
index f0e1824..1fecb88 100644
--- a/MoltenVK/MoltenVK/Vulkan/vulkan.mm
+++ b/MoltenVK/MoltenVK/Vulkan/vulkan.mm
@@ -174,6 +174,7 @@
 	MVKInstance* mvkInst = new MVKInstance(pCreateInfo);
 	*pInstance = mvkInst->getVkInstance();
 	VkResult rslt = mvkInst->getConfigurationResult();
+	if (rslt < 0) { *pInstance = nullptr; mvkInst->destroy(); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -313,6 +314,7 @@
 	MVKDevice* mvkDev = new MVKDevice(mvkPD, pCreateInfo);
 	*pDevice = mvkDev->getVkDevice();
 	VkResult rslt = mvkDev->getConfigurationResult();
+	if (rslt < 0) { *pDevice = nullptr; mvkDev->destroy(); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -428,6 +430,7 @@
 	MVKDeviceMemory* mvkMem = mvkDev->allocateMemory(pAllocateInfo, pAllocator);
 	VkResult rslt = mvkMem->getConfigurationResult();
 	*pMem = (VkDeviceMemory)((rslt == VK_SUCCESS) ? mvkMem : VK_NULL_HANDLE);
+    if (rslt != VK_SUCCESS) { mvkDev->freeMemory(mvkMem, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -618,6 +621,7 @@
 	MVKFence* mvkFence = mvkDev->createFence(pCreateInfo, pAllocator);
 	*pFence = (VkFence)mvkFence;
 	VkResult rslt = mvkFence->getConfigurationResult();
+	if (rslt < 0) { *pFence = VK_NULL_HANDLE; mvkDev->destroyFence(mvkFence, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -680,6 +684,7 @@
 	MVKSemaphore* mvkSem4 = mvkDev->createSemaphore(pCreateInfo, pAllocator);
 	*pSemaphore = (VkSemaphore)mvkSem4;
 	VkResult rslt = mvkSem4->getConfigurationResult();
+	if (rslt < 0) { *pSemaphore = VK_NULL_HANDLE; mvkDev->destroySemaphore(mvkSem4, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -706,6 +711,7 @@
 	MVKEvent* mvkEvent = mvkDev->createEvent(pCreateInfo, pAllocator);
 	*pEvent = (VkEvent)mvkEvent;
 	VkResult rslt = mvkEvent->getConfigurationResult();
+	if (rslt < 0) { *pEvent = VK_NULL_HANDLE; mvkDev->destroyEvent(mvkEvent, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -765,6 +771,7 @@
 	MVKQueryPool* mvkQP = mvkDev->createQueryPool(pCreateInfo, pAllocator);
 	*pQueryPool = (VkQueryPool)mvkQP;
 	VkResult rslt = mvkQP->getConfigurationResult();
+	if (rslt < 0) { *pQueryPool = VK_NULL_HANDLE; mvkDev->destroyQueryPool(mvkQP, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -808,6 +815,7 @@
 	MVKBuffer* mvkBuff = mvkDev->createBuffer(pCreateInfo, pAllocator);
 	*pBuffer = (VkBuffer)mvkBuff;
 	VkResult rslt = mvkBuff->getConfigurationResult();
+	if (rslt < 0) { *pBuffer = VK_NULL_HANDLE; mvkDev->destroyBuffer(mvkBuff, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -834,6 +842,7 @@
     MVKBufferView* mvkBuffView = mvkDev->createBufferView(pCreateInfo, pAllocator);
     *pView = (VkBufferView)mvkBuffView;
     VkResult rslt = mvkBuffView->getConfigurationResult();
+	if (rslt < 0) { *pView = VK_NULL_HANDLE; mvkDev->destroyBufferView(mvkBuffView, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -860,6 +869,7 @@
 	MVKImage* mvkImg = mvkDev->createImage(pCreateInfo, pAllocator);
 	*pImage = (VkImage)mvkImg;
 	VkResult rslt = mvkImg->getConfigurationResult();
+	if (rslt < 0) { *pImage = VK_NULL_HANDLE; mvkDev->destroyImage(mvkImg, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -898,6 +908,7 @@
 	MVKImageView* mvkImgView = mvkDev->createImageView(pCreateInfo, pAllocator);
 	*pView = (VkImageView)mvkImgView;
 	VkResult rslt = mvkImgView->getConfigurationResult();
+	if (rslt < 0) { *pView = VK_NULL_HANDLE; mvkDev->destroyImageView(mvkImgView, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -924,6 +935,7 @@
 	MVKShaderModule* mvkShdrMod = mvkDev->createShaderModule(pCreateInfo, pAllocator);
 	*pShaderModule = (VkShaderModule)mvkShdrMod;
 	VkResult rslt = mvkShdrMod->getConfigurationResult();
+	if (rslt < 0) { *pShaderModule = VK_NULL_HANDLE; mvkDev->destroyShaderModule(mvkShdrMod, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -950,6 +962,7 @@
 	MVKPipelineCache* mvkPLC = mvkDev->createPipelineCache(pCreateInfo, pAllocator);
 	*pPipelineCache = (VkPipelineCache)mvkPLC;
 	VkResult rslt = mvkPLC->getConfigurationResult();
+	if (rslt < 0) { *pPipelineCache = VK_NULL_HANDLE; mvkDev->destroyPipelineCache(mvkPLC, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -1043,6 +1056,7 @@
 	MVKPipelineLayout* mvkPLL = mvkDev->createPipelineLayout(pCreateInfo, pAllocator);
 	*pPipelineLayout = (VkPipelineLayout)mvkPLL;
 	VkResult rslt = mvkPLL->getConfigurationResult();
+	if (rslt < 0) { *pPipelineLayout = VK_NULL_HANDLE; mvkDev->destroyPipelineLayout(mvkPLL, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -1069,6 +1083,7 @@
 	MVKSampler* mvkSamp = mvkDev->createSampler(pCreateInfo, pAllocator);
 	*pSampler = (VkSampler)mvkSamp;
 	VkResult rslt = mvkSamp->getConfigurationResult();
+	if (rslt < 0) { *pSampler = VK_NULL_HANDLE; mvkDev->destroySampler(mvkSamp, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -1095,6 +1110,7 @@
 	MVKDescriptorSetLayout* mvkDSL = mvkDev->createDescriptorSetLayout(pCreateInfo, pAllocator);
 	*pSetLayout = (VkDescriptorSetLayout)mvkDSL;
 	VkResult rslt = mvkDSL->getConfigurationResult();
+	if (rslt < 0) { *pSetLayout = VK_NULL_HANDLE; mvkDev->destroyDescriptorSetLayout(mvkDSL, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -1121,6 +1137,7 @@
 	MVKDescriptorPool* mvkDP = mvkDev->createDescriptorPool(pCreateInfo, pAllocator);
 	*pDescriptorPool = (VkDescriptorPool)mvkDP;
 	VkResult rslt = mvkDP->getConfigurationResult();
+	if (rslt < 0) { *pDescriptorPool = VK_NULL_HANDLE; mvkDev->destroyDescriptorPool(mvkDP, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -1198,6 +1215,7 @@
 	MVKFramebuffer* mvkFB = mvkDev->createFramebuffer(pCreateInfo, pAllocator);
 	*pFramebuffer = (VkFramebuffer)mvkFB;
 	VkResult rslt = mvkFB->getConfigurationResult();
+	if (rslt < 0) { *pFramebuffer = VK_NULL_HANDLE; mvkDev->destroyFramebuffer(mvkFB, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -1224,6 +1242,7 @@
 	MVKRenderPass* mvkRendPass = mvkDev->createRenderPass(pCreateInfo, pAllocator);
 	*pRenderPass = (VkRenderPass)mvkRendPass;
 	VkResult rslt = mvkRendPass->getConfigurationResult();
+	if (rslt < 0) { *pRenderPass = VK_NULL_HANDLE; mvkDev->destroyRenderPass(mvkRendPass, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -1261,6 +1280,7 @@
 	MVKCommandPool* mvkCmdPool = mvkDev->createCommandPool(pCreateInfo, pAllocator);
 	*pCmdPool = (VkCommandPool)mvkCmdPool;
 	VkResult rslt = mvkCmdPool->getConfigurationResult();
+	if (rslt < 0) { *pCmdPool = VK_NULL_HANDLE; mvkDev->destroyCommandPool(mvkCmdPool, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -2146,6 +2166,10 @@
                                                           pAllocator);
     *pDescriptorUpdateTemplate = (VkDescriptorUpdateTemplate)mvkDUT;
     VkResult rslt = mvkDUT->getConfigurationResult();
+    if (rslt < 0) {
+        *pDescriptorUpdateTemplate = VK_NULL_HANDLE;
+        mvkDev->destroyDescriptorUpdateTemplate(mvkDUT, pAllocator);
+    }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -2194,6 +2218,10 @@
 	MVKSamplerYcbcrConversion* mvkSampConv = mvkDev->createSamplerYcbcrConversion(pCreateInfo, pAllocator);
 	*pYcbcrConversion = (VkSamplerYcbcrConversion)mvkSampConv;
 	VkResult rslt = mvkSampConv->getConfigurationResult();
+    if (rslt < 0) {
+        *pYcbcrConversion = VK_NULL_HANDLE;
+        mvkDev->destroySamplerYcbcrConversion(mvkSampConv, pAllocator);
+    }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -2267,6 +2295,7 @@
 	MVKRenderPass* mvkRendPass = mvkDev->createRenderPass(pCreateInfo, pAllocator);
 	*pRenderPass = (VkRenderPass)mvkRendPass;
 	VkResult rslt = mvkRendPass->getConfigurationResult();
+    if (rslt < 0) { *pRenderPass = VK_NULL_HANDLE; mvkDev->destroyRenderPass(mvkRendPass, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -2423,6 +2452,7 @@
     MVKSwapchain* mvkSwpChn = mvkDev->createSwapchain(pCreateInfo, pAllocator);
     *pSwapchain = (VkSwapchainKHR)(mvkSwpChn);
     VkResult rslt = mvkSwpChn->getConfigurationResult();
+    if (rslt < 0) { *pSwapchain = VK_NULL_HANDLE; mvkDev->destroySwapchain(mvkSwpChn, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -2647,6 +2677,7 @@
 	MVKDebugReportCallback* mvkDRCB = mvkInst->createDebugReportCallback(pCreateInfo, pAllocator);
 	*pCallback = (VkDebugReportCallbackEXT)mvkDRCB;
 	VkResult rslt = mvkDRCB->getConfigurationResult();
+    if (rslt < 0) { *pCallback = VK_NULL_HANDLE; mvkInst->destroyDebugReportCallback(mvkDRCB, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -2814,6 +2845,7 @@
 	MVKDebugUtilsMessenger* mvkDUM = mvkInst->createDebugUtilsMessenger(pCreateInfo, pAllocator);
 	*pMessenger = (VkDebugUtilsMessengerEXT)mvkDUM;
 	VkResult rslt = mvkDUM->getConfigurationResult();
+    if (rslt < 0) { *pMessenger = VK_NULL_HANDLE; mvkInst->destroyDebugUtilsMessenger(mvkDUM, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -2890,6 +2922,7 @@
 	MVKSurface* mvkSrfc = mvkInst->createSurface(pCreateInfo, pAllocator);
 	*pSurface = (VkSurfaceKHR)mvkSrfc;
 	VkResult rslt = mvkSrfc->getConfigurationResult();
+    if (rslt < 0) { *pSurface = VK_NULL_HANDLE; mvkInst->destroySurface(mvkSrfc, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
@@ -2934,6 +2967,7 @@
     MVKSurface* mvkSrfc = mvkInst->createSurface(pCreateInfo, pAllocator);
     *pSurface = (VkSurfaceKHR)mvkSrfc;
     VkResult rslt = mvkSrfc->getConfigurationResult();
+    if (rslt < 0) { *pSurface = VK_NULL_HANDLE; mvkInst->destroySurface(mvkSrfc, pAllocator); }
 	MVKTraceVulkanCallEnd();
 	return rslt;
 }
diff --git a/MoltenVK/include/vulkan-portability b/MoltenVK/include/vulkan-portability
deleted file mode 120000
index cfe721d..0000000
--- a/MoltenVK/include/vulkan-portability
+++ /dev/null
@@ -1 +0,0 @@
-../../External/Vulkan-Portability/include/vulkan
\ No newline at end of file
diff --git a/fetchDependencies b/fetchDependencies
index 08a2cb0..28bd8d7 100755
--- a/fetchDependencies
+++ b/fetchDependencies
@@ -252,18 +252,6 @@
 	update_repo ${REPO_NAME} ${REPO_URL} ${REPO_REV}
 fi
 
-# ----------------- Vulkan-Portability -------------------
-
-echo
-echo ========== Vulkan-Portability ==========
-echo
-
-REPO_NAME=Vulkan-Portability
-REPO_URL="https://github.com/KhronosGroup/${REPO_NAME}.git"
-REPO_REV=$(cat "${EXT_REV_DIR}/${REPO_NAME}_repo_revision")
-
-update_repo ${REPO_NAME} ${REPO_URL} ${REPO_REV}
-
 
 # ----------------- SPIRV-Cross -------------------