GPU 프로그래밍

각 스레드의 번호를 i로 하여, 위와 같이 32요소 간격으로 배열을 읽으면 a[i]를 읽는 경우보다 느려진다. 주된 이유는?

float x = a[i * 32];
  1. A인접한 스레드의 주소가 128바이트 떨어져 병합 접근이 되지 않고, 스레드마다 다른 메모리 덩어리를 읽게 되기 때문
  2. B곱셈 계산에 시간이 걸리기 때문
  3. C32요소 간격으로 읽으면 값을 올바르게 읽을 수 없기 때문
정답과 해설 보기

정답: A. 인접한 스레드의 주소가 128바이트 떨어져 병합 접근이 되지 않고, 스레드마다 다른 메모리 덩어리를 읽게 되기 때문

float는 4바이트이므로 a[i*32]에서는 인접한 스레드의 주소가 128바이트 떨어집니다. 워프의 32개 스레드는 32개의 서로 다른 덩어리를 읽고, 각 덩어리에서 4바이트만 사용합니다. a[i]라면 32개 스레드가 128바이트의 하나의 덩어리로 끝나므로, 옮기는 데이터 양에 수십 배의 차이가 납니다. 2차원 배열에서는 스레드 번호를 안쪽(연속된 주소)의 첨자에 대응시키는 것이 기본입니다.

이 분야 연습하기 분야로 돌아가기

GPU 프로그래밍의 다른 문제

  1. 입자의 위치 x, y, z를 가진 데이터를 GPU에서 처리할 때, 메모리 병합 접근 측면에서 유리한 배치는?
  2. CUDA 공유 메모리(shared memory)의 특징으로 올바른 것은?
  3. CUDA의 __syncthreads()의 역할로 올바른 것은?
  4. CUDA에서 __syncthreads()를 "if (threadIdx.x < 16)" 안에 써서 일부 스레드만 통과하도록 했다. 일어날 수 있는 일로 올바른 것은?
  5. CUDA 공유 메모리의 뱅크 충돌에 대한 설명으로 올바른 것은?

이 분야의 비교