Siraj Razick

Posted on Nov 4 AI-assisted

Ray Tracing on macOS: A Comprehensive Guide & Hands-On Tutorial

#gamedev #programming #softwaredevelopment #tutorial

Few notes that helped me while porting Quake engine to support Ray Tracing on Mac, sharing so others can benefit.

original game is here : https://buymeacoffee.com/plexydesk/e/593896

Overview
Prerequisites
Core Concepts
Hands-On Tutorial: Step-by-Step
Official Sample Analysis
Troubleshooting
References

===

1. Overview

What is Ray Tracing in Metal?

Ray tracing in Apple's Metal API is a hardware-accelerated rendering technique that simulates the physical behavior of light by tracing rays through a scene. Introduced with Metal 3 (WWDC 2022), Metal provides a native ray tracing subsystem built on top of acceleration structures, enabling real-time lighting, shadows, reflections, and global illumination effects.

Unlike traditional rasterization—which projects triangles onto a screen pixel by pixel—ray tracing casts rays from the camera (or light sources) into the scene, computing exact intersections with geometry. This produces physically accurate results including:

Soft shadows from area lights
Realistic reflections and refractions
Ambient occlusion and indirect lighting
Accurate depth-of-field effects

Supported Hardware

Key Point: Hardware-accelerated ray tracing requires an Apple Silicon Mac (M1 or later). On Intel Macs, ray tracing may fall back to CPU/software implementations, which are significantly slower.

The Metal Ray Tracing Pipeline

+======================+    +=========================+    +====================+
| Geometry Data |---->|| Acceleration Structure |---->|| Ray Tracing Pass   |
| (Vertices,     |    || (BVH Tree)               |    || (Compute Shader)  |
|  Indices)      |    ||                         |    ||                    |
+======================+    +=========================+    +====================+
                                                             |
                                                     +-------v-------+
                                                     | Output Image  |
                                                     | (Texture)     |
                                                     +---------------+

The workflow consists of three main stages:

Build the Acceleration Structure (AS) — Organize geometry for efficient ray intersection tests
Configure the Ray Tracing Command Encoder — Set up shaders, function tables, and resources
Dispatch the Ray Generation Routine — Execute the compute shader that traces rays

===

2. Prerequisites

Minimum Requirements

Checking Device Support

Before using ray tracing, always verify that the target device supports it:

// Swift — Check if the default device supports ray tracing
import Metal

let device = MTLCreateSystemDefaultDevice()!
if device.supportsFamily(.apple3) {
    print("Device supports Apple Silicon features including ray tracing")
} else {
    print("Ray tracing may not be hardware-accelerated on this device")
}

// Objective-C
#import <Metal/Metal.h>

id<MTLDevice> device = MTLCreateSystemDefaultDevice();
if ([device supportsFamily:MTLGPUFamilyApple3]) {
    NSLog(@"Device supports ray tracing");
}

The MTLGPUFamilyApple3 family identifier indicates Apple Silicon GPUs with full Metal 3 feature support, including hardware ray tracing.

===

3. Core Concepts

3.1 Acceleration Structures (AS)

An acceleration structure is a hierarchical data structure (typically a Bounding Volume Hierarchy, or BVH) that organizes scene geometry to enable fast ray-intersection queries. Without an AS, every ray would need to test against every triangle in the scene — O(n) complexity. With an AS, the search becomes approximately O(log n).

Key Classes

Building Triangle Geometry

// Swift — Create a triangle geometry descriptor
let triangleDescriptor = MTLAccelerationStructureTriangleGeometryDescriptor()
triangleDescriptor.triangleCount = numberOfTriangles
triangleDescriptor.vertexBuffer = vertexBuffer
triangleDescriptor.vertexBufferOffset = 0
triangleDescriptor.vertexStride = strideInBytes
triangleDescriptor.vertexFormat = .float3
triangleDescriptor.indexBuffer = indexBuffer
triangleDescriptor.indexBufferOffset = 0
triangleDescriptor.indexType = .uint16

// Objective-C
MTLAccelerationStructureTriangleGeometryDescriptor *triangleDescriptor =
    [MTLAccelerationStructureTriangleGeometryDescriptor new];
