GPU Thread compute

 https://m.blog.naver.com/PostView.naver?isHttpsRedirect=true&blogId=julie_eun1014&logNo=221116294215

1. 멀티 프로세서와 CUDA 코어

GPU의 구조는 코어 개수 * 코어 클락 으로 이루어졌습다.
따라서 GPU 코어는 CPU 코어와는 다른 구조를 가지는 것을 알 수 있습니다.

CPU는 "멀티 코어를 구성하는 전체 구성 요소"를 뜻하고,
GPU에서는 이러한 구성 요소를 멀티 프로세서 라고 부릅니다.
멀티 프로세서 : 계산 유닛, 로드스토어 유닛 등 GPU를 구성하는 기본 요소로 구성된 최소단위.
GPU는 멀티프로세서 내에 여러 개의 GPU 코어를 가지고 있음.
그 안에 계산 유닛을  GPU 코어 라고 합니다.
CPU코어와 GPU 코어의 역할은 동일합니다.

CPU는 하나의 칩당 멀티코어로 4~12개 정도의 코어를 가지고 있습니다.

GPU는 하나의 칩에 최대 15개의 멀티 프로세서를 가지고 있습니다.
이 멀티 프로세서는 각각 안에 192개의 코어(계산 유닛)을 가지고 있습니다.
따라서 하나의 칩에 총 코어가 2880개 입니다.

2. Block과 Thread의 개념

GPU 내에 여러 개의 멀티 프로세서가 존재하는데,
하나의 멀티 프로세서에서 작동되는 "스레드의 모음"을 블록이라고 합니다.
따라서 GPU안에서 동작하는 모든 스레드의 개수는
블록 * 블록 내의 스레드 개수 입니다.

스레드는 멀티 프로세서 내에서 작동되는 코어 하나를 스레드라고 합니다.
정확히는 코어 하나에 할당된 명령어입니다..

블록과 스레드는 하드웨어와 1:n으로 매칭되기 때문에
자원이 허용되는 한 하나의 멀티 프로세서에서 여러 개의 블록을 동시에 실행시킬 수 있습니다.
그러면 스레드도 마찬가지로 여러 개를 동시에 실행시킬 수 있겠죠?

3. Kernel 실행시키기

GPU 병렬 프로그래밍에서의 kernel(커널)은
GPU에서 병렬 실행되는 명령의 모음을 뜻합니다.

GPU 커널은 OpenMP의 Fork-Join 모델과 유사하기 때문에
fork를 통해 병렬화를 해야합니다.
fork(포크)란 멀티 프로세스 시스템에서 하나의 프로세스가 자신의 프로세스를 복제하여
여러 개의 프로세스를 생성하는 동작입니다.

멀티 스레드 프로그래밍에서는 하나의 프로세스에서 여러 개의 자식 스레드로 작업을 처리하기 때문에
fork가 필요합니다.


쨌든, 블록과 스레드를 포크하기 위해서는 GPU함수를 실행해야 합니다.
GPU 함수를 호출할 때는, C언어의 함수 호출문에 꺽기 기호명령을 추가해야 합니다. (<<<,>>>)
이 꺽기 기호를 "트리플 세브론"이라고 부릅니다.

func<<<A,B>>>();

A : 블록 개수
B : 스레드 개수
func이라는 이름을 가진 GPU함수를 호출할 때 쓰는 문법입니다.

그런데 GPU함수를 선언할 때 일반 함수처럼 선언을 할까요? 아닙니다.
__global__이라는 키워드를 앞에 붙여서 GPU함수임을 명시해줘야 합니다.

__global__void func(){
}

예를 들어서,
func<<<5,4>>>();
라고 호출문을 작성한다면
GPU에 5개의 블록이 생성되고, 각각의 블록은 스레드를 4개씩 가집니다.

블록과 스레드의 개수가 많아지면 블록 개수, 스레드 개수, 고유한 스레드 ID등이 필요합니다.
이를 위해서 CUDA에는 변수가 정의되어 있습니다.
(blockIdx, threadIdx, blockDim, gridDim)

여기서 이 변수를 이해하기 위해서 CUDA 헤더파일에 정의된 구조체를 살펴보아야 합니다.
바로 dim3인데요, 이 dim3로 선언한 변수는
변수명.x, 변수명.y, 변수명.z의 구조체를 가진 3차원의 형태로 표시할 수 있습니다.