개미 군단 멘털 모델

CPU = 문어, GPU = 개미 군단

이 quest 나머지가 딸깍 들어맞게 하는 elevator-pitch 비유:

CPU는 문어 — 똑똑하고 팔 여러 개지만 한 번에 도형 몇 개만 색칠. GPU는 수천 마리 개미가 우글거리는 군단. 각 개미는 더 단순한데, 다 같이 붙으면 문어가 한 페이지 끝내기 전에 색칠북 통째로 다 끝내.

매핑하면 그 직관이 구체적 하드웨어 구조가 돼.

하드웨어 계층 (macro → micro)

GPU — 군단 자체. 예: RTX 4090은 16,384 CUDA core.
GPC (Graphics Processing Cluster) — 도시 크기 구. RTX 4090은 7개.
SM (Streaming Multiprocessor) — 공동 작업장. RTX 4090은 SM 128개.
Warp — 한 SM에서 lockstep 행진하는 32마리 분대.
Thread — 개별 개미.

소프트웨어 / 실행 체인 (커널 관점)

Kernel — 수천 개미한테 broadcast하는 mission briefing.
Grid — 캠페인 전체 (이번 launch의 모든 block).
Block — 멤버끼리 대화 (shared memory) + 동기화 가능한 단일 분대.
Warp — block 안의 32 마리 분대, lockstep.
Thread — 최전선 개별 개미.

Block-to-SM 할당은 GPU 스케줄러 일이야. 1,000 block 요청하면 스케줄러가 SM 128개에 자원 비는 대로 나눠줘. 그래서 출력 순서가 non-deterministic — 군단이 self-schedule 해.

Code

커널 launch에 계층 매핑·cuda

__global__ void example(int *out, int n) {
    int gid = blockIdx.x * blockDim.x + threadIdx.x;
    //         ^^^^^^^^^   ^^^^^^^^^^^   ^^^^^^^^^^^
    //         grid 안     block 크기    block 안
    //         어느 block (per-block thread 수) thread index
    if (gid < n) out[gid] = gid;
}

int main() {
    int n = 4096;
    int threadsPerBlock = 256;        // 전형적 sweet spot
    int blocks = (n + threadsPerBlock - 1) / threadsPerBlock;
    example<<<blocks, threadsPerBlock>>>(d_out, n);
    //        ^^^^^^  ^^^^^^^^^^^^^^^^
    //        grid    block (block당 thread 수)
    cudaDeviceSynchronize();
}

Metal에서 같은 계층 — 같은 단어, 다른 spell·metal

kernel void example(
    device int *out [[buffer(0)]],
    constant uint &n [[buffer(1)]],
    uint gid [[thread_position_in_grid]])
{
    if (gid < n) out[gid] = int(gid);
}

// Host (Swift):
// encoder.dispatchThreadgroups(
//     MTLSize(width: blocks, height: 1, depth: 1),
//     threadsPerThreadgroup: MTLSize(width: 256, height: 1, depth: 1))

CPU = 문어, GPU = 개미 군단

하드웨어 계층 (macro → micro)

소프트웨어 / 실행 체인 (커널 관점)

Code

External links

Exercise

Progress

댓글 0