triangleDescriptor.triangleCount = numberOfTriangles;
triangleDescriptor.vertexBuffer = vertexBuffer;
triangleDescriptor.vertexBufferOffset = 0;
triangleDescriptor.vertexStride = sizeof(Vertex);
triangleDescriptor.vertexFormat = MTLVertexFormatFloat3;
triangleDescriptor.indexBuffer = indexBuffer;
triangleDescriptor.indexBufferOffset = 0;
triangleDescriptor.indexType = MTLIndexTypeUInt16;

Configuring the Overall AS Descriptor

// Swift — Configure the acceleration structure descriptor
let asDescriptor = MTLAccelerationStructureDescriptor()
asDescriptor.usage = .allowQuery // Options: .static, .dynamic, .allowQuery
asDescriptor.geometryDescriptors = [triangleDescriptor]

// Calculate required scratch buffer size
let scratchBufferSize = device.accelerationStructureScratchBufferWidth(
    descriptor: asDescriptor
)
let scratchBuffer = device.makeBuffer(length: Int(scratchBufferSize))!

The usage property controls how the AS behaves at runtime:

.static (default): The AS will not change after building. Best performance.
.dynamic: The AS may be updated/refitted. Slightly slower queries.
.allowQuery: Enables ray casting queries against the AS. Required for ray tracing.

3.2 Ray Generation Routines

A ray generation routine is a compute shader entry point marked with the [[kernel]] attribute. It is responsible for:

Computing the initial ray origin and direction for each pixel/thread
Invoking ray tracing via the intersector
Handling hit/miss cases and writing output color

// Metal Shading Language — Ray Generation Kernel
#include <metal_stdlib>
using namespace metal;

#define GEOMETRY_MASK_TRIANGLE 1
#define RAY_MASK_PRIMARY (GEOMETRY_MASK_TRIANGLE)

struct Uniforms {
    uint width;
    uint height;
    float4x4 viewMatrix;
    float4x4 projectionMatrix;
};

kernel void raygen(
    device unsigned char* outputImage [[buffer(0)]],
    constant Uniforms& uniforms [[buffer(1)]],
    device MTLAccelerationStructure* accelerationStructure [[buffer(2)]],
    uint3 gid [[thread_position_in_grid]]
) {
    uint x = gid.x;
    uint y = gid.y;

    if (x >= uniforms.width || y >= uniforms.height) return;

    // Compute ray origin and direction from screen coordinates
    float2 uv = float2(x, y) / float2(uniforms.width, uniforms.height) * 2.0 - 1.0;
    uv.y = -uv.y; // Flip Y

    // Simple pinhole camera ray setup
    float3 rayOrigin = float3(0.0, 1.0, 3.0);
    float3 rayDir = normalize(float3(uv, -1.0));

    // Define ray bounds
    float tmin = 0.01f;
    float tmax = 1000.0f;

    // Perform ray intersection query
    float3 hitPoint;
    float2 barycentrics;
    uint primitiveId;
    bool hit = accelerationStructure->intersect(
        rayOrigin, rayDir, tmin, tmax,
        RAY_MASK_PRIMARY, hitPoint, barycentrics, primitiveId
    );

    // Write result to output texture
    if (hit) {
        // Hit: shade the intersection point
        float3 color = float3(0.8, 0.2, 0.3); // Example hit color
        outputImage[(y * uniforms.width + x) * 4 + 0] = uint8_t(color.r * 255);
        outputImage[(y * uniforms.width + x) * 4 + 1] = uint8_t(color.g * 255);
        outputImage[(y * uniforms.width + x) * 4 + 2] = uint8_t(color.b * 255);
        outputImage[(y * uniforms.width + x) * 4 + 3] = 255;
    } else {
        // Miss: background color
        outputImage[(y * uniforms.width + x) * 4 + 0] = 0;
        outputImage[(y * uniforms.width + x) * 4 + 1] = 0;
        outputImage[(y * uniforms.width + x) * 4 + 2] = 0;
        outputImage[(y * uniforms.width + x) * 4 + 3] = 255;
    }
}

3.3 Intersection & Miss Shaders

Metal supports custom intersection functions that allow you to define bespoke intersection logic beyond standard triangle hits. These are marked with the [[intersection]] attribute.

// Custom sphere intersection function
struct SphereIntersection {
    bool accept [[accept_intersection]];
    float distance [[distance]];
};

constant bool useResourcesBuffer = true;

