Metal-specific functionalities

This chapter provides information for Metal-specific functionalities and behaviors in Slang.

Entry Point Parameter Handling

Slang performs several transformations on entry point parameters when targeting Metal:

• Struct parameters are flattened to eliminate nested structures
• Input parameters with varying inputs are packed into a single struct
• System value semantics are translated to Metal attributes
• Parameters without semantics are given automatic attribute indices

System-Value semantics

The system-value semantics are translated to the following Metal attributes:

SV semantic name	Metal attribute
SV_Position	[[position]]
SV_Coverage	[[sample_mask]]
SV_Depth	[[depth(any)]]
SV_DepthGreaterEqual	[[depth(greater)]]
SV_DepthLessEqual	[[depth(less)]]
SV_DispatchThreadID	[[thread_position_in_grid]]
SV_GroupID	[[threadgroup_position_in_grid]]
SV_GroupThreadID	[[thread_position_in_threadgroup]]
SV_GroupIndex	Calculated from SV_GroupThreadID and group extents
SV_InstanceID	[[instance_id]]
SV_IsFrontFace	[[front_facing]]
SV_PrimitiveID	[[primitive_id]]
SV_RenderTargetArrayIndex	[[render_target_array_index]]
SV_SampleIndex	[[sample_id]]
SV_Target<N>	[[color(N)]]
SV_VertexID	[[vertex_id]]
SV_ViewportArrayIndex	[[viewport_array_index]]
SV_StartVertexLocation	[[base_vertex]]
SV_StartInstanceLocation	[[base_instance]]

Custom semantics are mapped to user attributes:

• [[user(SEMANTIC_NAME)]] For non-system value semantics
• [[user(SEMANTIC_NAME_INDEX)]] When semantic has an index

Interpolation Modifiers

Slang maps interpolation modifiers to Metal’s interpolation attributes:

Slang Interpolation	Metal Attribute
nointerpolation	[[flat]]
noperspective	[[center_no_perspective]]
linear	[[sample_no_perspective]]
sample	[[sample_perspective]]
centroid	[[center_perspective]]

Resource Types

Resource types are translated with appropriate Metal qualifiers:

Slang Type	Metal Translation
Texture2D	texture2d
RWTexture2D	texture2d
ByteAddressBuffer	uint32_t device*
StructuredBuffer<T>	device* T
ConstantBuffer<T>	constant* T

Slang Type	Metal Translation
Texture1D	texture1d
Texture1DArray	texture1d_array
RWTexture1D	texture1d
RWTexture1DArray	texture1d_array
Texture2D	texture2d
Texture2DArray	texture2d_array
RWTexture2D	texture2d
RWTexture2DArray	texture2d_array
Texture3D	texture3d
RWTexture3D	texture3d
TextureCube	texturecube
TextureCubeArray	texturecube_array
Buffer<T>	device* T
RWBuffer<T>	device* T
ByteAddressBuffer	device* uint32_t
RWByteAddressBuffer	device* uint32_t
StructuredBuffer<T>	device* T
RWStructuredBuffer<T>	device* T
AppendStructuredBuffer<T>	device* T
ConsumeStructuredBuffer<T>	device* T
ConstantBuffer<T>	constant* T
SamplerState	sampler
SamplerComparisonState	sampler
RaytracingAccelerationStructure	(Not supported)
RasterizerOrderedTexture2D	texture2d [[raster_order_group(0)]]
RasterizerOrderedBuffer<T>	device* T [[raster_order_group(0)]]

Raster-ordered access resources receive the [[raster_order_group(0)]] attribute, for example texture2d<float, access::read_write> tex [[raster_order_group(0)]].

Array Types

Array types in Metal are declared using the array template:

Slang Type	Metal Translation
ElementType[Size]	array<ElementType, Size>

Matrix Layout

Metal exclusively uses column-major matrix layout. Slang automatically handles the translation of matrix operations to maintain correct semantics:

