3장: GPU 설계의 병렬 프로그래밍 모델
그래픽스 처리 장치(GPU)는 고정 기능 그래픽 가속기에서 다양한 응용 프로그램을 가속할 수 있는 고도로 병렬적이고 프로그래밍 가능한 컴퓨팅 엔진으로 발전해왔습니다. GPU의 대규모 병렬성을 효과적으로 활용할 수 있도록 NVIDIA CUDA, OpenCL, DirectCompute와 같은 여러 병렬 프로그래밍 모델과 API가 개발되었습니다. 이러한 프로그래밍 모델은 프로그래머가 GPU 하드웨어의 저수준 세부 사항을 숨기면서 그들의 응용 프로그램에서 병렬성을 표현할 수 있는 추상화를 제공합니다.
이 장에서는 GPU를 위한 병렬 프로그래밍 모델의 핵심 개념과 원칙을 탐색할 것입니다. SIMT(Single Instruction, Multiple Thread) 실행 모델, CUDA 프로그래밍 모델과 API, OpenCL 프레임워크에 초점을 맞출 것입니다. 또한 높은 성능과 효율을 달성하기 위해 알고리즘을 GPU 아키텍처에 매핑하는 기법에 대해 논의할 것입니다.
SIMT (Single Instruction, Multiple Thread) 실행 모델
SIMT 실행 모델은 현대 GPU가 대규모 병렬성을 달성하는 데 사용되는 근본적인 패러다임입니다. SIMT 모델에서 많은 수의 스레드가 병렬로 동일한 프로그램(커널이라 함)을 실행하지만, 각 스레드는 자신의 프로그램 카운터를 가지고 있어 스레드 ID와 작동하는 데이터에 따라 다른 실행 경로를 취할 수 있습니다.
커널 및 스레드 계층
GPU 커널은 많은 수의 스레드에 의해 병렬로 실행되는 함수입니다. 커널을 실행할 때 프로그래머는 생성할 스레드 수와 이들을 그리드, 블록(또는 협력 스레드 어레이 - CTA), 개별 스레드의 계층으로 구성하는 방법을 지정합니다.
- 그리드는 전체 문제 공간을 나타내며 하나 이상의 블록으로 구성됩니다.
- 블록은 서로 협력하고 공유 메모리와 장벽을 통해 동기화할 수 있는 스레드의 그룹입니다. 블록 내의 스레드는 동일한 GPU 코어(스트리밍 멀티프로세서라 함)에서 실행됩니다.이 마크다운 파일에 대한 한국어 번역은 다음과 같습니다. 코드의 경우 주석만 번역되며, 코드 자체는 번역되지 않습니다.
(계산 단위 또는 SOR)
- 각 스레드는 자신의 블록과 그리드 내에서 고유한 ID를 가지고 있으며, 이를 이용하여 메모리 주소를 계산하고 제어 흐름 결정을 내릴 수 있습니다.
이러한 계층적 구조를 통해 프로그래머는 데이터 병렬성(동일한 작업이 여러 데이터 요소에 적용되는 경우)과 태스크 병렬성(다른 작업이 병렬로 실행되는 경우)을 모두 표현할 수 있습니다.
그림 3.1은 SIMT 실행 모델에서의 스레드 계층 구조를 보여줍니다.
Grid
________________
/ / / / /
/ / / / /
/ / / / /
/ / / / /
/__/__/__/__/__/
| | | |
| | Block |
| | | |
스레드 스레드 ...그림 3.1: SIMT 실행 모델에서의 스레드 계층 구조.
SIMT 실행
SIMT 실행 모델에서, 각 스레드는 동일한 명령을 실행하지만 서로 다른 데이터를 다룹니다. 그러나 SIMD(Single Instruction, Multiple Data)와 달리, SIMT는 스레드가 독립적인 실행 경로를 가지고 분기 명령에서 분기할 수 있게 합니다.
워프(NVIDIA GPU에서는 32개 스레드의 그룹, AMD GPU에서는 64개 스레드의 그룹)가 분기 명령을 만나면, GPU 하드웨어는 워프 내 각 스레드의 분기 조건을 평가합니다. 모든 스레드가 같은 경로를 취하는 경우(수렴), 워프는 정상적으로 실행을 계속합니다. 그러나 일부 스레드가 다른 경로를 취하는 경우(분기), 워프는 두 개 이상의 서브 워프로 분할되며, 각각 다른 경로를 따릅니다. GPU 하드웨어는 분기된 경로들을 순차적으로 실행하며, 각 서브 워프에서 비활성 스레드를 마스킹합니다. 모든 경로가 완료되면 서브 워프가 다시 수렴하여 동기화된 상태로 실행을 계속합니다.
그림 3.2는 분기 제어 흐름이 있는 SIMT 실행을 보여줍니다.
워프
________________
/ / / / /
/ / / / /
/ / / / /
| | |
| 분기 |
| | |
/ \ / \ / \
/ X \ \
/ / \ \ \
/ \ \
/ \ \
/ \ \
/ \ \
\
```재수렴
그림 3.2: 분기 제어 흐름을 가진 SIMT 실행.
이 분기 처리 메커니즘은 SIMT가 SIMD보다 더 유연한 제어 흐름을 지원할 수 있게 해주지만, 분기가 발생할 때 SIMD 효율성이 저하되는 단점이 있습니다. 프로그래머는 최적의 성능을 달성하기 위해 워프 내의 분기를 최소화하려고 노력해야 합니다.
메모리 계층 구조
GPU는 병렬 워크로드의 높은 대역폭과 낮은 대기 시간 요구 사항을 지원하기 위해 복잡한 메모리 계층 구조를 가지고 있습니다. 메모리 계층 구조는 일반적으로 다음과 같이 구성됩니다:
- 글로벌 메모리: 가장 큰 크기지만 가장 느린 메모리 공간으로, 커널의 모든 스레드에서 접근할 수 있습니다. 글로벌 메모리는 일반적으로 높은 대역폭의 GDDR 또는 HBM 메모리로 구현됩니다.
- 공유 메모리: 블록의 모든 스레드가 공유하는 빠른 온칩 메모리 공간. 공유 메모리는 스레드 간 통신과 블록 내 데이터 공유에 사용됩니다.
- 상수 메모리: 읽기 전용 메모리 공간으로, 모든 스레드에 읽기 전용 데이터를 브로드캐스팅하는 데 사용됩니다.
- 텍스처 메모리: 공간 지역성이 최적화된 읽기 전용 메모리 공간으로, 텍스처 캐시를 통해 접근됩니다. 텍스처 메모리는 주로 그래픽 워크로드에 사용됩니다.
- 로컬 메모리: 각 스레드의 사적인 메모리 공간으로, 레지스터 스필링과 큰 데이터 구조에 사용됩니다. 로컬 메모리는 일반적으로 글로벌 메모리에 매핑됩니다.
메모리 계층 구조의 효과적인 활용은 GPU에서 높은 성능을 달성하는 데 매우 중요합니다. 프로그래머는 공유 메모리 사용을 극대화하고 글로벌 메모리 접근을 최소화하여 메모리 대기 시간과 대역폭 병목 현상을 해결해야 합니다.
그림 3.3은 GPU 메모리 계층 구조를 보여줍니다.
| 공유 |
| 메모리 |
____________
|
____________
| |
| 로컬 |
| 메모리 |
____________그림 3.3: GPU 메모리 계층 구조.
CUDA 프로그래밍 모델 및 API
CUDA(Compute Unified Device Architecture)는 NVIDIA가 일반 목적의 GPU 컴퓨팅을 위해 개발한 병렬 컴퓨팅 플랫폼 및 프로그래밍 모델입니다. CUDA는 C, C++, Fortran과 같은 표준 프로그래밍 언어에 대한 확장 기능을 제공하여 프로그래머들이 병렬 처리를 표현하고 NVIDIA GPU의 계산 능력을 활용할 수 있게 합니다.
CUDA 프로그래밍 모델
CUDA 프로그래밍 모델은 커널이라는 개념을 기반으로 합니다. 커널은 GPU에서 병렬로 실행되는 함수입니다. 프로그래머는 실행할 스레드의 수와 이들을 블록으로 구성하는 방법을 지정합니다.
CUDA는 병렬 프로그래밍을 지원하기 위해 다음과 같은 주요 추상화 개념을 도입했습니다:
- 스레드: CUDA의 기본 실행 단위입니다. 각 스레드는 자체 프로그램 카운터, 레지스터 및 로컬 메모리를 가집니다.
- 블록: 협력하고 동기화할 수 있는 스레드 그룹입니다. 블록 내의 스레드는 동일한 스트리밍 멀티프로세서에서 실행되며 공유 메모리를 통해 통신할 수 있습니다.
- 그리드: 동일한 커널을 실행하는 블록 집합입니다. 그리드는 전체 문제 공간을 나타내며 1, 2 또는 3차원일 수 있습니다.
CUDA는 또한 스레드가 자신을 식별하고 스레드 계층 구조 내에서의 위치를 기반으로 메모리 주소를 계산할 수 있도록 하는 내장 변수(예: threadIdx, blockIdx, blockDim, gridDim)를 제공합니다.
그림 3.4는 CUDA 프로그래밍 모델을 보여줍니다.
그리드
________________
/ / / / /
/ / / / /
/ / / / /
/ / / / /
/__/__/__/__/__/
| | | |
| | 블록 |
| | | |
스레드 스레드 ...그림 3.4: CUDA 프로그래밍 모델.
CUDA 메모리 계층CUDA는 프로그래머에게 GPU 메모리 계층을 노출하여 데이터 배치와 이동에 대한 명시적인 제어를 허용합니다. CUDA의 주요 메모리 공간은 다음과 같습니다:
- 전역 메모리: 커널의 모든 스레드가 접근할 수 있으며, 커널 실행 간에 유지됩니다. 전역 메모리는 지연 시간이 가장 높으며, 일반적으로 큰 데이터 구조에 사용됩니다.
- 공유 메모리: 블록의 모든 스레드가 공유하는 빠른 온칩 메모리. 공유 메모리는 스레드 간 통신과 블록 내 데이터 공유에 사용됩니다.
- 상수 메모리: 모든 스레드에 읽기 전용 데이터를 브로드캐스트하는 데 사용되는 읽기 전용 메모리 공간. 상수 메모리는 캐시되어 지연 시간이 낮습니다.
- 텍스처 메모리: 공간 지역성에 최적화된 읽기 전용 메모리 공간이며, 텍스처 캐시를 통해 접근됩니다. 텍스처 메모리는 주로 그래픽스 워크로드에 사용됩니다.
- 로컬 메모리: 각 스레드의 개인 메모리 공간으로, 레지스터 스필링과 큰 데이터 구조에 사용됩니다. 로컬 메모리는 일반적으로 전역 메모리에 매핑됩니다.
프로그래머는 cudaMalloc, cudaMemcpy, cudaFree와 같은 CUDA 런타임 API를 사용하여 호스트(CPU) 및 디바이스(GPU) 메모리 간에 데이터를 할당하고 전송할 수 있습니다.
그림 3.5는 CUDA 메모리 계층을 보여줍니다.
____________
| |
| Global |
| Memory |
____________
|
____________
| |
| Constant |
| Memory |
____________
|
____________
| |
| Texture |
| Memory |
____________
|
|
____________
| |
| Shared |
| Memory |
____________
|
____________
| |
| Local |
| Memory |
____________그림 3.5: CUDA 메모리 계층.
CUDA 동기화와 협력
CUDA는 스레드 간 협력과 통신을 가능하게 하는 동기화와 협력 원시 연산을 제공합니다:
- 장벽 동기화: __syncthreads()여기는 한국어 번역본입니다:
s() 함수는 모든 스레드가 동일한 지점에 도달할 때까지 기다리는 장벽 역할을 합니다.
- 원자 연산: CUDA는 다른 스레드의 간섭 없이 공유 메모리나 전역 메모리에서 읽기-수정-쓰기 연산을 수행할 수 있는 원자 연산(예: atomicAdd, atomicExch)을 지원합니다.
- 워프 단위 원시: CUDA는 워프 내에서 효율적인 통신과 동기화를 가능하게 하는 워프 단위 기본 연산(예: __shfl, __ballot)을 제공합니다.
동기화와 조정 원시의 적절한 사용은 CUDA에서 올바르고 효율적인 병렬 프로그램을 작성하는 데 필수적입니다.
예제 3.1은 벡터 덧셈을 수행하는 간단한 CUDA 커널을 보여줍니다.
__global__ void vectorAdd(int *a, int *b, int *c, int n) {
// 스레드 ID 계산
int i = blockIdx.x * blockDim.x + threadIdx.x;
// 배열 범위 내에 있는 경우에만 연산 수행
if (i < n) {
c[i] = a[i] + b[i];
}
}
int main() {
int *a, *b, *c;
int n = 1024;
// 호스트 메모리 할당
a = (int*)malloc(n * sizeof(int));
b = (int*)malloc(n * sizeof(int));
c = (int*)malloc(n * sizeof(int));
// 입력 벡터 초기화
for (int i = 0; i < n; i++) {
a[i] = i;
b[i] = i * 2;
}
// 디바이스 메모리 할당
int *d_a, *d_b, *d_c;
cudaMalloc(&d_a, n * sizeof(int));
cudaMalloc(&d_b, n * sizeof(int));
cudaMalloc(&d_c, n * sizeof(int));
// 호스트에서 디바이스로 입력 벡터 복사
cudaMemcpy(d_a, a, n * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(d_b, b, n * sizeof(int), cudaMemcpyHostToDevice);
// 커널 실행
int blockSize = 256;
int numBlocks = (n + blockSize - 1) / blockSize;
vectorAdd<<<numBlocks,blockSize>>>(d_a, d_b, d_c, n);
// 디바이스에서 호스트로 결과 벡터 복사
cudaMemcpy(c, d_c, n * sizeof(int), cudaMemcpyDeviceToHost);
// 디바이스 메모리 해제
cudaFree(d_a);
cudaFree(d_b);
cudaFree(d_c);
// 호스트 메모리 해제
free(a);
free(b);
free(c);
return 0;
}
```Here is the Korean translation of the provided Markdown file, with the comments in the code translated, but the code itself not translated:
# CUDA 코드 예제
```c
int main() {
// 입력 벡터 a, b와 출력 벡터 c의 크기를 설정합니다.
int numElements = 1000000;
size_t size = numElements * sizeof(float);
// 호스트 메모리 할당
float *a, *b, *c;
a = (float *)malloc(size);
b = (float *)malloc(size);
c = (float *)malloc(size);
// 디바이스 메모리 할당
float *d_a, *d_b, *d_c;
cudaMalloc(&d_a, size);
cudaMalloc(&d_b, size);
cudaMalloc(&d_c, size);
// 입력 데이터 초기화
for (int i = 0; i < numElements; i++) {
a[i] = rand() / (float)RAND_MAX;
b[i] = rand() / (float)RAND_MAX;
}
// 입력 데이터를 디바이스로 복사
cudaMemcpy(d_a, a, size, cudaMemcpyHostToDevice);
cudaMemcpy(d_b, b, size, cudaMemcpyHostToDevice);
// 커널 실행 파라미터 설정
int blockSize = 256;
int numBlocks = (numElements + blockSize - 1) / blockSize;
// 벡터 덧셈 커널 실행
vectorAdd<<<numBlocks, blockSize>>>(d_a, d_b, d_c, numElements);
// 결과를 호스트로 복사
cudaMemcpy(c, d_c, size, cudaMemcpyDeviceToHost);
// 메모리 해제
free(a);
free(b);
free(c);
cudaFree(d_a);
cudaFree(d_b);
cudaFree(d_c);
return 0;
}이 CUDA 코드는 numBlocks 개의 블록과 각 블록당 blockSize 개의 스레드로 vectorAdd 커널을 실행합니다. 커널은 입력 벡터 a와 b의 원소 단위 덧셈을 수행하고 그 결과를 벡터 c에 저장합니다. <<<...>>>
구문은 커널을 실행할 때 그리드와 블록의 차원을 지정하는 데 사용됩니다.
CUDA 스트림과 이벤트
CUDA 스트림과 이벤트는 커널 실행과 메모리 작업의 동시 실행 및 동기화를 가능하게 합니다:
- 스트림: 순서대로 실행되는 일련의 작업(커널 실행, 메모리 복사)입니다. 다른 스트림들은 동시에 실행될 수 있어 계산과 메모리 전송을 겹쳐서 수행할 수 있습니다.
- 이벤트: 특정 작업의 완료를 기록하는 마커입니다. 이벤트는 동기화와 타이밍 목적으로 사용될 수 있습니다.
스트림과 이벤트를 활용하면 프로그래머는 계산과 메모리 전송을 겹쳐서 수행함으로써 CUDA 애플리케이션의 성능을 최적화할 수 있습니다.
예제 3.2는 CUDA 스트림을 사용하여 커널 실행과 메모리 전송을 겹쳐서 수행하는 방법을 보여줍니다.
// 두 개의 스트림 생성
cudaStream_t stream1, stream2;
cudaStreamCreate(&stream1);
cudaStreamCreate(&stream2);
// 입력 데이터를 비동기적으로 디바이스로 복사
cudaMemcpyAsync(d_a, a, size, cudaMemcpyHostToDevice, stream1);
cudaMemcpyAsync(d_b, b, size, cudaMemcpyHostToDevice, stream2);
// 다른 스트림에서 커널 실행
kernelA<<<blocks, threads, 0, stream1>>>(d_a);
kernelB<<<blocks, threads, 0, stream2>>>(d_b);
// 결과를 비동기적으로 호스트로 복사
cudaMemcpyAsync(a, d_a, size, cudaMemcpyDeviceToHost, stream1);
cudaMemcpyAsync(b, d_b, size, cudaMemcpyDeviceToHost, stream2);
// 스트림 동기화
cudaStreamSynchronize(stream1);
cudaStreamSynchronize(stream2);이 예제에서는 두 개의 CUDA 스트림을 생성합니다. 각 스트림을 사용하여 입력 데이터를 비동기적으로 디바이스로 복사합니다. 그런 다음 다른 스트림에서 커널을 실행하고, 결과를 비동기적으로 호스트로 복사합니다. 마지막으로 두 스트림을 동기화합니다.다음은 제공된 마크다운 파일의 한국어 번역입니다. 코드에 대해서는 주석만 번역하고 코드 자체는 번역하지 않습니다.
OpenCL 프레임워크
OpenCL(Open Computing Language)은 CPU, GPU, FPGA 및 기타 가속기와 같은 이기종 플랫폼에서 병렬 프로그래밍을 위한 무료 표준입니다. OpenCL은 개발자가 이식 가능하고 효율적인 병렬 코드를 작성할 수 있는 통합 프로그래밍 모델과 API 세트를 제공합니다.
OpenCL 프로그래밍 모델
OpenCL 프로그래밍 모델은 CUDA와 유사하지만 용어와 추상화에서 몇 가지 차이가 있습니다:
- 커널: OpenCL 디바이스에서 많은 수의 작업 항목(스레드)에 의해 병렬로 실행되는 함수.
- 작업 항목: OpenCL에서 실행의 기본 단위로, CUDA의 스레드와 유사합니다.
- 작업 그룹: 로컬 메모리를 통해 데이터를 동기화하고 공유할 수 있는 작업 항목의 집합. CUDA의 스레드 블록과 유사합니다.
- NDRange: 커널 실행을 위한 색인 공간과 작업 항목 조직을 정의합니다. 1차원, 2차원 또는 3차원일 수 있습니다.
OpenCL은 CUDA와 유사한 계층적 메모리 모델도 정의합니다:
- 글로벌 메모리: 모든 작업 그룹의 모든 작업 항목이 접근할 수 있는 메모리, CUDA의 글로벌 메모리와 유사합니다.
- 로컬 메모리: 작업 그룹 내의 모든 작업 항목이 공유할 수 있는 메모리, CUDA의 공유 메모리와 유사합니다.
- 개인 메모리: 단일 작업 항목에 전용인 메모리, CUDA의 레지스터와 유사합니다.
- 상수 메모리: 모든 작업 항목이 읽기 전용으로 접근할 수 있는 메모리.
OpenCL 커널은 OpenCL 런타임에 의해 런타임에 컴파일됩니다. 호스트 프로그램은 사용 가능한 OpenCL 디바이스를 쿼리하고, 적절한 디바이스를 선택하고, 컨텍스트를 생성하고, 해당 디바이스에 대해 커널을 빌드할 수 있습니다. 이를 통해 OpenCL 애플리케이션은 다양한 하드웨어 플랫폼에서 높은 이식성을 가질 수 있습니다.
예제 3.3은 예제 3.1의 CUDA 예제와 유사한 OpenCL 커널을 보여줍니다.Here is the Korean translation of the provided markdown file:
__kernel void vector_add(
__global const int *a,
__global const int *b,
__global int *c,
int n) {
int i = get_global_id(0);
if (i < n) {
c[i] = a[i] + b[i];
}
}__kernel 키워드는 OpenCL 커널 함수를 정의합니다. __global 키워드는 포인터가 전역 메모리를 가리킨다는 것을 나타냅니다. get_global_id 함수는 현재 워크-항목의 전역 인덱스를 반환하며, 이는 입력 및 출력 벡터의 메모리 주소를 계산하는 데 사용됩니다.
GPU 아키텍처에 알고리즘 매핑하기
GPU 아키텍처에 알고리즘을 효율적으로 매핑하는 것은 높은 성능을 달성하는 데 중요합니다. 핵심 고려 사항은 다음과 같습니다:
-
충분한 병렬성 노출: 알고리즘을 GPU의 병렬 처리 기능을 완전히 활용할 수 있도록 많은 세부 스레드로 분해해야 합니다.
-
분기 발산 최소화: 워프/웨이브프론트 내의 발산 제어 흐름은 직렬화와 SIMD 효율성 저하로 이어질 수 있습니다. 알고리즘은 분기 발산을 최소화하도록 구조화되어야 합니다.
-
메모리 계층 활용: 전역 메모리 액세스는 비용이 큽니다. 알고리즘은 공유 메모리와 레지스터의 사용을 극대화하여 전역 메모리 액세스를 줄여야 합니다. 또한 데이터는 병합된 메모리 액세스가 가능하도록 메모리에 배치되어야 합니다.
-
계산과 메모리 액세스의 균형: 알고리즘은 산술 연산과 메모리 연산의 비율이 높아야 메모리 대기 시간을 효과적으로 숨길 수 있고 높은 계산 처리량을 달성할 수 있습니다.
-
호스트-디바이스 데이터 전송 최소화: 호스트와 디바이스 메모리 간 데이터 전송은 느립니다. 알고리즘은 가능한 한 GPU에서 최대한 많은 계산을 수행하여 이러한 전송을 최소화해야 합니다.
GPU 커널 개발 시 일반적으로 사용되는 병렬 알고리즘 설계 패턴은 다음과 같습니다:
-
맵: 각 스레드가 다른 데이터 요소에 동일한 작업을 수행하여 대용량 데이터 세트의 간단한 병렬 처리를 가능하게 합니다.
-
리듀스: 병렬 리덕션은 큰 입력 데이터 세트에서 단일 값(예: 합계, 최대값)을 효율적으로 계산하는 데 사용됩니다.스레드는 로컬 감소를 수행하고, 이를 결합하여 최종 결과를 생성합니다.
-
스캔: 접두사 합계로도 알려진 스캔은 배열의 요소에 대한 실행 합계를 계산하는 데 사용됩니다. 효율적인 병렬 스캔 알고리즘은 많은 GPU 가속 애플리케이션의 핵심 구성 요소입니다.
-
스텐실: 각 스레드는 인접한 데이터 요소를 기반으로 값을 계산합니다. 스텐실 계산은 과학 시뮬레이션 및 이미지 처리 애플리케이션에서 일반적입니다.
-
수집/산포: 스레드는 전역 메모리의 임의 위치에서 읽기(수집) 또는 쓰기(산포)를 수행합니다. 효율성을 위해서는 데이터 레이아웃과 액세스 패턴의 주의가 필요합니다.
결론
CUDA와 OpenCL과 같은 GPU 프로그래밍 모델은 현대 GPU의 병렬 처리 기능을 개발자에게 노출하여 다양한 애플리케이션을 가속화할 수 있도록 합니다. 이러한 프로그래밍 모델은 세부적인 병렬 워크로드를 GPU 하드웨어에 효율적으로 매핑할 수 있는 추상화를 제공합니다.
GPU 코드의 고성능을 작성하기 위해서는 이러한 프로그래밍 모델에서 제공하는 실행 모델, 메모리 계층 구조 및 동기화 원시 요소에 대한 이해가 필수적입니다. 개발자는 스레드 구성, 분기 분산, 메모리 액세스 패턴 및 알고리즘 설계와 같은 요소를 신중하게 고려해야 합니다.
GPU 아키텍처가 지속적으로 발전함에 따라, 프로그래밍 모델과 도구도 새로운 하드웨어 기능과 기능을 효과적으로 활용할 수 있도록 발전해야 합니다. 프로그래밍 언어 설계, 컴파일러 최적화 및 자동 조정과 같은 분야의 지속적인 연구가 이기종 컴퓨팅 시대에 프로그래머 생산성과 성능 이식성을 개선하는 데 필수적일 것입니다.