AuroraMiddleware/SPIRV-Cross
1
0
Fork 0
Code Issues Pull Requests Packages Projects Releases Wiki Activity

Merge pull request #1144 from cdavis5e/msl-arg-buff-storage-image-ios

Browse Source 
MSL: Force storage images on iOS to use discrete descriptors.
This commit is contained in:
Hans-Kristian Arntzen 2019-09-05 21:33:25 +02:00 committed by GitHub
commit 537bee3cfa
No known key found for this signature in database
GPG Key ID: 4AEE18F83AFDEB23
5 changed files with 48 additions and 9 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

11
reference/opt/shaders-msl/comp/argument-buffers-image-load-store.ios.msl2.argument.comp Normal file
View File

@ -0,0 +1,11 @@
#include <metal_stdlib>
#include <simd/simd.h>
using namespace metal;
kernel void main0(texture2d<float, access::write> uImage [[texture(0)]], texture2d<float> uImageRead [[texture(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
{
int2 _17 = int2(gl_GlobalInvocationID.xy);
uImage.write(uImageRead.read(uint2(_17)), uint2(_17));
}

11
reference/shaders-msl/comp/argument-buffers-image-load-store.ios.msl2.argument.comp Normal file
View File

@ -0,0 +1,11 @@
#include <metal_stdlib>
#include <simd/simd.h>
using namespace metal;
kernel void main0(texture2d<float, access::write> uImage [[texture(0)]], texture2d<float> uImageRead [[texture(1)]], uint3 gl_GlobalInvocationID [[thread_position_in_grid]])
{
int2 coord = int2(gl_GlobalInvocationID.xy);
uImage.write(uImageRead.read(uint2(coord)), uint2(coord));
}

10
shaders-msl/comp/argument-buffers-image-load-store.ios.msl2.argument.comp Normal file
View File

@ -0,0 +1,10 @@
#version 450
layout(set = 0, binding = 1, r32f) writeonly uniform image2D uImage;
layout(set = 0, binding = 2, r32f) readonly uniform image2D uImageRead;
void main()
{
ivec2 coord = ivec2(gl_GlobalInvocationID.xy);
imageStore(uImage, coord, imageLoad(uImageRead, coord));
}

13
spirv_msl.cpp
View File

@ -8161,15 +8161,19 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args)
SmallVector<Resource> resources;
ir.for_each_typed_id<SPIRVariable>([&](uint32_t, SPIRVariable &var) {
ir.for_each_typed_id<SPIRVariable>([&](uint32_t var_id, SPIRVariable &var) {
if ((var.storage == StorageClassUniform || var.storage == StorageClassUniformConstant ||
var.storage == StorageClassPushConstant || var.storage == StorageClassStorageBuffer) &&
!is_hidden_variable(var))
{
auto &type = get_variable_data_type(var);
uint32_t var_id = var.self;
if (var.storage != StorageClassPushConstant)
// Very specifically, image load-store in argument buffers are disallowed on MSL on iOS.
// But we won't know when the argument buffer is encoded whether this image will have
// a NonWritable decoration. So just use discrete arguments for all storage images
// on iOS.
if (!(msl_options.is_ios() && type.basetype == SPIRType::Image && type.image.sampled == 2) &&
var.storage != StorageClassPushConstant)
{
uint32_t desc_set = get_decoration(var_id, DecorationDescriptorSet);
if (descriptor_set_is_argument_buffer(desc_set))
@ -11034,10 +11038,13 @@ void CompilerMSL::analyze_argument_buffers()
else if (!constexpr_sampler)
{
// constexpr samplers are not declared as resources.
if (!msl_options.is_ios() || type.basetype != SPIRType::Image || type.image.sampled != 2)
{
add_resource_name(var_id);
resources_in_set[desc_set].push_back(
{ &var, to_name(var_id), type.basetype, get_metal_resource_index(var, type.basetype), 0 });
}
}
// Check if this descriptor set needs a swizzle buffer.
if (needs_swizzle_buffer_def && is_sampled_image_type(type))

6
spirv_msl.hpp
View File

@ -54,9 +54,9 @@ struct MSLVertexAttr
// Matches the binding index of a MSL resource for a binding within a descriptor set.
// Taken together, the stage, desc_set and binding combine to form a reference to a resource
// descriptor used in a particular shading stage.
// If using MSL 2.0 argument buffers, and the descriptor set is not marked as a discrete descriptor set,
// the binding reference we remap to will become an [[id(N)]] attribute within
// the "descriptor set" argument buffer structure.
// If using MSL 2.0 argument buffers, the descriptor set is not marked as a discrete descriptor set,
// and (for iOS only) the resource is not a storage image (sampled != 2), the binding reference we
// remap to will become an [[id(N)]] attribute within the "descriptor set" argument buffer structure.
// For resources which are bound in the "classic" MSL 1.0 way or discrete descriptors, the remap will become a
// [[buffer(N)]], [[texture(N)]] or [[sampler(N)]] depending on the resource types used.
struct MSLResourceBinding