7

array があるとしint * dataます。各スレッドはこの配列の 1 つの要素にアクセスします。この配列はすべてのスレッド間で共有されるため、グローバル メモリ内に保存されます。

テストカーネルを作成しましょう:

 __global__ void test(int *data, int a, int b, int c){ ... }

dataを使用してこの配列にメモリを割り当てたため、配列がグローバルメモリにあることは確かですcudaMalloc。他の変数については、メモリを割り当てずに整数をカーネル関数に直接渡す例をいくつか見てきました。私の場合、そのような変数はa bとcです。

私が間違っていなければ、cudaMalloc3 つの整数ごとに 4 バイトを割り当てるために直接呼び出すことはしませんが、CUDA が自動的にそれを行うため、最終的に変数a bとcグローバル メモリに割り当てられます。

現在、これらの変数は補助的なものにすぎず、スレッドはそれらを読み取るだけで、他には何もありません。

私の質問は、これらの変数を共有メモリに転送した方がよいでしょうか?

たとえば、スレッドを10含むブロックがある場合、各ブロックの共有メモリに数値を格納するには、バイトの読み取りが必要になると思います。102410*3 = 304

共有メモリがなく、各スレッドがこれら 3 つの変数すべてを 1 回読み取る必要がある場合、グローバル メモリの読み取りの合計量1024*10*3 = 30720は非常に非効率的になります。

ここに問題があります。私はCUDAに少し慣れていないので、各スレッドがグローバルメモリからこれらの変数を読み取ってロードすることなく、変数のメモリと各ブロックの共有メモリに転送できるかどうかわかりa bませcんしたがって、最終的にグローバル メモリ読み取りの合計量は であり1024*10*3 = 30720、 ではありません10*3 = 30。

次のWeb サイトに、この例があります。

 __global__ void staticReverse(int *d, int n)
 {
    __shared__ int s[64];
    int t = threadIdx.x;
    int tr = n-t-1;
    s[t] = d[t];
    __syncthreads();
   d[t] = s[tr];
 }

ここで、各スレッドは共有変数内に異なるデータをロードしますs。したがって、各スレッドは、そのインデックスに従って、指定されたデータを共有メモリ内にロードします。

私の場合、変数のみを共有メモリにロードしたいと考えてa bいcます。これらの変数は常に同じで、変更されないため、スレッド自体とは何の関係もありません。これらは補助的なものであり、各スレッドが何らかのアルゴリズムを実行するために使用されています。

この問題にどのようにアプローチすればよいですか？total_amount_of_blocks*3グローバルメモリの読み取りのみを行うことでこれを達成することは可能ですか?

4

1 に答える 1

13

GPU ランタイムは、ユーザーが何もする必要なく、既にこれを最適に実行しています (そして、CUDA で引数の受け渡しがどのように機能するかについてのあなたの仮定は正しくありません)。これは現在起こっていることです：

  • 計算機能 1.0/1.1/1.2/1.3 デバイスでは、カーネル引数はランタイムによって共有メモリに渡されます。
  • コンピューティング機能 2.x/3.x/4.x/5.x/6.x デバイスでは、カーネル引数は、ランタイムによって予約済み定数メモリ バンク (ブロードキャスト専用のキャッシュを持つ) に渡されます。

したがって、仮想カーネルでは

__global__ void test(int *data, int a, int b, int c){ ... }

data、a、b、およびcはすべて、共有メモリまたは定数メモリ (GPU アーキテクチャに応じて) の各ブロックに値によって自動的に渡されます。あなたが提案したことを行う利点はありません。

于 2013-05-26T06:21:35.727 に答える