[[intersection(bounding_box, triangle_data, instancing)]]
SphereIntersection sphereIntersectionFunction(
    float3 origin [[origin]],
    float3 direction [[direction]],
    float minDistance [[min_distance]],
    float maxDistance [[max_distance]],
    unsigned int primitiveId [[primitive_id]],
    unsigned int geometryIndex [[geometry_intersection_function_table_offset]],
    device void* resources [[buffer(0), function_constant(useResourcesBuffer)]]
) {
    // Mathematical sphere intersection test
    float3 oc = origin - spheres[primitiveId].origin;
    float a = dot(direction, direction);
    float b = 2.0f * dot(oc, direction);
    float c = dot(oc, oc) - spheres[primitiveId].radiusSquared;
    float disc = b * b - 4.0f * a * c;

    SphereIntersection ret;
    if (disc >= 0.0f) {
        float sqrtDisc = sqrt(disc);
        float t = (-b - sqrtDisc) / (2.0f * a);
        if (t >= minDistance && t <= maxDistance) {
            ret.accept = true;
            ret.distance = t;
        } else {
            ret.accept = false;
        }
    } else {
        ret.accept = false;
    }
    return ret;
}

The intersection function keywords control what data Metal passes between pipeline stages:

When you omit keywords, Metal stores less intermediate data, potentially improving performance.

===

4. Hands-On Tutorial: Step-by-Step

Step 1: Project Setup

Open Xcode → File → New → Project
Choose macOS → App (or iOS → App)
Select Swift or Objective-C as the language
In your project settings, ensure Metal Framework is linked:
- Go to Target → General → Frameworks, Libraries, and Embedded Content
- Add Metal.framework and MetalPerformanceShaders.framework
Create a .metal file for your shaders:
- File → New → File → Metal Shader File
- Name it RayTracing.metal

Step 2: Device Detection

Before creating any ray tracing resources, check device capability:

import Metal

class RayTracingRenderer {
    let device: MTLDevice
    let commandQueue: MTLCommandQueue

    init() throws {
        guard let device = MTLCreateSystemDefaultDevice() else {
            throw NSError(domain: "RayTracing", code: 1,
                          userInfo: [NSLocalizedDescriptionKey:
                             "No Metal-capable device found"])
        }
        self.device = device

        // Verify Apple Silicon support
        guard device.supportsFamily(.apple3) else {
            throw NSError(domain: "RayTracing", code: 2,
                          userInfo: [NSLocalizedDescriptionKey:
                             "Device does not support hardware ray tracing"])
        }

        guard let commandQueue = device.makeCommandQueue() else {
            throw NSError(domain: "RayTracing", code: 3,
                          userInfo: [NSLocalizedDescriptionKey:
                             "Failed to create command queue"])
        }
        self.commandQueue = commandQueue
    }
}

Step 3: Building the Acceleration Structure

This is the most critical step. You must build the AS before any ray tracing can occur.

// Swift — Complete AS build workflow

func buildAccelerationStructure() throws -> MTLAccelerationStructure {
    // 1. Create the triangle geometry descriptor
    let triangleDescriptor = MTLAccelerationStructureTriangleGeometryDescriptor()
    triangleDescriptor.triangleCount = triangleCount
    triangleDescriptor.vertexBuffer = vertexBuffer
    triangleDescriptor.vertexBufferOffset = 0
    triangleDescriptor.vertexStride = MemoryLayout<Vertex>.stride
    triangleDescriptor.vertexFormat = .float3
    triangleDescriptor.indexBuffer = indexBuffer
    triangleDescriptor.indexBufferOffset = 0
    triangleDescriptor.indexType = .uint16

    // 2. Configure the overall AS descriptor
    let asDescriptor = MTLAccelerationStructureDescriptor()
    asDescriptor.usage = .allowQuery
    asDescriptor.geometryDescriptors = [triangleDescriptor]

    // 3. Create the acceleration structure buffer
    let asSize = device.accelerationStructureSize(descriptor: asDescriptor)
    guard let accelerationStructure = device.makeAccelerationStructure(
        descriptor: asDescriptor
    ) else {
        throw NSError(domain: "RayTracing", code: 4,
                      userInfo: [NSLocalizedDescriptionKey:
                         "Failed to create acceleration structure"])
    }

    // 4. Allocate scratch buffer (temporary working memory)
    let scratchBufferSize = device.accelerationStructureScratchBufferWidth(
        descriptor: asDescriptor
    )
    let scratchBuffer = device.makeBuffer(length: Int(scratchBufferSize))!

    // 5. Encode the build command
    guard let commandBuffer = commandQueue.makeCommandBuffer() else {
        throw NSError(domain: "RayTracing", code: 5,
                      userInfo: [NSLocalizedDescriptionKey:
                         "Failed to make command buffer"])
    }

    guard let encoder = commandBuffer.makeAccelerationStructureCommandEncoder() else {
        throw NSError(domain: "RayTracing", code: 6,
                      userInfo: [NSLocalizedDescriptionKey:
                         "Failed to make acceleration structure command encoder"])
    }

    // The build method signature:
    // build(accelerationStructure:descriptor:scratchBuffer:scratchBufferOffset:)
    encoder.build(
        accelerationStructure: accelerationStructure,
        descriptor: asDescriptor,
        scratchBuffer: scratchBuffer,
        scratchBufferOffset: 0
    )

    encoder.endEncoding()
    commandBuffer.commit()
    commandBuffer.waitUntilCompleted()

    return accelerationStructure
}

