---
layout: user-guide
permalink: /user-guide/metal-target-specific
---

# 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

### Limitation: `isParameterLocationUsed` for varying inputs

Because Metal entry point varying inputs are packed into a single `[[stage_in]]` struct parameter during legalization, `IMetadata::isParameterLocationUsed` cannot distinguish between used and unused individual varying inputs. The metadata is recorded at the struct level, so if any varying input is used, all varying input locations covered by the struct will be reported as used. This differs from SPIR-V, where each varying input becomes a separate global parameter that can be individually tracked.

This limitation does not affect non-varying resource types (e.g. descriptor table slots), which are tracked individually.

## 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_FragInvocationCount`    | `(Not supported)`                                    |
| `SV_FragSize`               | `(Not supported)`                                    |
| `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_PointSize`              | `[[point_size]]`                                     |
| `SV_PointCoord`             | `[[point_coord]]`                                    |
| `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]]`                                  |
| `SV_VulkanInstanceID`       | `[[instance_id]]`                                    |
| `SV_VulkanSamplePosition`   | `(Not supported)`                                    |
| `SV_VulkanVertexID`         | `[[vertex_id]]`                                      |

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, which is the same as for task/mesh shaders generally in Slang.

```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.

```cpp
#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...

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

ParameterBlock<MyParameters> gObj;

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

... results in this Metal output:

```cpp
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:

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

Will be flattened to:

```cpp
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:

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

becomes:

```c++
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:

```slang
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.

## SubpassInput

Slang supports `SubpassInput` and `SubpassInputMS` on Metal, lowered to Metal's framebuffer fetch. The `[[vk::input_attachment_index(N)]]` attribute maps to Metal's `[[color(N)]]` fragment input, allowing a fragment shader to read the current value in the render target within the same render pass.

```slang
[[vk::input_attachment_index(0)]] SubpassInput<float4> color;
[[vk::input_attachment_index(1)]] SubpassInput<float4> prevBloom;

[shader("fragment")]
float4 main() : SV_Target
{
    return color.SubpassLoad() + prevBloom.SubpassLoad();
}
```

This generates a Metal fragment function with `[[color(N)]]` parameters. The `SV_Target` return value is wrapped in a Slang-generated output struct:

```cpp
struct pixelOutput { float4 output [[color(0)]]; };

[[fragment]] pixelOutput main(float4 color [[color(0)]], float4 prevBloom [[color(1)]]);
```

Note that `[[vk::input_attachment_index(N)]]` always maps to `[[color(N)]]` regardless of what the attachment semantically holds — Metal framebuffer fetch reads from color attachments only, not from `[[depth]]`/`[[stencil]]` attachments.

### Restrictions

- **Fragment stage only.** `SubpassInput` can only be used in fragment entry points.
- **Direct entry-point parameters only.** `SubpassInput` cannot be placed inside a `ParameterBlock`. It must be a global-scope declaration directly referenced by the entry point.
- **No cross-function references.** `SubpassInput` must be referenced only from the fragment entry point or from helper functions that are inlined into it. Mark any helper that calls `SubpassLoad` with `[ForceInline]` to guarantee inlining; `[noinline]` helpers will fail with error E56108.
- **`SubpassInputMS` per-sample limitation.** Metal does not support per-sample reads. When using `SubpassInputMS`, the sample index passed to `SubpassLoad` is ignored and the resolved value is returned. The compiler emits warning E56107 at each call site where a sample index is provided.

## 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 between 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:

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

Translates to:

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