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

MSL: Force storage images on iOS to use discrete descriptors.

Browse Source 
Writable textures cannot use argument buffers on iOS. They must be
passed as arguments directly to the shader function. Since we won't know
if a given storage image will have the `NonWritable` decoration at the
time we encode the argument buffer, we must therefore pass all storage
images as discrete arguments. Previously, we were throwing an error if
we encountered an argument buffer with a writable texture in it on iOS.
This commit is contained in:
Chip Davis 2019-09-04 13:57:17 -05:00
parent f24654db8c
commit 103817009c
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));
}

19
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,9 +11038,12 @@ void CompilerMSL::analyze_argument_buffers()
else if (!constexpr_sampler)
{
// constexpr samplers are not declared as resources.
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 });
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.

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