1

次の CUDA コードのグローバル メモリ アクセス数を取得するために nvprof を使用しています。カーネル内のロード数は 36 (d_In 配列へのアクセス) であり、カーネル内のストア数は 36+36 (d_Out 配列および d_rows 配列へのアクセス) です。そのため、グローバル メモリ ロードの合計数は 36 で、グローバル メモリ ストアの数は 72 です。ただし、nvprof CUDA プロファイラーでコードをプロファイリングすると、次のようにレポートされます。 (CGMA)比)

      1                gld_transactions        Global Load Transactions           6           6           6
      1                gst_transactions       Global Store Transactions          11          11          11
      1            l2_read_transactions            L2 Read Transactions         133         133         133
      1           l2_write_transactions           L2 Write Transactions          24          24          24


#include <stdio.h>
#include "cuda_profiler_api.h"
__constant__ int crows;

__global__ void kernel(double *d_In, double *d_Out, int *d_rows){
        int tx=threadIdx.x;
        int bx=blockIdx.x;
        int n=bx*blockDim.x+tx;
        if(n < 36){
                d_Out[n]=d_In[n]+1;
                d_rows[n]=crows;
        }
        return;
}

int main(int argc,char **argv){

     double I[36]={1,5,9,2,6,10,3,7,11,4,8,12,13,17,21,14,18,22,15,19,23,16,20,24,25,29,33,26,30,34,27,31,35,28,32,36};

     double *d_In;
     double *d_Out;
     int *d_rows;

     double Iout[36];
     int rows=5;
     int h_rows[36];

     cudaMemcpyToSymbol(crows,&rows,sizeof(int));
     cudaMalloc(&d_In,sizeof(double)*36);
     cudaMalloc(&d_Out,sizeof(double)*36);
     cudaMalloc(&d_rows,sizeof(int)*36);

     cudaMemcpy(d_In,I,sizeof(double)*36,cudaMemcpyHostToDevice);

     dim3 dimGrid(4,1,1);
     dim3 dimBlock(10,1,1);

     cudaProfilerStart();
     kernel<<<dimGrid,dimBlock>>>(d_In,d_Out,d_rows);
     cudaProfilerStop();

     cudaMemcpy(Iout,d_Out,sizeof(double)*36,cudaMemcpyDeviceToHost);
      cudaMemcpy(h_rows,d_rows,sizeof(int)*36,cudaMemcpyDeviceToHost);


    int i;
     for(i=0;i<36;i++)
       printf("%f %d\n",Iout[i],h_rows[i]);


}

誰かが私を助けることができますか？ありがとうございました

4

1 に答える 1

2

「誰か助けてくれませんか」よりも具体的な質問をするのが通例です。示されているコードには浮動小数点演算 (+、* など) がないため、計算する CGMA はありません (ゼロです)。

メモリ トランザクションに関して、コードには 4 つのスレッドブロックがあります。

 dim3 dimGrid(4,1,1);

各スレッドブロックは、個別のマルチプロセッサで実行できます。各ブロックに 10 個のスレッドがあります。次のコード行:

            d_Out[n]=d_In[n]+1;

スレッドを処理するために、少なくとも 1 つのグローバル ロード トランザクション ( d_In) と 1 つのグローバル ストア トランザクション ( ) を生成します。d_Out4 番目のブロックには、アクティブなスレッドのグローバル インデックス ( n) が 30 ～ 35 のスレッドがあります。このブロックが上記のコード行を実行すると、2 つのグローバル ロード トランザクションと 2 つのグローバル ストア トランザクションが生成されます。これは、スレッドが要求を処理するために 2 つのキャッシュラインが必要なためです。したがって、この 1 行のコードで 5 つのグローバル ロード トランザクションと 5 つのグローバル ストア トランザクションが生成される可能性があります。

同様の理由で、次のコード行:

            d_rows[n]=crows;

5 つの追加のグローバル ストア トランザクションを生成する可能性があります。したがって、プロファイラーの出力は次のとおりです。

  1                gld_transactions        Global Load Transactions           6           6           6
  1                gst_transactions       Global Store Transactions          11

6 つのグローバル ロード トランザクションのうち 5 つ、11 のグローバル ストア トランザクションのうち 10 について説明したと思います。願わくば、これでこれらの数字の由来を理解していただければ幸いです。

于 2014-09-23T14:06:17.833 に答える