Metal Shading Language이 뭔데

C++11에 attribute 괄호 추가 — 그리고 architectural twist 하나

Metal Shading Language (MSL)이 .metal 파일 안에 쓰는 거야. CUDA 개발자가 보면 즉시 친숙해 — function qualifier로 GPU entry 표시, attribute 괄호로 SIMT index 노출, address-space 키워드로 memory region 명시. 그래픽스 개발자가 보면 GLSL의 더 lean하고 typed한 사촌처럼 보여.

가장 중요한 확장 셋:

kernel function qualifier — CPU가 launch하는 GPU entry. CUDA의 __global__ 쌍둥이.
Attribute 괄호 — [[thread_position_in_grid]], [[threadgroup_position_in_grid]], [[thread_index_in_threadgroup]]. Metal의 threadIdx + blockIdx 답.
Address-space 키워드 — device, threadgroup, constant, thread — 포인터가 어느 메모리 region에 사는지 명시. CUDA는 __shared__/__constant__로 추론, Metal은 type의 일부로 만들어.

나머지는 표준 C++11 + Apple 확장이야: packed vector (float4, half3), threadgroup barrier, math built-in. 전체 MSL spec이 PDF 한 권 — 오후에 다 읽혀.

눈에 띄지 않게 숨어있는 architectural twist: Apple Silicon에선 device 메모리랑 CPU-visible 메모리가 같은 물리 RAM이야. 다음 레슨 통째로 이거 다룰 거.

Code

Metal kernel — CUDA랑 같은 모양, 다른 어휘·metal

#include <metal_stdlib>
using namespace metal;

// 'kernel' = CUDA의 __global__
// device   = GPU-visible 메모리 포인터
// constant = dispatch당 한 번 전달되는 read-only uniform
// [[thread_position_in_grid]] = threadIdx + blockIdx 합친 거
kernel void vector_add(
    const device float *A     [[buffer(0)]],
    const device float *B     [[buffer(1)]],
    device       float *C     [[buffer(2)]],
    constant     uint  &N     [[buffer(3)]],
    uint gid                  [[thread_position_in_grid]])
{
    if (gid < N) C[gid] = A[gid] + B[gid];
}

외워둘 만한 흔한 MSL built-in·metal

// thread별 정체성
uint gid     [[thread_position_in_grid]];        // global thread index
uint tid     [[thread_index_in_threadgroup]];    // threadgroup 안 local index
uint tg_idx  [[threadgroup_position_in_grid]];   // 어느 threadgroup인가
uint tg_size [[threads_per_threadgroup]];        // 내 threadgroup 크기

// SIMD-group (Apple의 warp, 32 thread)
uint simd_lane [[thread_index_in_simdgroup]];
uint simd_id   [[simdgroup_index_in_threadgroup]];

Metal Shading Language이 뭔데

C++11에 attribute 괄호 추가 — 그리고 architectural twist 하나

Code

External links

Exercise

Progress

댓글 0