diff --git a/main.cpp b/main.cpp index d8aff1523..36cd79a65 100644 --- a/main.cpp +++ b/main.cpp @@ -664,6 +664,7 @@ struct CLIArguments bool msl_raw_buffer_tese_input = false; bool msl_multi_patch_workgroup = false; bool msl_vertex_for_tessellation = false; + bool msl_for_mesh_pipeline = false; uint32_t msl_additional_fixed_sample_mask = 0xffffffff; bool msl_arrayed_subpass_input = false; uint32_t msl_r32ui_linear_texture_alignment = 4; @@ -900,23 +901,24 @@ static void print_help_msl() "\t[--msl-enable-frag-output-mask ]:\n\t\tOnly selectively enable fragment outputs. Useful if pipeline does not enable fragment output for certain locations, as pipeline creation might otherwise fail.\n" "\t[--msl-no-clip-distance-user-varying]:\n\t\tDo not emit user varyings to emulate gl_ClipDistance in fragment shaders.\n" "\t[--msl-add-shader-input ]:\n\t\tSpecify the format of the shader input at .\n" - "\t\t can be 'any32', 'any16', 'u16', 'u8', or 'other', to indicate a 32-bit opaque value, 16-bit opaque value, 16-bit unsigned integer, 8-bit unsigned integer, " - "or other-typed variable. is the vector length of the variable, which must be greater than or equal to that declared in the shader. can be 'vertex', " + "\t\t can be 'i32', 'i16', 'i8', 'u32', 'u16', 'u8', 'float', 'half', or 'other',\n\t\tto indicate a 32/16/8-bit integer (i) or unsigned integer (u), floating point, half-precision floating point, " + "or other-typed variable.\n\t\t'any16' or 'any32' can also be used to specify opaque 16-bit or 32-bit value.\n\t\t is the vector length of the variable, which must be greater than or equal to that declared in the shader. can be 'vertex', " "'primitive', or 'patch' to indicate a per-vertex, per-primitive, or per-patch variable.\n" "\t\tUseful if shader stage interfaces don't match up, as pipeline creation might otherwise fail.\n" "\t[--msl-add-shader-output ]:\n\t\tSpecify the format of the shader output at .\n" - "\t\t can be 'any32', 'any16', 'u16', 'u8', or 'other', to indicate a 32-bit opaque value, 16-bit opaque value, 16-bit unsigned integer, 8-bit unsigned integer, " - "or other-typed variable. is the vector length of the variable, which must be greater than or equal to that declared in the shader. can be 'vertex', " + "\t\t can be 'i32', 'i16', 'i8', 'u32', 'u16', 'u8', 'float', 'half', or 'other',\n\t\tto indicate a 32/16/8-bit integer (i) or unsigned integer (u), floating point, half-precision floating point, " + "or other-typed variable.\n\t\t'any16' or 'any32' can also be used to specify opaque 16-bit or 32-bit value.\n\t\t is the vector length of the variable, which must be greater than or equal to that declared in the shader. can be 'vertex', " "'primitive', or 'patch' to indicate a per-vertex, per-primitive, or per-patch variable.\n" "\t\tUseful if shader stage interfaces don't match up, as pipeline creation might otherwise fail.\n" "\t[--msl-shader-input ]:\n\t\tSpecify the format of the shader input at .\n" - "\t\t can be 'any32', 'any16', 'u16', 'u8', or 'other', to indicate a 32-bit opaque value, 16-bit opaque value, 16-bit unsigned integer, 8-bit unsigned integer, " - "or other-typed variable. is the vector length of the variable, which must be greater than or equal to that declared in the shader." - "\t\tEquivalent to --msl-add-shader-input with a rate of 'vertex'.\n" + "\t[--msl-shader-attribute ]:\n\t\tSpecify the vertex attribute at .\n" + "\t\t can be 'i32', 'i16', 'i8', 'u32', 'u16', 'u8', 'float', 'half', or 'other',\n\t\tto indicate a 32/16/8-bit integer (i) or unsigned integer (u), floating point, half-precision floating point, " + "or other-typed variable.\n\t\t'any16' or 'any32' can also be used to specify opaque 16-bit or 32-bit value.\n\t\t is the vector length of the variable, which must be greater than or equal to that declared in the shader." + "\n\t\tEquivalent to --msl-add-shader-input with a rate of 'vertex'.\n" "\t[--msl-shader-output ]:\n\t\tSpecify the format of the shader output at .\n" - "\t\t can be 'any32', 'any16', 'u16', 'u8', or 'other', to indicate a 32-bit opaque value, 16-bit opaque value, 16-bit unsigned integer, 8-bit unsigned integer, " - "or other-typed variable. is the vector length of the variable, which must be greater than or equal to that declared in the shader." - "\t\tEquivalent to --msl-add-shader-output with a rate of 'vertex'.\n" + "\t\t can be 'i32', 'i16', 'i8', 'u32', 'u16', 'u8', 'float', 'half', or 'other',\n\t\tto indicate a 32/16/8-bit integer (i) or unsigned integer (u), floating point, half-precision floating point, " + "or other-typed variable.\n\t\t'any16' or 'any32' can also be used to specify opaque 16-bit or 32-bit value.\n\t\t is the vector length of the variable, which must be greater than or equal to that declared in the shader." + "\n\t\tEquivalent to --msl-add-shader-output with a rate of 'vertex'.\n" "\t[--msl-raw-buffer-tese-input]:\n\t\tUse raw buffers for tessellation evaluation input.\n" "\t\tThis allows the use of nested structures and arrays.\n" "\t\tIn a future version of SPIRV-Cross, this will become the default.\n" @@ -926,6 +928,7 @@ static void print_help_msl() "\t\tIn a future version of SPIRV-Cross, this will become the default.\n" "\t[--msl-vertex-for-tessellation]:\n\t\tWhen handling a vertex shader, marks it as one that will be used with a new-style tessellation control shader.\n" "\t\tThe vertex shader is output to MSL as a compute kernel which outputs vertices to the buffer in the order they are received, rather than in index order as with --msl-capture-output normally.\n" + "\t[--msl-for-mesh-pipeline]:\n\t\tWhen handling a vertex shader, marks it as one that will be used in a mesh pipeline in conjunction with a geometry shader.\n" "\t[--msl-additional-fixed-sample-mask ]:\n" "\t\tSet an additional fixed sample mask. If the shader outputs a sample mask, then the final sample mask will be a bitwise AND of the two.\n" "\t[--msl-arrayed-subpass-input]:\n\t\tAssume that images of dimension SubpassData have multiple layers. Layered input attachments are accessed relative to BuiltInLayer.\n" @@ -1219,6 +1222,7 @@ static string compile_iteration(const CLIArguments &args, std::vector msl_opts.raw_buffer_tese_input = args.msl_raw_buffer_tese_input; msl_opts.multi_patch_workgroup = args.msl_multi_patch_workgroup; msl_opts.vertex_for_tessellation = args.msl_vertex_for_tessellation; + msl_opts.for_mesh_pipeline = args.msl_for_mesh_pipeline; msl_opts.additional_fixed_sample_mask = args.msl_additional_fixed_sample_mask; msl_opts.arrayed_subpass_input = args.msl_arrayed_subpass_input; msl_opts.r32ui_linear_texture_alignment = args.msl_r32ui_linear_texture_alignment; @@ -1554,6 +1558,34 @@ static string compile_iteration(const CLIArguments &args, std::vector return ret; } +static MSLShaderVariableFormat parse_format(const char *text) +{ + MSLShaderVariableFormat format; + if (strcmp(text, "i8") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_INT8; + else if (strcmp(text, "i16") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_INT16; + else if (strcmp(text, "i32") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_INT32; + else if (strcmp(text, "u8") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_UINT8; + else if (strcmp(text, "u16") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_UINT16; + else if (strcmp(text, "u32") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_UINT32; + else if (strcmp(text, "float") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_FLOAT; + else if (strcmp(text, "half") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_HALF; + else if (strcmp(text, "any16") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_ANY16; + else if (strcmp(text, "any32") == 0) + format = MSL_SHADER_VARIABLE_FORMAT_ANY32; + else + format = MSL_SHADER_VARIABLE_FORMAT_OTHER; + return format; +} + static int main_inner(int argc, char *argv[]) { CLIArguments args; @@ -1685,16 +1717,7 @@ static int main_inner(int argc, char *argv[]) // Make sure next_uint() is called in-order. input.location = parser.next_uint(); const char *format = parser.next_value_string("other"); - if (strcmp(format, "any32") == 0) - input.format = MSL_SHADER_VARIABLE_FORMAT_ANY32; - else if (strcmp(format, "any16") == 0) - input.format = MSL_SHADER_VARIABLE_FORMAT_ANY16; - else if (strcmp(format, "u16") == 0) - input.format = MSL_SHADER_VARIABLE_FORMAT_UINT16; - else if (strcmp(format, "u8") == 0) - input.format = MSL_SHADER_VARIABLE_FORMAT_UINT8; - else - input.format = MSL_SHADER_VARIABLE_FORMAT_OTHER; + input.format = parse_format(format); input.vecsize = parser.next_uint(); const char *rate = parser.next_value_string("vertex"); if (strcmp(rate, "primitive") == 0) @@ -1710,16 +1733,7 @@ static int main_inner(int argc, char *argv[]) // Make sure next_uint() is called in-order. output.location = parser.next_uint(); const char *format = parser.next_value_string("other"); - if (strcmp(format, "any32") == 0) - output.format = MSL_SHADER_VARIABLE_FORMAT_ANY32; - else if (strcmp(format, "any16") == 0) - output.format = MSL_SHADER_VARIABLE_FORMAT_ANY16; - else if (strcmp(format, "u16") == 0) - output.format = MSL_SHADER_VARIABLE_FORMAT_UINT16; - else if (strcmp(format, "u8") == 0) - output.format = MSL_SHADER_VARIABLE_FORMAT_UINT8; - else - output.format = MSL_SHADER_VARIABLE_FORMAT_OTHER; + output.format = parse_format(format); output.vecsize = parser.next_uint(); const char *rate = parser.next_value_string("vertex"); if (strcmp(rate, "primitive") == 0) @@ -1730,21 +1744,26 @@ static int main_inner(int argc, char *argv[]) output.rate = MSL_SHADER_VARIABLE_RATE_PER_VERTEX; args.msl_shader_outputs.push_back(output); }); + cbs.add("--msl-shader-attribute", [&args](CLIParser &parser) { + MSLShaderInterfaceVariable input; + // Make sure next_uint() is called in-order. + input.location = parser.next_uint(); + const char *format = parser.next_value_string("other"); + input.format = parse_format(format); + input.vecsize = parser.next_uint(); + + input.offset = parser.next_uint(); + input.stride = parser.next_uint(); + input.binding = parser.next_uint(); + + args.msl_shader_inputs.push_back(input); + }); cbs.add("--msl-shader-input", [&args](CLIParser &parser) { MSLShaderInterfaceVariable input; // Make sure next_uint() is called in-order. input.location = parser.next_uint(); const char *format = parser.next_value_string("other"); - if (strcmp(format, "any32") == 0) - input.format = MSL_SHADER_VARIABLE_FORMAT_ANY32; - else if (strcmp(format, "any16") == 0) - input.format = MSL_SHADER_VARIABLE_FORMAT_ANY16; - else if (strcmp(format, "u16") == 0) - input.format = MSL_SHADER_VARIABLE_FORMAT_UINT16; - else if (strcmp(format, "u8") == 0) - input.format = MSL_SHADER_VARIABLE_FORMAT_UINT8; - else - input.format = MSL_SHADER_VARIABLE_FORMAT_OTHER; + input.format = parse_format(format); input.vecsize = parser.next_uint(); args.msl_shader_inputs.push_back(input); }); @@ -1753,22 +1772,14 @@ static int main_inner(int argc, char *argv[]) // Make sure next_uint() is called in-order. output.location = parser.next_uint(); const char *format = parser.next_value_string("other"); - if (strcmp(format, "any32") == 0) - output.format = MSL_SHADER_VARIABLE_FORMAT_ANY32; - else if (strcmp(format, "any16") == 0) - output.format = MSL_SHADER_VARIABLE_FORMAT_ANY16; - else if (strcmp(format, "u16") == 0) - output.format = MSL_SHADER_VARIABLE_FORMAT_UINT16; - else if (strcmp(format, "u8") == 0) - output.format = MSL_SHADER_VARIABLE_FORMAT_UINT8; - else - output.format = MSL_SHADER_VARIABLE_FORMAT_OTHER; + output.format = parse_format(format); output.vecsize = parser.next_uint(); args.msl_shader_outputs.push_back(output); }); cbs.add("--msl-raw-buffer-tese-input", [&args](CLIParser &) { args.msl_raw_buffer_tese_input = true; }); cbs.add("--msl-multi-patch-workgroup", [&args](CLIParser &) { args.msl_multi_patch_workgroup = true; }); cbs.add("--msl-vertex-for-tessellation", [&args](CLIParser &) { args.msl_vertex_for_tessellation = true; }); + cbs.add("--msl-for-mesh-pipeline", [&args](CLIParser &) { args.msl_for_mesh_pipeline = true; }); cbs.add("--msl-additional-fixed-sample-mask", [&args](CLIParser &parser) { args.msl_additional_fixed_sample_mask = parser.next_hex_uint(); }); cbs.add("--msl-arrayed-subpass-input", [&args](CLIParser &) { args.msl_arrayed_subpass_input = true; }); diff --git a/reference/shaders-msl/geom/basic.msl31.geom b/reference/shaders-msl/geom/basic.msl31.geom new file mode 100644 index 000000000..a26256c7d --- /dev/null +++ b/reference/shaders-msl/geom/basic.msl31.geom @@ -0,0 +1,159 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" +#pragma clang diagnostic ignored "-Wmissing-braces" + +#include +#include +#include + +using namespace metal; + +template +struct spvUnsafeArray +{ + T elements[Num ? Num : 1]; + + thread T& operator [] (size_t pos) thread + { + return elements[pos]; + } + constexpr const thread T& operator [] (size_t pos) const thread + { + return elements[pos]; + } + + device T& operator [] (size_t pos) device + { + return elements[pos]; + } + constexpr const device T& operator [] (size_t pos) const device + { + return elements[pos]; + } + + constexpr const constant T& operator [] (size_t pos) const constant + { + return elements[pos]; + } + + threadgroup T& operator [] (size_t pos) threadgroup + { + return elements[pos]; + } + constexpr const threadgroup T& operator [] (size_t pos) const threadgroup + { + return elements[pos]; + } +}; + +template +struct spvMeshStream +{ + using mesh_t = metal::mesh; + thread mesh_t &meshOut; + int currentVertex = 0; + int currentIndex = 0; + int currentVertexInPrimitive = 0; + int currentPrimitive = 0; + thread P &primitiveData; + thread V &vertexData; + spvMeshStream(thread mesh_t &_meshOut, thread V &_v, thread P &_p) : meshOut(_meshOut), primitiveData(_p), vertexData(_v) + { + } + ~spvMeshStream() + { + meshOut.set_primitive_count(currentPrimitive); + } + int VperP() + { + if (T == metal::topology::triangle) return 3; + else if (T == metal::topology::line) return 2; + else /* if (T == metal::topology::point) */ return 1; + } + void EndPrimitive() + { + currentVertexInPrimitive = 0; + } + void EmitVertex() + { + meshOut.set_vertex(currentVertex++, vertexData); + currentVertexInPrimitive++; + if (currentVertexInPrimitive >= VperP()) + { + if (T == metal::topology::triangle) meshOut.set_index(currentIndex++, currentVertex-3); + if (T == metal::topology::triangle || T == metal::topology::line) meshOut.set_index(currentIndex++, currentVertex-2); + meshOut.set_index(currentIndex++, currentVertex-1); + meshOut.set_primitive(currentPrimitive++, primitiveData); + } + } +}; +struct VertexData +{ + float3 normal; + float4 pos; +}; + +struct main0_out_2 +{ +}; +struct main0_out_1 +{ + float3 vNormal [[user(locn0)]]; + float4 gl_Position [[position]]; +}; + +struct main0_out_2_1 +{ +}; +struct main0_in +{ + spvUnsafeArray vin; + spvUnsafeArray pos; +}; + +enum { VERTEX_COUNT = 3, PRIMITIVE_COUNT = 1 }; +using mesh_stream_t = spvMeshStream; +void main0(mesh_stream_t::mesh_t spvMeshOut, main0_in in) +{ + main0_out_1 out = {}; + main0_out_2_1 out_1 = {}; + mesh_stream_t meshStream(spvMeshOut, out, out_1); + out.gl_Position = in.pos[0]; + out.vNormal = in.vin[0].normal; + meshStream.EmitVertex(); + out.gl_Position = in.pos[1]; + out.vNormal = in.vin[1].normal; + meshStream.EmitVertex(); + out.gl_Position = in.pos[2]; + out.vNormal = in.vin[2].normal; + meshStream.EmitVertex(); + meshStream.EndPrimitive(); +} + +struct Payload +{ + struct + { + struct + { + VertexData vin [[user(locn0)]]; + float4 pos [[user(locn2)]]; + } in; + } vertices[3]; +}; +[[mesh]] void main0(mesh_stream_t::mesh_t outputMesh, const object_data Payload &payload [[payload]], + +uint lid [[thread_index_in_threadgroup]], uint tid [[threadgroup_position_in_grid]]) +{ + main0_in in; + const unsigned long vertexCount = 3; + for (unsigned long i = 0; i < vertexCount; ++i) + { + auto out = payload.vertices[i]; + if (i < sizeof(in.pos) / sizeof(in.pos[0])) + in.pos[i] = out.in.pos; + if (i < sizeof(in.vin) / sizeof(in.vin[0])) + in.vin[i] = out.in.vin; + } + main0(outputMesh, in + ); +} diff --git a/reference/shaders-msl/geom/lines.msl31.geom b/reference/shaders-msl/geom/lines.msl31.geom new file mode 100644 index 000000000..b91271709 --- /dev/null +++ b/reference/shaders-msl/geom/lines.msl31.geom @@ -0,0 +1,152 @@ +#pragma clang diagnostic ignored "-Wmissing-prototypes" +#pragma clang diagnostic ignored "-Wmissing-braces" + +#include +#include +#include + +using namespace metal; + +template +struct spvUnsafeArray +{ + T elements[Num ? Num : 1]; + + thread T& operator [] (size_t pos) thread + { + return elements[pos]; + } + constexpr const thread T& operator [] (size_t pos) const thread + { + return elements[pos]; + } + + device T& operator [] (size_t pos) device + { + return elements[pos]; + } + constexpr const device T& operator [] (size_t pos) const device + { + return elements[pos]; + } + + constexpr const constant T& operator [] (size_t pos) const constant + { + return elements[pos]; + } + + threadgroup T& operator [] (size_t pos) threadgroup + { + return elements[pos]; + } + constexpr const threadgroup T& operator [] (size_t pos) const threadgroup + { + return elements[pos]; + } +}; + +template +struct spvMeshStream +{ + using mesh_t = metal::mesh; + thread mesh_t &meshOut; + int currentVertex = 0; + int currentIndex = 0; + int currentVertexInPrimitive = 0; + int currentPrimitive = 0; + thread P &primitiveData; + thread V &vertexData; + spvMeshStream(thread mesh_t &_meshOut, thread V &_v, thread P &_p) : meshOut(_meshOut), primitiveData(_p), vertexData(_v) + { + } + ~spvMeshStream() + { + meshOut.set_primitive_count(currentPrimitive); + } + int VperP() + { + if (T == metal::topology::triangle) return 3; + else if (T == metal::topology::line) return 2; + else /* if (T == metal::topology::point) */ return 1; + } + void EndPrimitive() + { + currentVertexInPrimitive = 0; + } + void EmitVertex() + { + meshOut.set_vertex(currentVertex++, vertexData); + currentVertexInPrimitive++; + if (currentVertexInPrimitive >= VperP()) + { + if (T == metal::topology::triangle) meshOut.set_index(currentIndex++, currentVertex-3); + if (T == metal::topology::triangle || T == metal::topology::line) meshOut.set_index(currentIndex++, currentVertex-2); + meshOut.set_index(currentIndex++, currentVertex-1); + meshOut.set_primitive(currentPrimitive++, primitiveData); + } + } +}; +struct VertexData +{ + float3 normal; + float4 position; +}; + +struct main0_out_2 +{ +}; +struct main0_out_1 +{ + float3 vNormal [[user(locn0)]]; + float4 gl_Position [[position]]; +}; + +struct main0_out_2_1 +{ +}; +struct main0_in +{ + spvUnsafeArray vin; +}; + +enum { VERTEX_COUNT = 2, PRIMITIVE_COUNT = 1 }; +using mesh_stream_t = spvMeshStream; +void main0(mesh_stream_t::mesh_t spvMeshOut, main0_in in) +{ + main0_out_1 out = {}; + main0_out_2_1 out_1 = {}; + mesh_stream_t meshStream(spvMeshOut, out, out_1); + out.gl_Position = in.vin[0].position; + out.vNormal = in.vin[0].normal; + meshStream.EmitVertex(); + out.gl_Position = in.vin[1].position; + out.vNormal = in.vin[1].normal; + meshStream.EmitVertex(); + meshStream.EndPrimitive(); +} + +struct Payload +{ + struct + { + struct + { + VertexData vin [[user(locn0)]]; + } in; + } vertices[2]; +}; +[[mesh]] void main0(mesh_stream_t::mesh_t outputMesh, const object_data Payload &payload [[payload]], + +uint lid [[thread_index_in_threadgroup]], uint tid [[threadgroup_position_in_grid]]) +{ + main0_in in; + const unsigned long vertexCount = 2; + for (unsigned long i = 0; i < vertexCount; ++i) + { + auto out = payload.vertices[i]; + if (i < sizeof(in.vin) / sizeof(in.vin[0])) + in.vin[i] = out.in.vin; + } + main0(outputMesh, in + ); +} diff --git a/reference/shaders-msl/tesc/arrayed-block-io.multi-patch.tesc b/reference/shaders-msl/tesc/arrayed-block-io.multi-patch.tesc index c11c7410c..f54a07412 100644 --- a/reference/shaders-msl/tesc/arrayed-block-io.multi-patch.tesc +++ b/reference/shaders-msl/tesc/arrayed-block-io.multi-patch.tesc @@ -68,7 +68,7 @@ struct main0_patchOut struct main0_in { float3 in_tc_attr; - ushort2 m_179; + ushort2 m_182; }; kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], constant uint* spvIndirectParams [[buffer(29)]], device main0_patchOut* spvPatchOut [[buffer(27)]], device MTLQuadTessellationFactorsHalf* spvTessLevel [[buffer(26)]], device main0_in* spvIn [[buffer(22)]]) diff --git a/reference/shaders-msl/tesc/matrix-output.multi-patch.tesc b/reference/shaders-msl/tesc/matrix-output.multi-patch.tesc index 98b9dd052..9cc8bc781 100644 --- a/reference/shaders-msl/tesc/matrix-output.multi-patch.tesc +++ b/reference/shaders-msl/tesc/matrix-output.multi-patch.tesc @@ -13,7 +13,7 @@ struct main0_out struct main0_in { float3 in_tc_attr; - ushort2 m_103; + ushort2 m_106; }; kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device main0_out* spvOut [[buffer(28)]], constant uint* spvIndirectParams [[buffer(29)]], device MTLQuadTessellationFactorsHalf* spvTessLevel [[buffer(26)]], device main0_in* spvIn [[buffer(22)]]) diff --git a/reference/shaders-msl/tesc/struct-output.multi-patch.tesc b/reference/shaders-msl/tesc/struct-output.multi-patch.tesc index eaab245c1..620ec7842 100644 --- a/reference/shaders-msl/tesc/struct-output.multi-patch.tesc +++ b/reference/shaders-msl/tesc/struct-output.multi-patch.tesc @@ -20,7 +20,7 @@ struct main0_out struct main0_in { float3 in_tc_attr; - ushort2 m_107; + ushort2 m_110; }; kernel void main0(uint3 gl_GlobalInvocationID [[thread_position_in_grid]], device main0_out* spvOut [[buffer(28)]], constant uint* spvIndirectParams [[buffer(29)]], device MTLQuadTessellationFactorsHalf* spvTessLevel [[buffer(26)]], device main0_in* spvIn [[buffer(22)]]) diff --git a/reference/shaders-msl/tesc/water_tess.multi-patch.tesc b/reference/shaders-msl/tesc/water_tess.multi-patch.tesc index 0cd540ca1..e795d025c 100644 --- a/reference/shaders-msl/tesc/water_tess.multi-patch.tesc +++ b/reference/shaders-msl/tesc/water_tess.multi-patch.tesc @@ -24,7 +24,7 @@ struct main0_patchOut struct main0_in { float3 vPatchPosBase; - ushort2 m_430; + ushort2 m_433; }; static inline __attribute__((always_inline)) diff --git a/reference/shaders-msl/vert/basic.msl31.for-mesh.vert b/reference/shaders-msl/vert/basic.msl31.for-mesh.vert new file mode 100644 index 000000000..95e97a5cd --- /dev/null +++ b/reference/shaders-msl/vert/basic.msl31.for-mesh.vert @@ -0,0 +1,84 @@ +#include +#include +#include + +using namespace metal; + +struct UBO +{ + float4x4 uMVP; +}; + +struct main0_out +{ + float3 vNormal [[user(locn0)]]; + float4 gl_Position [[position]]; +}; + +struct main0_in +{ + float4 aVertex; + float3 aNormal; +}; + +main0_out main0(main0_in in, constant UBO& _16) +{ + main0_out out = {}; + out.gl_Position = _16.uMVP * in.aVertex; + out.vNormal = in.aNormal; + return out; +} + +struct Payload +{ + main0_out vertices[3]; +}; +struct DrawInfo +{ + int32_t indexed; + int32_t indexSize; + int64_t indexBuffer; +}; +[[object]] void main0(object_data Payload &payload [[payload]], mesh_grid_properties meshGridProperties, constant DrawInfo *drawInfo [[buffer(20)]], +device uchar *vb29 [[buffer(29)]], +device uchar *vb30 [[buffer(30)]], +constant UBO& _16 [[buffer(0)]], +uint3 positionInGrid [[thread_position_in_grid]]) +{ + int startingIndex = positionInGrid.x * 3; + int vertexCount = 3; + int instanceIndex = positionInGrid.y; + for (int i = 0; i < vertexCount; ++i) + { + uint vertexIndex; + if (drawInfo->indexed) + { + + if (drawInfo->indexSize == 1) { + vertexIndex = ((constant uchar *)drawInfo->indexBuffer)[startingIndex + i]; + if (vertexIndex == 0xff) { + return; + } + } else if (drawInfo->indexSize == 2) { + vertexIndex = ((constant ushort *)drawInfo->indexBuffer)[startingIndex + i]; + if (vertexIndex == 0xffff) { + return; + } + } else { + vertexIndex = ((constant uint *)drawInfo->indexBuffer)[startingIndex + i]; + if (vertexIndex == 0xffffffff) { + return; + } + } + } + else vertexIndex = startingIndex + i; + main0_in in; + in.aVertex = *(device packed_float4 *)(vb30 + 0 + vertexIndex * 16); + in.aNormal = *(device packed_float3 *)(vb29 + 0 + vertexIndex * 12); + payload.vertices[i] = main0( + in + , _16 + ); + } + meshGridProperties.set_threadgroups_per_grid(uint3(1, 1, 1)); +} diff --git a/shaders-msl/geom/basic.msl31.geom b/shaders-msl/geom/basic.msl31.geom new file mode 100644 index 000000000..ea193fecc --- /dev/null +++ b/shaders-msl/geom/basic.msl31.geom @@ -0,0 +1,31 @@ +#version 310 es +#extension GL_EXT_geometry_shader : require + +layout(triangles, invocations = 4) in; +layout(triangle_strip, max_vertices = 3) out; + +layout(location = 0) in VertexData { + vec3 normal; + vec4 pos; +} vin[]; + +layout(location = 2) in vec4 pos[]; + +layout(location = 0) out vec3 vNormal; + +void main() +{ + gl_Position = pos[0]; + vNormal = vin[0].normal; + EmitVertex(); + + gl_Position = pos[1]; + vNormal = vin[1].normal; + EmitVertex(); + + gl_Position = pos[2]; + vNormal = vin[2].normal; + EmitVertex(); + + EndPrimitive(); +} diff --git a/shaders-msl/geom/lines.msl31.geom b/shaders-msl/geom/lines.msl31.geom new file mode 100644 index 000000000..0191ebcff --- /dev/null +++ b/shaders-msl/geom/lines.msl31.geom @@ -0,0 +1,25 @@ +#version 310 es +#extension GL_EXT_geometry_shader : require + +layout(lines) in; +layout(line_strip, max_vertices = 2) out; + +layout(location = 0) in VertexData { + vec3 normal; + vec4 position; +} vin[]; + +layout(location = 0) out vec3 vNormal; + +void main() +{ + gl_Position = vin[0].position; + vNormal = vin[0].normal; + EmitVertex(); + + gl_Position = vin[1].position; + vNormal = vin[1].normal; + EmitVertex(); + + EndPrimitive(); +} diff --git a/shaders-msl/vert/basic.msl31.for-mesh.vert b/shaders-msl/vert/basic.msl31.for-mesh.vert new file mode 100644 index 000000000..8191dc2d0 --- /dev/null +++ b/shaders-msl/vert/basic.msl31.for-mesh.vert @@ -0,0 +1,17 @@ +#version 310 es + +layout(std140) uniform UBO +{ + uniform mat4 uMVP; +}; + +layout(location = 0) in vec4 aVertex; +layout(location = 1) in vec3 aNormal; + +layout(location = 0) out vec3 vNormal; + +void main() +{ + gl_Position = uMVP * aVertex; + vNormal = aNormal; +} diff --git a/spirv_cross_c.h b/spirv_cross_c.h index 0d8e6e10a..534411233 100644 --- a/spirv_cross_c.h +++ b/spirv_cross_c.h @@ -296,10 +296,17 @@ typedef enum spvc_msl_shader_variable_format SPVC_MSL_SHADER_VARIABLE_FORMAT_OTHER = 0, SPVC_MSL_SHADER_VARIABLE_FORMAT_UINT8 = 1, SPVC_MSL_SHADER_VARIABLE_FORMAT_UINT16 = 2, - SPVC_MSL_SHADER_VARIABLE_FORMAT_ANY16 = 3, - SPVC_MSL_SHADER_VARIABLE_FORMAT_ANY32 = 4, + SPVC_MSL_SHADER_VARIABLE_FORMAT_UINT32 = 3, + SPVC_MSL_SHADER_VARIABLE_FORMAT_FLOAT = 4, + SPVC_MSL_SHADER_VARIABLE_FORMAT_INT8 = 5, + SPVC_MSL_SHADER_VARIABLE_FORMAT_INT16 = 6, + SPVC_MSL_SHADER_VARIABLE_FORMAT_INT32 = 7, + SPVC_MSL_SHADER_VARIABLE_FORMAT_HALF = 8, + + // Deprecated aliases. + SPVC_MSL_SHADER_VARIABLE_FORMAT_ANY16 = SPVC_MSL_SHADER_VARIABLE_FORMAT_UINT16, + SPVC_MSL_SHADER_VARIABLE_FORMAT_ANY32 = SPVC_MSL_SHADER_VARIABLE_FORMAT_UINT32, - /* Deprecated names. */ SPVC_MSL_VERTEX_FORMAT_OTHER = SPVC_MSL_SHADER_VARIABLE_FORMAT_OTHER, SPVC_MSL_VERTEX_FORMAT_UINT8 = SPVC_MSL_SHADER_VARIABLE_FORMAT_UINT8, SPVC_MSL_VERTEX_FORMAT_UINT16 = SPVC_MSL_SHADER_VARIABLE_FORMAT_UINT16, diff --git a/spirv_msl.cpp b/spirv_msl.cpp index 8471e34f3..1a6ad0c89 100644 --- a/spirv_msl.cpp +++ b/spirv_msl.cpp @@ -36,6 +36,22 @@ static const uint32_t k_unknown_location = ~0u; static const uint32_t k_unknown_component = ~0u; static const char *force_inline = "static inline __attribute__((always_inline))"; + +static bool builtin_is_per_primitive_mesh_output(BuiltIn builtin) +{ + switch (builtin) + { + case BuiltInLayer: + case BuiltInViewportIndex: + case BuiltInPrimitiveId: + case BuiltInCullPrimitiveEXT: + return true; + default: break; + } + + return false; +} + CompilerMSL::CompilerMSL(std::vector spirv_) : CompilerGLSL(std::move(spirv_)) { @@ -244,11 +260,13 @@ void CompilerMSL::build_implicit_builtins() active_input_builtins.get(BuiltInSubgroupGtMask)); bool need_multiview = get_execution_model() == ExecutionModelVertex && !msl_options.view_index_from_device_index && msl_options.multiview_layered_rendering && - (msl_options.multiview || active_input_builtins.get(BuiltInViewIndex)); + (msl_options.multiview || active_input_builtins.get(BuiltInViewIndex)) && + !msl_options.for_mesh_pipeline; bool need_dispatch_base = msl_options.dispatch_base && get_execution_model() == ExecutionModelGLCompute && (active_input_builtins.get(BuiltInWorkgroupId) || active_input_builtins.get(BuiltInGlobalInvocationId)); - bool need_grid_params = get_execution_model() == ExecutionModelVertex && msl_options.vertex_for_tessellation; + bool need_grid_params = get_execution_model() == ExecutionModelVertex && msl_options.vertex_for_tessellation && + !msl_options.for_mesh_pipeline; bool need_vertex_base_params = need_grid_params && (active_input_builtins.get(BuiltInVertexId) || active_input_builtins.get(BuiltInVertexIndex) || @@ -934,9 +952,9 @@ void CompilerMSL::build_implicit_builtins() } // If we're returning a struct from a vertex-like entry point, we must return a position attribute. - bool need_position = (get_execution_model() == ExecutionModelVertex || is_tese_shader()) && + bool need_position = (get_execution_model() == ExecutionModelVertex || is_tese_shader() || get_execution_model() == ExecutionModelGeometry) && !capture_output_to_buffer && !get_is_rasterization_disabled() && - !active_output_builtins.get(BuiltInPosition); + !active_output_builtins.get(BuiltInPosition) && !msl_options.for_mesh_pipeline; if (need_position) { @@ -1463,8 +1481,399 @@ void CompilerMSL::emit_entry_point_declarations() } } +static int vertex_count_in_primitive(CompilerMSL::Options::PrimitiveTopology top) +{ + if (top == CompilerMSL::Options::PrimitiveTopology::TriangleStrip) + return 3; + else if (top == CompilerMSL::Options::PrimitiveTopology::Triangles) + return 3; + else if (top == CompilerMSL::Options::PrimitiveTopology::Points) + return 1; + + return 0; +} + +static const char* get_vertex_loader_component_suffix(uint32_t elements) +{ + switch (elements) + { + case 1: return ""; + case 2: return "2"; + case 3: return "3"; + case 4: return "4"; + default: + SPIRV_CROSS_THROW("Invalid component count: " + std::to_string(elements)); + return "INVALID_COMPONENT_COUNT"; + } +} + +static const char* get_normalization_string(MSLShaderVariableFormat type, bool normalized) +{ + if (!normalized) + return ""; + switch (type) + { + case MSL_SHADER_VARIABLE_FORMAT_UINT16: return " * (1.f/65535.f)"; + case MSL_SHADER_VARIABLE_FORMAT_UINT8: return " * (1.f/255.f)"; + // TODO: Proper positive sint normalization + case MSL_SHADER_VARIABLE_FORMAT_INT16: return " * (1.f/32768.f)"; + case MSL_SHADER_VARIABLE_FORMAT_INT8: return " * (1.f/128.f)"; + default: return ""; + } +} + +static const char *get_variable_format_string(MSLShaderVariableFormat format) +{ + switch (format) + { + case MSL_SHADER_VARIABLE_FORMAT_INT8: return "byte"; + case MSL_SHADER_VARIABLE_FORMAT_UINT8: return "ubyte"; + case MSL_SHADER_VARIABLE_FORMAT_INT16: return "short"; + case MSL_SHADER_VARIABLE_FORMAT_UINT16: return "ushort"; + case MSL_SHADER_VARIABLE_FORMAT_INT32: return "int"; + case MSL_SHADER_VARIABLE_FORMAT_UINT32: return "uint"; + case MSL_SHADER_VARIABLE_FORMAT_HALF: return "half"; + case MSL_SHADER_VARIABLE_FORMAT_FLOAT: return "float"; + default: + SPIRV_CROSS_THROW("Format not handled: " + std::to_string(format)); + return "INVALID_TYPE"; + } +} + +int CompilerMSL::get_primitive_vertex_count() +{ + auto &execution = get_entry_point(); + + if (execution.flags.get(ExecutionModeInputPoints)) + return 1; + if (execution.flags.get(ExecutionModeInputLines)) + return 2; + if (execution.flags.get(ExecutionModeInputLinesAdjacency)) + return 2; + if (execution.flags.get(ExecutionModeTriangles)) + return 3; + if (execution.flags.get(ExecutionModeInputTrianglesAdjacency)) + return 3; + + return 1; +} + +void CompilerMSL::emit_mesh_wrapper() +{ + auto &execution = get_entry_point(); + + if (execution.model == ExecutionModelVertex) + { + auto out_var_type = get_variable_data_type(get(stage_out_var_id)); + + // Emit the payload struct + statement("struct Payload"); + begin_scope(); + statement(join(type_to_glsl(out_var_type), " vertices[", std::to_string(vertex_count_in_primitive(msl_options.input_primitive_type)), "];")); + end_scope(";"); + + // Emit struct with info about the draw call. + statement("struct DrawInfo"); + begin_scope(); + + statement("int32_t indexed;"); + statement("int32_t indexSize;"); + statement("int64_t indexBuffer;"); + + end_scope(";"); + + // Object entry point. + statement("[[object]] void ", execution.name, "(object_data Payload &payload [[payload]], mesh_grid_properties meshGridProperties, constant DrawInfo *drawInfo [[buffer(", std::to_string(get_msl_options().draw_info_index),")]],"); + + bool vertex_bindings[32] = {false}; + for (auto si: inputs_by_location) + { + if (si.second.builtin != spv::BuiltInMax) continue; + vertex_bindings[si.second.binding] = true; + } + + for (int i = 0; i < 32; ++i) + { + if (!vertex_bindings[i]) continue; + std::string binding = std::to_string(i); + statement("device uchar *vb", binding, " [[buffer(", binding, ")]],"); + } + + // Disable for_mesh_pipeline temporarily so that args get their [[attributes]]. + msl_options.for_mesh_pipeline = false; + string object_arguments; + entry_point_args_discrete_descriptors(object_arguments); + msl_options.for_mesh_pipeline = true; + + if (!object_arguments.empty()) object_arguments += ","; + statement(object_arguments); + + statement("uint3 positionInGrid [[thread_position_in_grid]])"); + + begin_scope(); + + if (msl_options.input_primitive_type == Options::PrimitiveTopology::TriangleStrip) + { + statement("int startingIndex = positionInGrid.x;"); + statement("int vertexCount = 3;"); + } + else if (msl_options.input_primitive_type == Options::PrimitiveTopology::Triangles) + { + statement("int startingIndex = positionInGrid.x * 3;"); + statement("int vertexCount = 3;"); + } + else if (msl_options.input_primitive_type == Options::PrimitiveTopology::Points) + { + statement("int startingIndex = positionInGrid.x;"); + statement("int vertexCount = 1;"); + } + else + { + SPIRV_CROSS_THROW("Input primitive type not supported"); + } + + statement("int instanceIndex = positionInGrid.y;"); + + statement("for (int i = 0; i < vertexCount; ++i)"); + begin_scope(); + statement("uint vertexIndex;"); + + statement("if (drawInfo->indexed)"); + begin_scope(); + + statement(R"END( + if (drawInfo->indexSize == 1) { + vertexIndex = ((constant uchar *)drawInfo->indexBuffer)[startingIndex + i]; + if (vertexIndex == 0xff) { + return; + } + } else if (drawInfo->indexSize == 2) { + vertexIndex = ((constant ushort *)drawInfo->indexBuffer)[startingIndex + i]; + if (vertexIndex == 0xffff) { + return; + } + } else { + vertexIndex = ((constant uint *)drawInfo->indexBuffer)[startingIndex + i]; + if (vertexIndex == 0xffffffff) { + return; + } + })END"); + + end_scope(); + statement("else vertexIndex = startingIndex + i;"); + + if (stage_in_var_id) + { + auto in_var_type = get_variable_data_type(get(stage_in_var_id)); + statement(type_to_glsl(in_var_type), " ", to_name(stage_in_var_id), ";"); + } + + ir.for_each_typed_id([&](uint32_t id, SPIRVariable &var) + { + if (var.storage != StorageClassInput) + return; + + auto &type = get(var.basetype); + if (has_decoration(type.self, DecorationBlock)) + return; + + if (!interface_variable_exists_in_entry_point(var.self)) + return; + + if (is_hidden_variable(var, true)) + return; + + if (is_builtin_variable(var)) + return; + + uint32_t location = get_decoration(id, DecorationLocation); + + LocationComponentPair key; + key.location = location; + key.component = 0; + + auto payload_it = inputs_by_location.find(key); + if (payload_it == inputs_by_location.end()) + return; + + auto variable = payload_it->second; + std::string name = to_name(id); + + SPIRType& parent_type = get(type.parent_type); + std::string parent_type_str = type_to_glsl(get(type.parent_type), var.self); + uint32_t load_elements = std::min(variable.vecsize, parent_type.vecsize); + + std::string type_name = get_variable_format_string(variable.format); + std::string component = type_name + get_vertex_loader_component_suffix(load_elements); + std::string packed = variable.vecsize > 1 ? "packed_" : ""; + std::string load_string = "*(device " + packed + component + " *)(vb" + std::to_string(variable.binding) + " + " + std::to_string(variable.offset) + " + vertexIndex * " + std::to_string(variable.stride) + ")"; + if (load_elements != parent_type.vecsize) + { + component = type_name + get_vertex_loader_component_suffix(parent_type.vecsize); + load_string = component + "(" + load_string; + for (uint32_t i = load_elements; i < parent_type.vecsize; i++) + load_string += (i == 3) ? ", 1" : ", 0"; + load_string += ")"; + } + + if (parent_type_str != component) + load_string = parent_type_str + "(" + load_string + ")"; + statement(to_name(var.self), " = ", load_string, get_normalization_string(variable.format, variable.normalized), ";"); + }); + + statement("payload.vertices[i] = ", execution.name, "("); + bool need_comma = false; + if (stage_in_var_id) + { + need_comma = true; + statement(to_name(stage_in_var_id)); + } + + auto resources = get_sorted_entry_point_args(false); + + for (auto &resource : resources) + { + statement(need_comma ? ", " : "", resource.name); + need_comma = true; + } + + ir.for_each_typed_id([&](uint32_t var_id, SPIRVariable &var) + { + if (var.storage == StorageClassInput && is_builtin_variable(var)) + { + uint32_t builtin = get_decoration(var_id, DecorationBuiltIn); + + switch (builtin) + { + case BuiltInInstanceIndex: { + statement(need_comma ? ", " : "", "instanceIndex"); + need_comma = true; + break; + } + case BuiltInBaseInstance: { + statement(need_comma ? ", " : "", "0"); + need_comma = true; + break; + } + case BuiltInVertexIndex: { + statement(need_comma ? ", " : "", "vertexIndex"); + need_comma = true; + break; + } + default: { + statement(need_comma ? ", " : "", "0 /* Unhandled builtin ", builtin, " */"); + need_comma = true; + } break; + } + } + }); + + statement(");"); + + end_scope(); + + statement("meshGridProperties.set_threadgroups_per_grid(uint3(1, 1, 1));"); + + end_scope(); + } + else + { + assert(execution.model == ExecutionModelGeometry); + + // Emit the payload struct + statement("struct Payload"); + begin_scope(); + statement("struct"); + begin_scope(); + + statement("struct "); + begin_scope(); + + auto &ib_var = get(stage_in_var_id); + auto &ib_type = get_variable_data_type(ib_var); + + msl_options.for_mesh_pipeline = false; + int i = 0; + for (auto &member : ib_type.member_types) + { + auto type = get(member); + auto parent = get(type.parent_type); + emit_struct_member(ib_type, type.parent_type, i); + i++; + } + msl_options.for_mesh_pipeline = true; + + int vertex_count = get_primitive_vertex_count(); + + end_scope_decl("in"); + end_scope_decl(join("vertices[", std::to_string(vertex_count), "]")); + end_scope(";"); + + // Mesh entry point. + + statement("[[mesh]] void ", execution.name, "(mesh_stream_t::mesh_t outputMesh, const object_data Payload &payload [[payload]],"); + + // Geometry bindings + + msl_options.for_mesh_pipeline = false; + + string mesh_arguments; + entry_point_args_discrete_descriptors(mesh_arguments); + msl_options.for_mesh_pipeline = true; + if (!mesh_arguments.empty()) mesh_arguments += ","; + statement(mesh_arguments); + + statement("uint lid [[thread_index_in_threadgroup]], uint tid [[threadgroup_position_in_grid]])"); + + begin_scope(); + + auto in_var_type = get_variable_data_type(get(stage_in_var_id)); + statement(type_to_glsl(in_var_type), " ", to_name(stage_in_var_id), ";"); + + statement("const unsigned long vertexCount = ", vertex_count,";"); + statement("for (unsigned long i = 0; i < vertexCount; ++i)"); + begin_scope(); + + statement("auto out = payload.vertices[i];"); + + ir.for_each_typed_id([&](uint32_t id, SPIRVariable &var) + { + std::string name = to_name(id); + + if (var.storage != StorageClassInput) + return; + + if (!interface_variable_exists_in_entry_point(var.self)) + return; + + if (is_hidden_variable(var, true)) + return; + + statement("if (i < sizeof(", name, ") / sizeof(", name, "[0]))"); + statement("\t", name, "[i] = out.", name, ";"); + }); + + end_scope(); + + statement(join(execution.name, "(outputMesh, ", to_name(stage_in_var_id))); + + auto resources = get_sorted_entry_point_args(false); + + for (auto &resource : resources) + statement(", ", resource.name); + + statement(");"); + end_scope(); + } +} + + string CompilerMSL::compile() { + auto &execution = get_entry_point(); + if (execution.model == ExecutionModelGeometry) + msl_options.for_mesh_pipeline = true; + replace_illegal_entry_point_names(); ir.fixup_reserved_names(); @@ -1567,6 +1976,9 @@ string CompilerMSL::compile() // Do output first to ensure out. is declared at top of entry function. qual_pos_var_name = ""; stage_out_var_id = add_interface_block(StorageClassOutput); + if (execution.model == ExecutionModelGeometry) + stage_out_mesh_primitive_var_id = add_interface_block(StorageClassOutput, false, true); + patch_stage_out_var_id = add_interface_block(StorageClassOutput, true); stage_in_var_id = add_interface_block(StorageClassInput); if (is_tese_shader()) @@ -1593,6 +2005,16 @@ string CompilerMSL::compile() // the loop, so the hooks aren't added multiple times. fix_up_shader_inputs_outputs(); + if (execution.model == ExecutionModelGeometry) + { + auto &entry_func = get(ir.default_entry_point); + + entry_func.fixup_hooks_in.push_back([=]() + { + statement("mesh_stream_t meshStream(spvMeshOut, ", to_name(stage_out_var_id),", ", to_name(stage_out_mesh_primitive_var_id) ,");"); + }); + } + // If we are using argument buffers, we create argument buffer structures for them here. // These buffers will be used in the entry point, not the individual resources. if (msl_options.argument_buffers) @@ -1622,8 +2044,45 @@ string CompilerMSL::compile() emit_custom_functions(); emit_specialization_constants_and_structs(); emit_resources(); + + if (execution.model == ExecutionModelGeometry) + { + auto output_primitives = execution.output_primitives; + auto output_vertices = execution.output_vertices; + + auto vertex_type = type_to_glsl(get_variable_data_type(get(stage_out_var_id))); + auto prim_type = type_to_glsl(get_variable_data_type(get(stage_out_mesh_primitive_var_id))); + + string output_topology = ""; + + if (execution.flags.get(ExecutionModeOutputPoints)) + { + output_topology = "point"; + if (!output_vertices) output_vertices = 1; + if (!output_primitives) output_primitives = output_vertices; + } + else if (execution.flags.get(ExecutionModeOutputLineStrip)) + { + output_topology = "line"; + if (!output_vertices) output_vertices = 2; + if (!output_primitives) output_primitives = output_vertices - 1; + } + else if (execution.flags.get(ExecutionModeOutputTriangleStrip)) + { + output_topology = "triangle"; + if (!output_vertices) output_vertices = 3; + if (!output_primitives) output_primitives = output_vertices - 2; + } + + statement("enum { VERTEX_COUNT = ", std::to_string(output_vertices), ", PRIMITIVE_COUNT = ", std::to_string(output_primitives), " };"); + statement("using mesh_stream_t = spvMeshStream<", vertex_type, ", ", prim_type, ", VERTEX_COUNT, PRIMITIVE_COUNT, metal::topology::", output_topology, ">;"); + } + emit_function(get(ir.default_entry_point), Bitset()); + if (msl_options.for_mesh_pipeline) + emit_mesh_wrapper(); + pass_count++; } while (is_forcing_recompilation()); @@ -3391,7 +3850,8 @@ bool CompilerMSL::variable_storage_requires_stage_io(spv::StorageClass storage) return !capture_output_to_buffer; else if (storage == StorageClassInput) return !(is_tesc_shader() && msl_options.multi_patch_workgroup) && - !(is_tese_shader() && msl_options.raw_buffer_tese_input); + !(is_tese_shader() && msl_options.raw_buffer_tese_input) && + get_execution_model() != ExecutionModelGeometry; else return false; } @@ -3686,7 +4146,7 @@ void CompilerMSL::add_variable_to_interface_block(StorageClass storage, const st is_builtin = false; // MSL does not allow matrices or arrays in input or output variables, so need to handle it specially. - if ((!is_builtin || attribute_load_store) && storage_is_stage_io && is_composite_type) + if (!msl_options.for_mesh_pipeline && (!is_builtin || attribute_load_store) && storage_is_stage_io && is_composite_type) { add_composite_variable_to_interface_block(storage, ib_var_ref, ib_type, var, meta); } @@ -3746,11 +4206,11 @@ void CompilerMSL::fix_up_interface_member_indices(StorageClass storage, uint32_t // Add an interface structure for the type of storage, which is either StorageClassInput or StorageClassOutput. // Returns the ID of the newly added variable, or zero if no variable was added. -uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch) +uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch, bool mesh_primitive) { // Accumulate the variables that should appear in the interface struct. SmallVector vars; - bool incl_builtins = storage == StorageClassOutput || is_tessellation_shader(); + bool incl_builtins = storage == StorageClassOutput || is_tessellation_shader() || get_execution_model() == ExecutionModelGeometry; bool has_seen_barycentric = false; InterfaceBlockMeta meta; @@ -3792,6 +4252,14 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch) bi_type == BuiltInFragDepth || bi_type == BuiltInFragStencilRefEXT || bi_type == BuiltInSampleMask; + if (get_execution_model() == ExecutionModelGeometry) + { + if (mesh_primitive && (!is_builtin || !builtin_is_per_primitive_mesh_output(bi_type))) + return; + + if (!mesh_primitive && is_builtin && builtin_is_per_primitive_mesh_output(bi_type)) return; + } + // These builtins are part of the stage in/out structs. bool is_interface_block_builtin = builtin_is_stage_in_out || (is_tese_shader() && !msl_options.raw_buffer_tese_input && @@ -3912,7 +4380,7 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch) // If no variables qualify, leave. // For patch input in a tessellation evaluation shader, the per-vertex stage inputs // are included in a special patch control point array. - if (vars.empty() && + if (vars.empty() && !mesh_primitive && !(!msl_options.raw_buffer_tese_input && storage == StorageClassInput && patch && stage_in_var_id)) return 0; @@ -4014,7 +4482,10 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch) { auto &blk = get(blk_id); if (blk.terminator == SPIRBlock::Return || (blk.terminator == SPIRBlock::Kill && blk_id == entry_func.blocks.back())) - blk.return_value = rtn_id; + { + if (get_execution_model() != ExecutionModelGeometry) // Geometry shaders don't return the output structure, it's emitted to the mesh stream. + blk.return_value = rtn_id; + } } vars_needing_early_declaration.push_back(ib_var_id); } @@ -4154,11 +4625,12 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch) switch (input.second.format) { case MSL_SHADER_VARIABLE_FORMAT_UINT16: + case MSL_SHADER_VARIABLE_FORMAT_INT16: + case MSL_SHADER_VARIABLE_FORMAT_HALF: case MSL_SHADER_VARIABLE_FORMAT_ANY16: type.basetype = SPIRType::UShort; type.width = 16; break; - case MSL_SHADER_VARIABLE_FORMAT_ANY32: default: type.basetype = SPIRType::UInt; type.width = 32; @@ -4212,11 +4684,12 @@ uint32_t CompilerMSL::add_interface_block(StorageClass storage, bool patch) switch (output.second.format) { case MSL_SHADER_VARIABLE_FORMAT_UINT16: + case MSL_SHADER_VARIABLE_FORMAT_INT16: + case MSL_SHADER_VARIABLE_FORMAT_HALF: case MSL_SHADER_VARIABLE_FORMAT_ANY16: type.basetype = SPIRType::UShort; type.width = 16; break; - case MSL_SHADER_VARIABLE_FORMAT_ANY32: default: type.basetype = SPIRType::UInt; type.width = 32; @@ -4471,6 +4944,26 @@ uint32_t CompilerMSL::ensure_correct_input_type(uint32_t type_id, uint32_t locat } } + case MSL_SHADER_VARIABLE_FORMAT_UINT32: + { + switch (type.basetype) + { + case SPIRType::UShort: + case SPIRType::UInt: + if (num_components > type.vecsize) + return build_extended_vector_type(type_id, num_components); + else + return type_id; + + case SPIRType::Int: + return build_extended_vector_type(type_id, num_components > type.vecsize ? num_components : type.vecsize, + SPIRType::UInt); + + default: + SPIRV_CROSS_THROW("Vertex attribute type mismatch between host and shader"); + } + } + default: if (num_components > type.vecsize) type_id = build_extended_vector_type(type_id, num_components); @@ -5237,6 +5730,9 @@ void CompilerMSL::emit_header() statement("#include "); statement("#include "); + if (msl_options.for_mesh_pipeline) + statement("#include "); + for (auto &header : header_lines) statement(header); @@ -5433,6 +5929,67 @@ void CompilerMSL::emit_custom_templates() statement(""); break; + case SPVFuncImplEmitVertex: + statement("template"); + statement("struct spvMeshStream"); + begin_scope(); + + statement("using mesh_t = metal::mesh;"); + statement("thread mesh_t &meshOut;"); + + statement("int currentVertex = 0;"); + statement("int currentIndex = 0;"); + statement("int currentVertexInPrimitive = 0;"); + statement("int currentPrimitive = 0;"); + statement("thread P &primitiveData;"); + statement("thread V &vertexData;"); + + statement("spvMeshStream(thread mesh_t &_meshOut, thread V &_v, thread P &_p) : meshOut(_meshOut), primitiveData(_p), vertexData(_v)"); + begin_scope(); + end_scope(); + + statement("~spvMeshStream()"); + begin_scope(); + statement("meshOut.set_primitive_count(currentPrimitive);"); + end_scope(); + + statement("int VperP()"); + begin_scope(); + statement("if (T == metal::topology::triangle) return 3;"); + statement("else if (T == metal::topology::line) return 2;"); + statement("else /* if (T == metal::topology::point) */ return 1;"); + end_scope(); + + statement("void EndPrimitive()"); + begin_scope(); + statement("currentVertexInPrimitive = 0;"); + end_scope(); + + statement("void EmitVertex()"); + begin_scope(); + if (options.vertex.flip_vert_y) + { + statement("V v = vertexData;"); + statement("v.gl_Position.y = -v.gl_Position.y; // Invert Y-axis for Metal"); + statement("meshOut.set_vertex(currentVertex++, v);"); + } + else + { + statement("meshOut.set_vertex(currentVertex++, vertexData);"); + } + statement("currentVertexInPrimitive++;"); + statement("if (currentVertexInPrimitive >= VperP())"); + begin_scope(); + statement("if (T == metal::topology::triangle) meshOut.set_index(currentIndex++, currentVertex-3);"); + statement("if (T == metal::topology::triangle || T == metal::topology::line) meshOut.set_index(currentIndex++, currentVertex-2);"); + statement("meshOut.set_index(currentIndex++, currentVertex-1);"); + statement("meshOut.set_primitive(currentPrimitive++, primitiveData);"); + end_scope(); + end_scope(); + end_scope(";"); + + break; + default: break; } @@ -7411,6 +7968,8 @@ void CompilerMSL::emit_resources() // Emit the special [[stage_in]] and [[stage_out]] interface blocks which we created. emit_interface_block(stage_out_var_id); + if (stage_out_mesh_primitive_var_id) + emit_interface_block(stage_out_mesh_primitive_var_id); emit_interface_block(patch_stage_out_var_id); emit_interface_block(stage_in_var_id); emit_interface_block(patch_stage_in_var_id); @@ -9479,6 +10038,20 @@ void CompilerMSL::emit_instruction(const Instruction &instruction) break; } + case OpEmitVertex: + { + add_spv_func_and_recompile(SPVFuncImplEmitVertex); + statement("meshStream.EmitVertex();"); + break; + } + + case OpEndPrimitive: + { + add_spv_func_and_recompile(SPVFuncImplEmitVertex); + statement("meshStream.EndPrimitive();"); + break; + } + default: CompilerGLSL::emit_instruction(instruction); break; @@ -10420,6 +10993,14 @@ void CompilerMSL::emit_function_prototype(SPIRFunction &func, const Bitset &) if (processing_entry_point) { + auto &execution = get_entry_point(); + if (execution.model == ExecutionModelGeometry) + { + auto output_primitives = execution.output_primitives; + if (!output_primitives) output_primitives = execution.output_vertices - 2; + decl += "mesh_stream_t::mesh_t spvMeshOut, "; + } + if (msl_options.argument_buffers) decl += entry_point_args_argument_buffer(!func.arguments.empty()); else @@ -11930,8 +12511,9 @@ string CompilerMSL::member_attribute_qualifier(const SPIRType &type, uint32_t in case BuiltInInstanceId: case BuiltInInstanceIndex: case BuiltInBaseInstance: - if (msl_options.vertex_for_tessellation) + if (msl_options.vertex_for_tessellation || msl_options.for_mesh_pipeline) return ""; + return string(" [[") + builtin_qualifier(builtin) + "]]"; case BuiltInDrawIndex: @@ -11949,11 +12531,36 @@ string CompilerMSL::member_attribute_qualifier(const SPIRType &type, uint32_t in locn = get_member_location(type.self, index); if (locn != k_unknown_location) + { + if (msl_options.for_mesh_pipeline) + return ""; return string(" [[attribute(") + convert_to_string(locn) + ")]]"; + } + + if (msl_options.for_mesh_pipeline) + return ""; + } + + if (execution.model == ExecutionModelGeometry && type.storage == StorageClassInput && !msl_options.for_mesh_pipeline) // @Cleanup: different condition than for_mesh_pipeline? + { + if (is_builtin) + { + switch (builtin) + { + case BuiltInPosition: + case BuiltInLayer: + return string(" [[") + builtin_qualifier(builtin) + "]]" + (mbr_type.array.empty() ? "" : " "); + default: break; + } + } + + string loc_qual = member_location_attribute_qualifier(type, index); + if (!loc_qual.empty()) + return join(" [[", loc_qual, "]]"); } // Vertex and tessellation evaluation function outputs - if (((execution.model == ExecutionModelVertex && !msl_options.vertex_for_tessellation) || is_tese_shader()) && + if (((execution.model == ExecutionModelVertex && !msl_options.vertex_for_tessellation) || is_tese_shader() || (execution.model == ExecutionModelGeometry)) && type.storage == StorageClassOutput) { if (is_builtin) @@ -12392,6 +12999,10 @@ uint32_t CompilerMSL::get_or_allocate_builtin_output_member_location(spv::BuiltI // entry type if the current function is the entry point function string CompilerMSL::func_type_decl(SPIRType &type) { + auto &execution = get_entry_point(); + if (execution.model == ExecutionModelGeometry) + return "void"; + // The regular function return type. If not processing the entry point function, that's all we need string return_type = type_to_glsl(type) + type_to_array_glsl(type); if (!processing_entry_point) @@ -12404,10 +13015,16 @@ string CompilerMSL::func_type_decl(SPIRType &type) // Prepend a entry type, based on the execution model string entry_type; - auto &execution = get_entry_point(); switch (execution.model) { case ExecutionModelVertex: + if (msl_options.for_mesh_pipeline) + { + if (!msl_options.supports_msl_version(3, 0)) + SPIRV_CROSS_THROW("Mesh pipelines require MSL 3.0."); + entry_type = ""; + break; + } if (msl_options.vertex_for_tessellation && !msl_options.supports_msl_version(1, 2)) SPIRV_CROSS_THROW("Tessellation requires Metal 1.2."); entry_type = msl_options.vertex_for_tessellation ? "kernel" : "vertex"; @@ -12441,6 +13058,9 @@ string CompilerMSL::func_type_decl(SPIRType &type) break; } + if (entry_type.empty()) + return return_type; + return entry_type + " " + return_type; } @@ -12621,7 +13241,9 @@ string CompilerMSL::entry_point_arg_stage_in() auto &type = get_variable_data_type(var); add_resource_name(var.self); - decl = join(type_to_glsl(type), " ", to_name(var.self), " [[stage_in]]"); + decl = join(type_to_glsl(type), " ", to_name(var.self)); + if (!msl_options.for_mesh_pipeline) + decl += " [[stage_in]]"; } return decl; @@ -12640,7 +13262,8 @@ bool CompilerMSL::is_direct_input_builtin(BuiltIn bi_type) case BuiltInInstanceId: case BuiltInInstanceIndex: case BuiltInBaseInstance: - return get_execution_model() != ExecutionModelVertex || !msl_options.vertex_for_tessellation; + return get_execution_model() != ExecutionModelVertex || !msl_options.vertex_for_tessellation || + msl_options.for_mesh_pipeline; // Tess. control function in case BuiltInPosition: case BuiltInPointSize: @@ -12741,16 +13364,19 @@ void CompilerMSL::entry_point_args_builtin(string &ep_args) else ep_args += builtin_type_decl(bi_type, var_id) + " " + to_expression(var_id); - ep_args += string(" [[") + builtin_qualifier(bi_type); - if (bi_type == BuiltInSampleMask && get_entry_point().flags.get(ExecutionModePostDepthCoverage)) + if (!msl_options.for_mesh_pipeline) { - if (!msl_options.supports_msl_version(2)) - SPIRV_CROSS_THROW("Post-depth coverage requires MSL 2.0."); - if (msl_options.is_macos() && !msl_options.supports_msl_version(2, 3)) - SPIRV_CROSS_THROW("Post-depth coverage on Mac requires MSL 2.3."); - ep_args += ", post_depth_coverage"; + ep_args += string(" [[") + builtin_qualifier(bi_type); + if (bi_type == BuiltInSampleMask && get_entry_point().flags.get(ExecutionModePostDepthCoverage)) + { + if (!msl_options.supports_msl_version(2)) + SPIRV_CROSS_THROW("Post-depth coverage requires MSL 2.0."); + if (msl_options.is_macos() && !msl_options.supports_msl_version(2, 3)) + SPIRV_CROSS_THROW("Post-depth coverage on Mac requires MSL 2.3."); + ep_args += ", post_depth_coverage"; + } + ep_args += "]]"; } - ep_args += "]]"; builtin_declaration = false; } } @@ -13003,7 +13629,8 @@ string CompilerMSL::entry_point_args_argument_buffer(bool append_comma) claimed_bindings.set(buffer_binding); ep_args += get_argument_address_space(var) + " " + type_to_glsl(type) + "& " + to_restrict(id, true) + to_name(id); - ep_args += " [[buffer(" + convert_to_string(buffer_binding) + ")]]"; + if (!msl_options.for_mesh_pipeline) + ep_args += " [[buffer(" + convert_to_string(buffer_binding) + ")]]"; next_metal_resource_index_buffer = max(next_metal_resource_index_buffer, buffer_binding + 1); } @@ -13039,23 +13666,12 @@ const MSLConstexprSampler *CompilerMSL::find_constexpr_sampler(uint32_t id) cons return nullptr; } -void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) +SmallVector CompilerMSL::get_sorted_entry_point_args(bool add_names) { // Output resources, sorted by resource index & type // We need to sort to work around a bug on macOS 10.13 with NVidia drivers where switching between shaders // with different order of buffers can result in issues with buffer assignments inside the driver. - struct Resource - { - SPIRVariable *var; - SPIRVariable *descriptor_alias; - string name; - SPIRType::BaseType basetype; - uint32_t index; - uint32_t plane; - uint32_t secondary_index; - }; - - SmallVector resources; + SmallVector resources; entry_point_bindings.clear(); ir.for_each_typed_id([&](uint32_t var_id, SPIRVariable &var) { @@ -13092,7 +13708,7 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) // and it's being used as an alias (so we can emit void* instead). resource.descriptor_alias = resource.var; // Need to promote interlocked usage so that the primary declaration is correct. - if (interlocked_resources.count(var_id)) + if (add_names && interlocked_resources.count(var_id)) interlocked_resources.insert(resource.var->self); break; } @@ -13119,7 +13735,7 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) if (type.basetype == SPIRType::SampledImage) { - add_resource_name(var_id); + if (add_names) add_resource_name(var_id); uint32_t plane_count = 1; if (constexpr_sampler && constexpr_sampler->ycbcr_conversion_enable) @@ -13139,7 +13755,7 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) else if (!constexpr_sampler) { // constexpr samplers are not declared as resources. - add_resource_name(var_id); + if (add_names) add_resource_name(var_id); // Don't allocate resource indices for aliases. uint32_t resource_index = ~0u; @@ -13154,9 +13770,16 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) }); stable_sort(resources.begin(), resources.end(), - [](const Resource &lhs, const Resource &rhs) + [](const Entry_Point_Resource &lhs, const Entry_Point_Resource &rhs) { return tie(lhs.basetype, lhs.index) < tie(rhs.basetype, rhs.index); }); + return resources; +} + +void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) +{ + auto resources = get_sorted_entry_point_args(); + for (auto &r : resources) { auto &var = *r.var; @@ -13223,10 +13846,13 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) type_to_glsl(type) + "*>* "; } ep_args += to_restrict(var_id, true) + r.name + "_"; - ep_args += " [[buffer(" + convert_to_string(r.index) + ")"; - if (interlocked_resources.count(var_id)) - ep_args += ", raster_order_group(0)"; - ep_args += "]]"; + if (!msl_options.for_mesh_pipeline) + { + ep_args += " [[buffer(" + convert_to_string(r.index) + ")"; + if (interlocked_resources.count(var_id)) + ep_args += ", raster_order_group(0)"; + ep_args += "]]"; + } } else { @@ -13250,10 +13876,13 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) ep_args += ", "; ep_args += get_argument_address_space(var) + " " + type_to_glsl(type) + "& " + to_restrict(var_id, true) + r.name; - ep_args += " [[buffer(" + convert_to_string(r.index) + ")"; - if (interlocked_resources.count(var_id)) - ep_args += ", raster_order_group(0)"; - ep_args += "]]"; + if (!msl_options.for_mesh_pipeline) + { + ep_args += " [[buffer(" + convert_to_string(r.index) + ")"; + if (interlocked_resources.count(var_id)) + ep_args += ", raster_order_group(0)"; + ep_args += "]]"; + } } break; } @@ -13261,10 +13890,17 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) if (!ep_args.empty()) ep_args += ", "; ep_args += sampler_type(type, var_id) + " " + r.name; + if (is_runtime_size_array(type)) - ep_args += "_ [[buffer(" + convert_to_string(r.index) + ")]]"; - else - ep_args += " [[sampler(" + convert_to_string(r.index) + ")]]"; + ep_args += "_"; + + if (!msl_options.for_mesh_pipeline) + { + if (is_runtime_size_array(type)) + ep_args += " [[buffer(" + convert_to_string(r.index) + ")]]"; + else + ep_args += " [[sampler(" + convert_to_string(r.index) + ")]]"; + } break; case SPIRType::Image: { @@ -13280,13 +13916,19 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) ep_args += join(plane_name_suffix, r.plane); if (is_runtime_size_array(type)) - ep_args += "_ [[buffer(" + convert_to_string(r.index) + ")"; - else - ep_args += " [[texture(" + convert_to_string(r.index) + ")"; + ep_args += "_"; - if (interlocked_resources.count(var_id)) - ep_args += ", raster_order_group(0)"; - ep_args += "]]"; + if (!msl_options.for_mesh_pipeline) + { + if (is_runtime_size_array(type)) + ep_args += " [[buffer(" + convert_to_string(r.index) + ")"; + else + ep_args += " [[texture(" + convert_to_string(r.index) + ")"; + + if (interlocked_resources.count(var_id)) + ep_args += ", raster_order_group(0)"; + ep_args += "]]"; + } } else { @@ -13301,10 +13943,13 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) { ep_args += ", device atomic_" + type_to_glsl(get(basetype.image.type), 0); ep_args += "* " + r.name + "_atomic"; - ep_args += " [[buffer(" + convert_to_string(r.secondary_index) + ")"; - if (interlocked_resources.count(var_id)) - ep_args += ", raster_order_group(0)"; - ep_args += "]]"; + if (!msl_options.for_mesh_pipeline) + { + ep_args += " [[buffer(" + convert_to_string(r.secondary_index) + ")"; + if (interlocked_resources.count(var_id)) + ep_args += ", raster_order_group(0)"; + ep_args += "]]"; + } } break; } @@ -13333,10 +13978,13 @@ void CompilerMSL::entry_point_args_discrete_descriptors(string &ep_args) type_to_glsl(type, var_id) + "& " + r.name; else ep_args += type_to_glsl(type, var_id) + " " + r.name; - ep_args += " [[buffer(" + convert_to_string(r.index) + ")"; - if (interlocked_resources.count(var_id)) - ep_args += ", raster_order_group(0)"; - ep_args += "]]"; + if (!msl_options.for_mesh_pipeline) + { + ep_args += " [[buffer(" + convert_to_string(r.index) + ")"; + if (interlocked_resources.count(var_id)) + ep_args += ", raster_order_group(0)"; + ep_args += "]]"; + } break; } } @@ -15817,7 +16465,21 @@ string CompilerMSL::builtin_to_glsl(BuiltIn builtin, StorageClass storage) break; if (storage != StorageClassInput && current_function && (current_function->self == ir.default_entry_point) && !is_stage_output_builtin_masked(builtin)) + { + if (builtin_is_per_primitive_mesh_output(builtin) && get_execution_model() == ExecutionModelGeometry) + return stage_out_mesh_primitive_var_name + "." + CompilerGLSL::builtin_to_glsl(builtin, storage); + return stage_out_var_name + "." + CompilerGLSL::builtin_to_glsl(builtin, storage); + } + + if (current_function && get_execution_model() == ExecutionModelGeometry) + { + if (storage == StorageClassInput) + return stage_in_var_name + "." + CompilerGLSL::builtin_to_glsl(builtin, storage); + else if (storage == StorageClassGeneric) + return CompilerGLSL::builtin_to_glsl(builtin, storage); + } + break; case BuiltInSampleMask: @@ -16060,7 +16722,8 @@ string CompilerMSL::builtin_qualifier(BuiltIn builtin) } else if (execution.model == ExecutionModelKernel || execution.model == ExecutionModelGLCompute || execution.model == ExecutionModelTessellationControl || - (execution.model == ExecutionModelVertex && msl_options.vertex_for_tessellation)) + (execution.model == ExecutionModelVertex && + (msl_options.vertex_for_tessellation || msl_options.for_mesh_pipeline))) { // We are generating a Metal kernel function. if (!msl_options.supports_msl_version(2)) diff --git a/spirv_msl.hpp b/spirv_msl.hpp index 26167f673..6afb300bc 100644 --- a/spirv_msl.hpp +++ b/spirv_msl.hpp @@ -42,8 +42,15 @@ enum MSLShaderVariableFormat MSL_SHADER_VARIABLE_FORMAT_OTHER = 0, MSL_SHADER_VARIABLE_FORMAT_UINT8 = 1, MSL_SHADER_VARIABLE_FORMAT_UINT16 = 2, - MSL_SHADER_VARIABLE_FORMAT_ANY16 = 3, - MSL_SHADER_VARIABLE_FORMAT_ANY32 = 4, + MSL_SHADER_VARIABLE_FORMAT_UINT32 = 3, + MSL_SHADER_VARIABLE_FORMAT_FLOAT = 4, + MSL_SHADER_VARIABLE_FORMAT_INT8 = 5, + MSL_SHADER_VARIABLE_FORMAT_INT16 = 6, + MSL_SHADER_VARIABLE_FORMAT_INT32 = 7, + MSL_SHADER_VARIABLE_FORMAT_HALF = 8, + + MSL_SHADER_VARIABLE_FORMAT_ANY16 = 9, + MSL_SHADER_VARIABLE_FORMAT_ANY32 = 10, // Deprecated aliases. MSL_VERTEX_FORMAT_OTHER = MSL_SHADER_VARIABLE_FORMAT_OTHER, @@ -81,6 +88,10 @@ struct MSLShaderInterfaceVariable spv::BuiltIn builtin = spv::BuiltInMax; uint32_t vecsize = 0; MSLShaderVariableRate rate = MSL_SHADER_VARIABLE_RATE_PER_VERTEX; + uint32_t offset = 0; + uint32_t stride = 0; + uint32_t binding = 0; + bool normalized = false; }; // Matches the binding index of a MSL resource for a binding within a descriptor set. @@ -319,6 +330,7 @@ class CompilerMSL : public CompilerGLSL uint32_t shader_input_buffer_index = 22; uint32_t shader_index_buffer_index = 21; uint32_t shader_patch_input_buffer_index = 20; + uint32_t draw_info_index = 20; uint32_t shader_input_wg_index = 0; uint32_t device_index = 0; uint32_t enable_frag_output_mask = 0xffffffff; @@ -505,6 +517,15 @@ class CompilerMSL : public CompilerGLSL // Note: Only Apple's GPU compiler takes advantage of the lack of coherency, so make sure to test on Apple GPUs if you disable this. bool readwrite_texture_fences = true; + // Compile for use with a geometry shader. If set, vertex shaders will be compiled as [[object]] + // functions, and geometry shaders as [[mesh]]. + bool for_mesh_pipeline = false; + + enum class PrimitiveTopology + { + Triangles, TriangleStrip, Points + } input_primitive_type; + bool is_ios() const { return platform == iOS; @@ -808,6 +829,7 @@ class CompilerMSL : public CompilerGLSL SPVFuncImplVariableDescriptor, SPVFuncImplVariableSizedDescriptor, SPVFuncImplVariableDescriptorArray, + SPVFuncImplEmitVertex, }; // If the underlying resource has been used for comparison then duplicate loads of that resource must be too @@ -902,7 +924,7 @@ class CompilerMSL : public CompilerGLSL void extract_global_variables_from_function(uint32_t func_id, std::set &added_arg_ids, std::unordered_set &global_var_ids, std::unordered_set &processed_func_ids); - uint32_t add_interface_block(spv::StorageClass storage, bool patch = false); + uint32_t add_interface_block(spv::StorageClass storage, bool patch = false, bool mesh_primitive = false); uint32_t add_interface_block_pointer(uint32_t ib_var_id, spv::StorageClass storage); struct InterfaceBlockMeta @@ -964,6 +986,7 @@ class CompilerMSL : public CompilerGLSL void emit_interface_block(uint32_t ib_var_id); bool maybe_emit_array_assignment(uint32_t id_lhs, uint32_t id_rhs); uint32_t get_resource_array_size(uint32_t id) const; + void emit_mesh_wrapper(); void fix_up_shader_inputs_outputs(); @@ -973,6 +996,19 @@ class CompilerMSL : public CompilerGLSL std::string entry_point_arg_stage_in(); void entry_point_args_builtin(std::string &args); void entry_point_args_discrete_descriptors(std::string &args); + + struct Entry_Point_Resource + { + SPIRVariable *var; + SPIRVariable *descriptor_alias; + std::string name; + SPIRType::BaseType basetype; + uint32_t index; + uint32_t plane; + uint32_t secondary_index; + }; + + SmallVector get_sorted_entry_point_args(bool add_name = true); std::string append_member_name(const std::string &qualifier, const SPIRType &type, uint32_t index); std::string ensure_valid_name(std::string name, std::string pfx); std::string to_sampler_expression(uint32_t id); @@ -1101,6 +1137,7 @@ class CompilerMSL : public CompilerGLSL void ensure_builtin(spv::StorageClass storage, spv::BuiltIn builtin); void mark_implicit_builtin(spv::StorageClass storage, spv::BuiltIn builtin, uint32_t id); + int get_primitive_vertex_count(); std::string convert_to_f32(const std::string &expr, uint32_t components); @@ -1140,6 +1177,7 @@ class CompilerMSL : public CompilerGLSL VariableID tess_level_inner_var_id = 0; VariableID tess_level_outer_var_id = 0; VariableID stage_out_masked_builtin_type_id = 0; + VariableID stage_out_mesh_primitive_var_id = 0; // Handle HLSL-style 0-based vertex/instance index. enum class TriState @@ -1169,6 +1207,7 @@ class CompilerMSL : public CompilerGLSL std::string qual_pos_var_name; std::string stage_in_var_name = "in"; std::string stage_out_var_name = "out"; + std::string stage_out_mesh_primitive_var_name = "out_1"; std::string patch_stage_in_var_name = "patchIn"; std::string patch_stage_out_var_name = "patchOut"; std::string sampler_name_suffix = "Smplr"; diff --git a/test_shaders.py b/test_shaders.py index 887cb5b73..197450f7c 100755 --- a/test_shaders.py +++ b/test_shaders.py @@ -124,6 +124,8 @@ def msl_compiler_supports_version(version): def path_to_msl_standard(shader): if '.ios.' in shader: + if '.msl31.' in shader: + return '-std=metal3.1' if '.msl3.' in shader: return '-std=metal3.0' elif '.msl2.' in shader: @@ -143,6 +145,8 @@ def path_to_msl_standard(shader): else: return '-std=ios-metal1.2' else: + if '.msl31.' in shader: + return '-std=metal3.1' if '.msl3.' in shader: return '-std=metal3.0' elif '.msl2.' in shader: @@ -161,6 +165,8 @@ def path_to_msl_standard(shader): return '-std=macos-metal1.2' def path_to_msl_standard_cli(shader): + if '.msl31.' in shader: + return '30100' if '.msl3.' in shader: return '30000' elif '.msl2.' in shader: @@ -334,6 +340,26 @@ def cross_compile_msl(shader, spirv, opt, iterations, paths): msl_args.append('--msl-raw-buffer-tese-input') if '.for-tess.' in shader: msl_args.append('--msl-vertex-for-tessellation') + if '.for-mesh.' in shader: + msl_args.append('--msl-for-mesh-pipeline') + # Need to specify some attributes. + msl_args.append('--msl-shader-attribute') + msl_args.append('0') + msl_args.append('float') + msl_args.append('4') + + msl_args.append('0') + msl_args.append('16') + msl_args.append('30') + + msl_args.append('--msl-shader-attribute') + msl_args.append('1') + msl_args.append('float') + msl_args.append('3') + + msl_args.append('0') + msl_args.append('12') + msl_args.append('29') if '.fixed-sample-mask.' in shader: msl_args.append('--msl-additional-fixed-sample-mask') msl_args.append('0x00000022') @@ -874,7 +900,7 @@ def test_shader_msl(stats, shader, args, paths): # used as input to an invocation of spirv-cross to debug from Xcode directly. # To do so, build spriv-cross using `make DEBUG=1`, then run the spriv-cross # executable from Xcode using args: `--msl --entry main --output msl_path spirv_path`. -# print('SPRIV shader: ' + spirv) + print('SPRIV shader: ' + spirv) shader_is_msl22 = 'msl22' in joined_path shader_is_msl23 = 'msl23' in joined_path @@ -1033,12 +1059,14 @@ def main(): args.msl23 = False args.msl24 = False args.msl30 = False + args.msl31 = False if args.msl: print_msl_compiler_version() args.msl22 = msl_compiler_supports_version('-std=macos-metal2.2') args.msl23 = msl_compiler_supports_version('-std=macos-metal2.3') args.msl24 = msl_compiler_supports_version('-std=macos-metal2.4') args.msl30 = msl_compiler_supports_version('-std=metal3.0') + args.msl31 = msl_compiler_supports_version('-std=metal3.1') backend = 'glsl' if (args.msl or args.metal):