2

「CUDAbyexample」という本からCUDAを学んでいます。第4章には、Juliaフラクタルを生成するデモがあります。ショーケースはCPUとGPUの両方のバージョンを示しています。両方の場合の実行速度を確認する時間を追加することにしました。驚いたことに、CPUバージョンはGPUの3倍の速度で実行されます。

CPU Julia生成の合計時間：

745ミリ秒。

GPU Julia生成の合計時間：

2456ミリ秒。

では、何が起こっているのでしょうか？少なくともCUDAカーネルコードから、実行が並列であり、それぞれが1000x1000解像度の最終画像のピクセルを計算する1000ブロックに分散されていることは明らかです。

実装のソースコードは次のとおりです。

 #define N 10
 #define DIM 1000
 typedef unsigned char byte;

struct cuComplex {
   float   r;
   float   i;
   __host__ __device__ cuComplex( float a, float b ) : r(a), i(b)  {}
   __host__  __device__ float magnitude2( void ) {
          return r * r + i * i;
   }
   __host__ __device__ cuComplex operator*(const cuComplex& a) {
        return cuComplex(r*a.r - i*a.i, i*a.r + r*a.i);
   }
    __host__ __device__ cuComplex operator+(const cuComplex& a) {
        return cuComplex(r+a.r, i+a.i);
   }
};

 __device__ int juliaGPU(int x , int y){
    const float scale =1.3;
    float jx = scale * (float)(DIM/2 -x)/(DIM/2);
    float jy=  scale *(float)(DIM/2 -y)/(DIM/2);

    cuComplex c(-0.8 ,0.156);
    cuComplex a(jx ,jy);
    int i = 0;
    for(i=0; i <200;i++){
        a = a * a +c;
        if(a.magnitude2() >1000){

            return 0;
        }
    }
    return 1;

 }

 __global__ void kernelGPU(byte *ptr){
    int x = blockIdx.x;
    int y = blockIdx.y;
    int offset =x + y * gridDim.x;

    int juliaValue =juliaGPU(x , y);
    ptr[offset * 4 + 0]=255 * juliaValue;
    ptr[offset * 4 + 1]=0;
    ptr[offset * 4 + 2]=0;
    ptr[offset * 4 + 3]=255 ;
}


 struct DataBlock {
    unsigned char   *dev_bitmap;
};
 void juliaGPUTestSample(){
 DataBlock   data;
CPUBitmap bitmap(DIM,DIM);
byte *dev_bitmap; //memory on GPU 
HANDLE_ERROR(cudaMalloc((void**)&dev_bitmap , bitmap.image_size()) );
data.dev_bitmap =dev_bitmap;
dim3 grid(DIM,DIM);
int starTime=glutGet(GLUT_ELAPSED_TIME);

kernelGPU<<<grid ,1 >>>(dev_bitmap);
HANDLE_ERROR(cudaMemcpy(bitmap.get_ptr() , dev_bitmap ,bitmap.image_size() ,cudaMemcpyDeviceToHost ) );
int endTime=glutGet(GLUT_ELAPSED_TIME)-starTime;
printf("Total time %d\n:" ,endTime);
HANDLE_ERROR(cudaFree(dev_bitmap));

bitmap.display_and_exit();
 }

 int main(void){
juliaGPUTestSample();
return 1;

 }

これがCPUバージョンです：

///"cuComplex"構造体は上から同じです。

int julia (int x , int y){

const float scale = 1.3;
float jx = scale * (float)(DIM/2 -x)/(DIM/2);
float jy = scale * (float)(DIM/2 -y)/(DIM/2);

cuComplex c(-0.8 ,0.156);
cuComplex a(jx ,jy);

int i = 0;
for(i=0; i <200;i++){

    a = a * a +c;
    if(a.magnitude2() >1000){

        return 0;
    }
}

return 1;

}

void kernel(unsigned char *ptr){

for(int y = 0 ; y <DIM ;++y){
    for(int x = 0 ; x <DIM ; ++x){
        int offset =x + y * DIM;
        int juliaValue = julia(x , y);

        ptr[offset * 4 + 0 ] = juliaValue * 125;
        ptr[offset * 4 + 1 ] = juliaValue * x;
        ptr[offset * 4 + 2 ] = juliaValue * y;
        ptr[offset * 4 + 3 ] = 255 ;
    }
}

}
void juliaCPUTestSample(){

CPUBitmap bitmap(DIM ,DIM);
unsigned char *ptr = bitmap.get_ptr();
int starTime=glutGet(GLUT_ELAPSED_TIME);

kernel(ptr);

int endTime=glutGet(GLUT_ELAPSED_TIME)-starTime;
printf("Total time %d\n:" ,endTime);
bitmap.display_and_exit();

}

システム構成の更新：

Windows764ビット

CPU-Intel i7 -3770CPU 3.40GHz、16GB RAM

GPU-NVidia Quadro 4000

4

1 に答える 1

9

他の人はこれに気づきました。

まず、CPUとGPUのパフォーマンスの比較について話すときは、ハードウェアプラットフォームとソフトウェアを含むシステム構成について言及することをお勧めします。たとえば、コアi72.60GHzクアッドコアCPUとquadro1000MGPUを搭載し、RHEL6.2とcuda5.0を実行しているHPラップトップでコードを実行したところ、GPUで438、CPUで441のスコアが得られました。

第二に、そしてもっと重要なことに、その本のジュリアのサンプルはCUDAコーディングの比較的初期の例であるため、実際には最大のパフォーマンスを対象としているのではなく、これまでに説明した概念を説明しています。その本と他のさまざまなCUDAチュートリアル資料は、ブロックレベルでCUDAを使用した並列プログラミングを紹介することから始まります。これの表示はここにあります：

kernelGPU<<<grid ,1 >>>(dev_bitmap);

カーネル起動パラメーターは、いくつかの数（この場合は合計100万ブロック）<<<grid, 1>>>のグリッドが起動され、各ブロックが単一のスレッドを持つことを示します。これにより、スレッドブロックごとにスレッドを完全に補完するグリッドを起動する場合と比較して、たとえば、FermiクラスのGPUの能力がすぐに1/32に減少します。FermiクラスGPUの各SMには32個のスレッドプロセッサがあり、すべてロックステップで実行されます。16スレッドしかないブロックを起動すると、16スレッドプロセッサがコードを実行し、他の16スレッドプロセッサは何もしません（つまり、何も役に立ちません）。したがって、1つのスレッドのみを含むスレッドブロックは、32のスレッドプロセッサのうち1つのみを使用し、残りの31はアイドル状態になります。 grid

したがって、この特定のコードサンプルは、GPUの完全な並列機能を利用するように適切に設計されていません。この本でのCUDAの概念の説明が比較的早いことを考えると、これは理解できます。このコードをベンチマークしたり、GPUで高速コードを作成する方法の正当な表現として使用したりすることは著者の意図ではなかったと思います。

この1/32の係数に照らして、システムではCPUがわずか3倍高速であり、私のシステムではCPUとGPUのスループットが同等であるという考え（これらのいずれも特に高性能のCUDA GPUである可能性が高い） GPUをかなり良い光で示していると思います。GPUは、その機能の約97％を未使用にして、この戦いを戦っています。

于 2013-02-17T23:11:22.637 に答える