cuda-tutorial

2.1 ホストとデバイス

まずはじめに、GPUプログラミングを行う上で必要な概念である、ホストとデバイスについて述べていく。といってもそんなに難しくない。

そして、CUDAのコードのうち、CPU側の処理について書かれている部分をホストコード、GPU側の処理について書かれている部分をデバイスコードと言う。[1]

以下、典型的なベクトルの加算を表現したGPUのコードを記す。

#include <cuda.h>
#include <stdio.h>
#include <stdlib.h>

//ブロックの数
#define NB 16
//1ブロックあたりのスレッド数
#define NT 512

//デバイス側の関数: GPU側で動く並列処理用の関数
__global__ void initVec(float *vec_dev, float a){
  unsigned int i_global = blockIdx.x * blockDim.x + threadIdx.x;  
  vec_dev[i_global] = a * (float)i_global;
}
__global__ void addVec(float *vec1_dev, float *vec2_dev){
  unsigned int i_global = blockIdx.x * blockDim.x + threadIdx.x;
  vec1_dev[i_global] += vec2_dev[i_global];
}

//ホスト側の関数: CPU側で動く逐次処理用の関数
int main(){
  //以下逐次処理
  float *vec;
  float *vec1_dev, *vec2_dev;
  
  vec = (float*)malloc(NB * NT);
  cudaMalloc((void**)vec1_dev, NB * NT * sizeof(float));
  cudaMalloc((void**)vec2_dev, NB * NT * sizeof(float));
  
  //以下並列処理
  initVec<<<NB,NT>>>(vec1_dev, 1.);
  initVec<<<NB,NT>>>(vec2_dev, 2.);
  addVec<<<NB,NT>>>(vec1_dev, vec2_dev);
  
  //以下再び逐次処理
  ////デバイスからホストへのデータ通信
  cudaMemcpy(vec, vec1_dev, NB * NT * sizeof(float), cudaMemcpyDeviceToHost);
  
  for(unsigned int i = 0; i < NB * NT; i++){
    printf("%d: %f", i, vec[i]);
  }
  
  free(vec);
  cudaFree(vec1_dev);
  cudaFree(vec2_dev);
  
  return 0;
}

上のコードを例に取ると、__global__ void initVec() , __global__ void addVec() がデバイスコード、対して int main() がホストコードである。 ホストコードの部分はほぼC/C++と遜色ないが、 initVec<<<NB,NT>>>(); のように並列処理を呼ぶ部分や cudaMemcpy() のようにデバイス・ホストの通信をする部分があるのが純粋なC/C++と異なる。

引用、紹介

[1] Cyril Zeller, “CUDA C/C++ Basics Supercomputing 2011 Tutorial”