• Matrix multiplication is transformed to account for layout differences
• Matrix types are declared as matrix<T, Columns, Rows>, for example float3x4 is represented as matrix<float, 3, 4>

Mesh Shader Support

Mesh shaders can be targeted using the following types and syntax. The same as task/mesh shaders generally in Slang.

[outputtopology("triangle")]
[numthreads(12, 1, 1)]
void meshMain(
    in uint tig: SV_GroupIndex,
    in payload MeshPayload meshPayload,
    OutputVertices<Vertex, MAX_VERTS> verts,
    OutputIndices<uint3, MAX_PRIMS> triangles,
    OutputPrimitives<Primitive, MAX_PRIMS> primitives
    )

Header Inclusions and Namespace

When targeting Metal, Slang automatically includes the following headers, these are available to any intrinsic code.

#include <metal_stdlib>
#include <metal_math>
#include <metal_texture>
using namespace metal;

Parameter blocks and Argument Buffers

ParameterBlock values are translated into Argument Buffers potentially containing nested resources. For example this Slang code…

struct MyParameters
{
    int x;
    int y;
    StructuredBuffer<float> buffer1;
    RWStructuredBuffer<uint3> buffer2;
}

ParameterBlock<MyParameters> gObj;

void main(){ ... gObj ... }

… results in this Metal output:

struct MyParameters
{
    int x;
    int y;
    float device* buffer1;
    uint3 device* buffer2;
};

[[kernel]] void main(MyParameters constant* gObj [[buffer(1)]])

Struct Parameter Flattening

When targeting Metal, top-level nested struct parameters are automatically flattened. For example:

struct NestedStruct
{
    float2 uv;
};
struct InputStruct
{
    float4 position;
    float3 normal;
    NestedStruct nested;
};

Will be flattened to:

struct InputStruct
{
    float4 position;
    float3 normal;
    float2 uv;
};

Return Value Handling

Non-struct return values from entry points are automatically wrapped in a struct with appropriate semantics. For example:

float4 main() : SV_Target
{
    return float4(1,2,3,4);
}

becomes:

struct FragmentOutput
{
    float4 value : SV_Target;
};
FragmentOutput main()
{
    return { float4(1,2,3,4) };
}

Value Type Conversion

Metal enforces strict type requirements for certain operations. Slang automatically performs the following conversions:

• Vector size expansion (e.g., float2 to float4), for example when the user specified float2 but the semantic type in Metal is float4.
• Image store value expansion to 4-components

For example:

RWTexture2D<float2> tex;
tex[coord] = float2(1,2);  // Automatically expanded to float4(1,2,0,0)

Conservative Rasterization

Since Metal doesn’t support conservative rasterization, SV_InnerCoverage is always false.

Address Space Assignment

Metal requires explicit address space qualifiers. Slang automatically assigns appropriate address spaces:

Variable Type	Metal Address Space
Local Variables	thread
Global Variables	device
Uniform Buffers	constant
RW/Structured Buffers	device
Group Shared	threadgroup
Parameter Blocks	constant

Explicit Parameter Binding

The HLSL :register() semantic is respected when emitting Metal code.

Since metal does not differentiate a constant buffer, a shader resource (read-only) buffer and an unordered access buffer, Slang will map register(tN), register(uN) and register(bN) to [[buffer(N)]] when such register semantic is declared on a buffer typed parameter.

spaceN specifiers inside register semantics are ignored.

The [vk::location(N)] attributes on stage input/output parameters are respected.

Specialization Constants

Specialization constants declared with the [SpecializationConstant] or [vk::constant_id] attribute will be translated into a function_constant when generating Metal source. For example:

[vk::constant_id(7)]
const int a = 2;

Translates to:

constant int fc_a_0 [[function_constant(7)]];
constant int a_0 = is_function_constant_defined(fc_a_0) ? fc_a_0 : 2;