스레드 블록 아키텍처 Thread Block Architecture

CPU의 스레드

CPU환경에서 생성하는 스레드는 생성 즉시 실행 코드가 프로세서에서 실행된다. 생성한 스레드의 개수와 CPU 코어 수가 동일하면 최적의 효율을 나타냄. 

 

CPU의 태스크 분할 처리 과정

1. 스레드의 개수를 코어 수와 동일하게 생성
2. 해야 할 작업을 태스크 단위로 작게 분할하여 태스크 큐에 넣는다.
3. 스레드가 작업을 태스크 단위로 가져와서 처리
4. 스레드가 하나의 테스크를 완료하면 다음 태스크를 큐에서 가져와서 다시 작업을 수행 

그리드 블록 모델 Grid Block Model

CUDA에서는 스레드가 모여 블록을 이루고 블록이 모여 그리드를 이루게 된다.

하나의 블록은 1개에서부터 최대 512개까지 스레드를 가질 수 있다.  블록 안에 있는 스레드는 고유 아이디를 갖는다. 블록이 스레드를 512개까지 가지고 있을 수 있기 때문에 0번부터 511번까지 고유한 인덱스 번호를 지정받게 된다. 블록 안에 스레드를 배치하는 방법은 1차원, 2차원, 3차원으로 지정할 수 있다. 

스레드 인덱스는 'threadIdx'로 이름이 지정된 변수를 사용한다. 

//1차원으로 배치된 스레드 6개
//thread 0 : threadIdx.x = 0
__global__ void kernel<<<1,6>>>(int a, int b, int c);

 CUDA 스레드 모델에서 SP는 4개의 스레드를 동시에 실행할 수 있고, 하나의 블록은 하나의 SM과 대응하여 동작하게 된다. 하나의 블록에서 스래드를 생성시켜 프로그램을 실행시키면 1개의 SM에서만 동작하여  GPU의 성능을 전부 발휘하지 못하고 비효율적이 된다. 

 

스레드 블록의 개수 설정

커널 함수 호출 문법

__global__ void kernel<<<Dg, Db, Ns, S>>>();

Dg(Dimensions of the grid: type dim3)

그리드의 크기로 그리드 안의 블록 개수를 3차원적으로 지정. 

Db(Dimensions of the block : type dim3)

블록의 크기로 블록 안의 스레드 개수를 3차원적으로 지정. 

Ns(Number of bytes shared memory dynamically allocated/ block : type size_t)

각 블록에서 사용하는 공유 메모리의 크기 지정. 정수형으로 입력해야 하며 생략 가능. 생략하면 디폴트 값 0 대입

S(Associated cudaStream_t)

스트림은 비동기 처리를 하려고 사용. 실행 스트림 번호를 '0'이상의 정수로 지정. 생략 가능, 생략 시 스트림 번호 '0'으로 그 커널 함수가 실행된다. 스트림 번호 '0'은 비동기 처리를 하지 않고 동기 구동을 하는 것을 의미.

 

그리드

그리드 : 블록의 모임

블록 : 스레드의 모임

스레드 -> 블록 -> 그리드 

그리드는 커널이 실행될 때 스레드를 구성하는 최상위 집합체로 디바이스에서 커널이 실행되는 모듈, 응용 프로그램을 나타냄. 하나의 디바이스에서 실행되는 그리드는 동일한 시점에 하나만 존재할 수 있음.