Why the scratch buffer? The acceleration structure build process requires temporary GPU memory to construct the BVH tree internally. The scratchBuffer provides this workspace. Its required size is queried via accelerationStructureScratchBufferWidth(descriptor:). After the build completes, the scratch buffer can be reused for subsequent builds or refits.

Step 4: Defining the Ray Gen Shader

Create your Metal shader file (RayTracing.metal):

// RayTracing.metal
#include <metal_stdlib>
#include <simd/simd.h>
using namespace metal;

// Constants passed from the CPU
struct CameraParams {
    float4x4 viewMatrix;
    float4x4 projectionMatrix;
    float2 viewportSize;
};

struct RayPayload {
    float3 color;      // RGB color of the hit surface
    float3 normal;     // Surface normal at hit point
    float3 position;   // World-space position of the hit
    float distance;    // Distance from ray origin
    bool hit;          // Whether any geometry was hit
};

// Ray generation kernel — one thread per pixel
[[kernel]] void raygen(
    device RayPayload* payloads       [[buffer(0)]],
    device MTLAccelerationStructure* accelerationStructure [[buffer(1)]],
    constant CameraParams& params     [[buffer(2)]],
    uint2 gid                         [[thread_position_in_grid]]
) {
    uint x = gid.x;
    uint y = gid.y;

    if (x >= uint(params.viewportSize.x) || y >= uint(params.viewportSize.y))
        return;

    // --- Ray Setup ---
    // Map pixel coordinates to NDC [-1, 1]
    float2 ndc = (float2(x, y) / params.viewportSize) * 2.0 - 1.0;
    ndc.y = -ndc.y; // Flip Y for OpenGL-style coords

    // Build ray direction from camera matrices
    float4 clipPos = float4(ndc, 0.0, 1.0);
    float4 eyePos = inverse(params.projectionMatrix) * clipPos;
    eyePos = eyePos / eyePos.w;
    float4 worldPos = inverse(params.viewMatrix) * eyePos;
    float3 rayOrigin = params.viewMatrix[3].xyz; // Camera position
    float3 rayDirection = normalize(worldPos.xyz - rayOrigin);

    // --- Ray Query ---
    float tmin = 0.001f;
    float tmax = 1000.0f;
    uint rayFlags = 0xFF; // Test all geometry
    uint instanceMask = 0xFF;

    float3 hitPoint;
    float2 barycentrics;
    uint primitiveId;
    bool hit = accelerationStructure->intersect(
        rayOrigin, rayDirection, tmin, tmax,
        rayFlags, instanceMask,
        hitPoint, barycentrics, primitiveId
    );

    // --- Output ---
    RayPayload payload;
    payload.hit = hit;
    if (hit) {
        payload.position = hitPoint;
        payload.distance = length(hitPoint - rayOrigin);
        payload.color = float3(0.8f, 0.2f, 0.3f); // Placeholder shading
        payload.normal = float3(0.0f, 1.0f, 0.0f);
    } else {
        payload.color = float3(0.0f, 0.05f, 0.1f); // Sky color
        payload.normal = float3(0.0f);
        payload.position = float3(0.0f);
        payload.distance = tmax;
    }

    payloads[gid.y * uint(params.viewportSize.x) + gid.x] = payload;
}

