This is the first of three notes on ray tracing, covering the algorithm itself and the shape of Metal's API.
The rendering equation
Every physically based renderer is approximating the same equation (Kajiya, 1986):
Plain TextL_o(p, ω_o) = L_e(p, ω_o) + ∫_Ω f_r(p, ω_i, ω_o) · L_i(p, ω_i) · (n · ω_i) dω_i
In words: the radiance leaving point p in direction ω_o is the light p emits itself, plus the integral over the hemisphere Ω of incoming light weighted by the BRDF.
The equation is recursive: L_i is itself some other point's L_o. That recursion is the mathematical essence of global illumination — light that bounces once goes on to bounce again. Rasterization cannot handle it directly and approximates it with precomputation and screen-space tricks; ray tracing evaluates it.
Monte Carlo estimation
The integral has no closed form and must be solved numerically. Monte Carlo estimates it by random sampling:
Plain Text∫ f(x) dx ≈ (1/N) · Σ f(x_i) / p(x_i)
where p is the sampling probability density. The estimator is unbiased: its expectation equals the true integral. Its error shows up as noise, with a standard deviation falling as 1/√N.
That 1/√N is the fundamental difficulty of ray tracing: halving the noise takes four times the samples. The gap between 1 spp and 100 spp is not a factor of a hundred in compute — it is the reason the entire real-time ray tracing field exists.
Importance sampling
Since the sampling distribution p is ours to choose, it should be as close as possible in shape to the integrand. Sampling more where p is proportional to f reduces variance.
For the GGX specular BRDF, sample the half vector from its normal distribution:
MSLfloat3 importanceSampleGGX(float2 xi, float3 n, float roughness) {
float a = roughness * roughness;
float phi = 2.0 * M_PI_F * xi.x;
float cosTheta = sqrt((1.0 - xi.y) / (1.0 + (a * a - 1.0) * xi.y));
float sinTheta = sqrt(1.0 - cosTheta * cosTheta);
float3 h = float3(sinTheta * cos(phi), sinTheta * sin(phi), cosTheta);
return tangentToWorld(h, n);
}
For diffuse, sample the hemisphere cosine-weighted, because the rendering equation already contains a (n · ω_i) factor:
MSLfloat3 cosineSampleHemisphere(float2 xi, float3 n) {
float r = sqrt(xi.x);
float phi = 2.0 * M_PI_F * xi.y;
float3 d = float3(r * cos(phi), r * sin(phi), sqrt(max(0.0, 1.0 - xi.x)));
return tangentToWorld(d, n);
}
Cosine sampling and the (n · ω_i) factor cancel exactly, reducing the estimator to a plain average of BRDF values — which is why it is the standard choice.
Multiple importance sampling (MIS) goes further and combines strategies. Sampling the light has low variance for small lights; sampling the BRDF has low variance on smooth surfaces. MIS weights the two with the balance heuristic and takes the strength of each. It is the single most effective optimization in a path tracer.
Metal's acceleration structures
The core operation in ray tracing is "what did this ray hit". Testing every triangle is O(n) and not viable. An acceleration structure (a BVH) brings it to O(log n).
Metal's acceleration structures come in two levels.
The BLAS (bottom level) contains actual geometry:
Swiftlet geometry = MTLAccelerationStructureTriangleGeometryDescriptor()
geometry.vertexBuffer = vertexBuffer
geometry.vertexStride = MemoryLayout<Vertex>.stride
geometry.indexBuffer = indexBuffer
geometry.indexType = .uint32
geometry.triangleCount = triangleCount
let primitiveDesc = MTLPrimitiveAccelerationStructureDescriptor()
primitiveDesc.geometryDescriptors = [geometry]
The TLAS (top level) contains instances of BLASes, each with a transform:
Swiftlet instanceDesc = MTLInstanceAccelerationStructureDescriptor()
instanceDesc.instancedAccelerationStructures = blasArray
instanceDesc.instanceCount = instanceCount
instanceDesc.instanceDescriptorBuffer = instanceBuffer
The value of two levels is reuse and incremental update: ten thousand identical trees need one BLAS and ten thousand instance transforms; when an object moves, only the TLAS is rebuilt (which is fast) and the BLAS is untouched.
Only deforming geometry — skinned characters, cloth — needs its BLAS updated, and that can be a refit rather than a full rebuild. A refit keeps the BVH's tree structure and only updates the bounding boxes, which is far cheaper, at the cost of degrading tree quality as the geometry drifts from its original shape.
Building takes a scratch buffer and a GPU command:
Swiftlet sizes = device.accelerationStructureSizes(descriptor: primitiveDesc)
let accel = device.makeAccelerationStructure(size: sizes.accelerationStructureSize)!
let scratch = device.makeBuffer(length: sizes.buildScratchBufferSize, options: .storageModePrivate)!
let encoder = commandBuffer.makeAccelerationStructureCommandEncoder()!
encoder.build(accelerationStructure: accel, descriptor: primitiveDesc,
scratchBuffer: scratch, scratchBufferOffset: 0)
encoder.endEncoding()
Intersection: the intersector
Metal provides an intersector template for casting rays. It works in a compute kernel and also in fragment or vertex functions (inline ray tracing):
MSL#include <metal_raytracing>
using namespace metal::raytracing;
kernel void raytrace(instance_acceleration_structure accel [[buffer(0)]],
texture2d<float, access::write> output [[texture(0)]],
constant Uniforms &u [[buffer(1)]],
uint2 gid [[thread_position_in_grid]])
{
ray r;
r.origin = u.cameraPosition;
r.direction = computeRayDirection(gid, u);
r.min_distance = 0.001; // avoid self-intersection
r.max_distance = INFINITY;
intersector<instancing, triangle_data> i;
i.assume_geometry_type(geometry_type::triangle);
i.force_opacity(forced_opacity::opaque); // faster when there is no alpha test
intersection_result<instancing, triangle_data> hit = i.intersect(r, accel);
if (hit.type == intersection_type::none) {
output.write(sampleSky(r.direction), gid);
return;
}
float2 bary = hit.triangle_barycentric_coord;
// interpolate vertex attributes with bary, then shade ...
}
That min_distance = 0.001 is not optional. A ray leaving a surface has a real chance, thanks to floating-point error, of immediately hitting the very triangle it started from, which shows up as black speckling across the image — the ray tracing version of shadow acne.
The more robust approach is to offset the ray origin along the normal, scaling the offset with distance from the camera, because floating-point precision is itself relative.
force_opacity(opaque) tells the hardware that no any-hit callback is needed, so traversal can stop at the first hit. Alpha-tested geometry such as foliage cannot do this and needs an any-hit function supplied through an intersection function table.
The next note covers using these rays for shadows and lighting.