Compile the shader library:

// Swift — Compile and load the shader library
let defaultLibrary = device.makeDefaultLibrary()!
let raygenFunction = defaultLibrary.makeFunction(name: "raygen")!

// Create a compute pipeline state for the ray gen shader
let raygenPipeline = try device.makeComputePipelineState(
    function: raygenFunction
)

Step 5: Dispatching the Ray Tracing Command

Now that you have an acceleration structure and a compiled ray gen shader, encode the actual ray tracing work:

// Swift — Encode and dispatch the ray tracing pass

func renderFrame(outputTexture: MTLTexture) throws {
    guard let commandBuffer = commandQueue.makeCommandBuffer() else { return }

    // --- Ray Tracing Compute Pass ---
    guard let computeEncoder = commandBuffer.makeComputeCommandEncoder() else { return }

    // Set the pipeline state
    computeEncoder.setComputePipelineState(raygenPipeline)

    // Bind the acceleration structure (resource at buffer slot 1)
    computeEncoder.setAccelerationStructure(accelerationStructure, at: 1)

    // Bind the output texture
    computeEncoder.setTexture(outputTexture, at: 0)

    // Bind camera parameters
    var cameraParams = CameraParams(
        viewMatrix: viewMatrix,
        projectionMatrix: projectionMatrix,
        viewportSize: float2(Float(textureWidth), Float(textureHeight))
    )
    computeEncoder.setBytes(&cameraParams, length: MemoryLayout<CameraParams>.size, at: 2)

    // Calculate threadgroup dimensions (one thread per pixel)
    let threadsPerThreadgroup = MTLSize(width: 16, height: 16, depth: 1)
    let numThreadgroups = MTLSize(
        width: (textureWidth + 15) / 16,
        height: (textureHeight + 15) / 16,
        depth: 1
    )

    // Dispatch the ray gen kernel
    computeEncoder.dispatchThreadgroups(numThreadgroups, threadsPerThreadgroup: threadsPerThreadgroup)

    computeEncoder.endEncoding()

    // Present or blit the result
    // ... (presentation code depends on your view setup)

    commandBuffer.commit()
}

Complete Workflow Summary:

1. Create MTLAccelerationStructureTriangleGeometryDescriptor
   |
2. Configure MTLAccelerationStructureDescriptor with usage=.allowQuery
   |
3. Create MTLAccelerationStructure buffer on the device
   |
4. Allocate scratch buffer (query size first)
   |
5. Encode build command via MTLAccelerationStructureCommandEncoder
   |
6. Commit command buffer and wait
   |
7. Create compute pipeline from ray gen shader
   |
8. Encode compute pass: set AS, textures, parameters
   |
9. Dispatch threadgroups (one per pixel or tile)
   |
10. Commit and present

===

5. Official Sample Analysis

Analyzing the gfx-rs Metal Ray Tracing Sample

One of the most thorough open-source examples of Metal ray tracing is the gfx-rs metal-rs ray tracing sample, available at:

https://github.com/gfx-rs/metal-rs/tree/master/examples/raytracing

This sample demonstrates a complete path tracer implemented in Metal, including:

Key Implementation Details

1. Geometry Masks and Ray Types

#define GEOMETRY_MASK_TRIANGLE 1
#define GEOMETRY_MASK_SPHERE   2
#define GEOMETRY_MASK_LIGHT    4
#define GEOMETRY_MASK_GEOMETRY (GEOMETRY_MASK_TRIANGLE | GEOMETRY_MASK_SPHERE)
#define RAY_MASK_PRIMARY   (GEOMETRY_MASK_GEOMETRY | GEOMETRY_MASK_LIGHT)
#define RAY_MASK_SHADOW    GEOMETRY_MASK_GEOMETRY
#define RAY_MASK_SECONDARY GEOMETRY_MASK_GEOMETRY

The sample uses bitmasks to filter which geometry types each ray should intersect. Primary rays test everything (geometry + lights), shadow rays only test geometry, and secondary rays (reflections/refractions) also only test geometry.

2. Custom Intersection Functions

The sample implements custom sphere intersection functions using the [[intersection]] attribute:

[[intersection(bounding_box, triangle_data, instancing)]]
BoundingBoxIntersection sphereIntersectionFunction(
    float3 origin [[origin]],
    float3 direction [[direction]],
    float minDistance [[min_distance]],
    float maxDistance [[max_distance]],
    unsigned int primitiveId [[primitive_id]],
    unsigned int geometryIndex [[geometry_intersection_function_table_offset]],
    device void* resources [[buffer(0), function_constant(useResourcesBuffer)]]
)

This demonstrates how Metal allows you to define bespoke intersection logic for non-standard primitives (like spheres) while still leveraging the hardware BVH traversal for bounding box culling.

3. Per-Primitive Data

The sample shows two approaches to binding geometry data:

Per-primitive data: Each primitive has its own data buffer, accessed via the [[primitive_data]] attribute
Resources buffer: A single buffer indexed by geometry type

if (usePerPrimitiveData) {
    sphere = *(const device Sphere*)perPrimitiveData;
} else {
    device SphereResources& sphereResources =
        *(device SphereResources*)((device char*)resources + resourcesStride * geometryIndex);
    sphere = sphereResources.spheres[primitiveId];
}

4. Monte Carlo Sampling

The sample implements cosine-weighted hemisphere sampling for diffuse shading:

inline float3 sampleCosineWeightedHemisphere(float2 u) {
    float phi = 2.0f * M_PI_F * u.x;
    float cos_phi;
    float sin_phi = sincos(phi, cos_phi);
    float cos_theta = sqrt(u.y);
    float sin_theta = sqrt(1.0f - cos_theta * cos_theta);
    return float3(sin_theta * cos_phi, cos_theta, sin_theta * sin_phi);
}

5. Halton Sequence for Anti-Aliasing

float halton(unsigned int i, unsigned int d) {
    unsigned int b = primes[d];
    float f = 1.0f;
    float invB = 1.0f / b;
    float r = 0;
    while (i > 0) {
        f = f * invB;
        r = r + f * (i % b);
        i = i / b;
    }
    return r;
}

This low-discrepancy sequence generates more evenly distributed samples than pure random numbers, reducing noise in rendered images.

Apple's Official Samples

Apple also provides official ray tracing samples through the Apple Developer portal:

"Accelerating ray tracing using Metal" — Demonstrates basic ray tracing with motion blur
"Rendering reflections in real time using ray tracing" — Implements real-time reflections via ray-traced compute passes
"Enhance visual quality with curves using Metal ray tracing" — Uses ray tracing for curve rendering

These samples are accessible at: https://developer.apple.com/metal/sample-code/

===

6. Troubleshooting

Common Errors and Solutions

Debugging with Xcode GPU Frame Capture

Xcode provides powerful tools for debugging ray tracing pipelines:

Enable Metal Validation:
- Go to Product → Scheme → Edit Scheme
- Select Run → Options → Metal Validation
- Enable Validate Render/Compute Commands
GPU Frame Capture:
- Click the Debug → Start GPU Frame Capture menu
- Run your application
- Xcode will capture the entire GPU command stream
Inspect the Command Buffer:
- In the GPU debugger, expand the command buffer hierarchy
- Navigate to the MTLAccelerationStructureCommandEncoder to inspect build parameters
- View the acceleration structure's internal BVH representation
Shader Debugger:
- Set breakpoints in your Metal shader functions
- Inspect thread groups, thread IDs, and resource bindings
- Watch variables like hit, barycentrics, and primitiveId during ray intersection
Counter Sampling:

   // Sample GPU counters for performance profiling
   let counterSet = MTLCounterSet.deviceCounterSet(device: device)!
   let timestampCounter = counterSet.counter(at: 0)!
   // Attach counters to your command buffer for timing analysis

Verification Checklist

Before deploying, verify:

[ ] Device supports MTLGPUFamilyApple3
[ ] Acceleration structure build completed without errors
[ ] Scratch buffer size was correctly queried
[ ] Ray gen shader compiles successfully
[ ] Threadgroup dimensions cover all pixels
[ ] Output texture format matches display requirements
[ ] All buffers are properly retained and released
[ ] No race conditions between AS updates and ray queries

===

7. References

Official Apple Documentation

Official Sample Code

Community Resources

WWDC Sessions

===

DEV Community