c++ - 違い - 共有メモリ windows 確認方法




共有メモリの割り当て (4)

CUDAは動的共有メモリの割り当てをサポートしています。 次のようにカーネルを宣言した場合:

__global__ void Kernel(const int count)
{
    extern __shared__ int a[];
}

カーネル起動の第3引数として必要なバイト数を渡します

Kernel<<< gridDim, blockDim, a_size >>>(count)

実行時にaをサイズ変更することができます。 ランタイムは、ブロックごとに1つの動的に宣言された割り当てのみをサポートしていることに注意してください。 もっと必要な場合は、その単一の割り当て内のオフセットへのポインタを使用する必要があります。 また、共有メモリが32ビットワードを使用し、共有メモリ割り当てのタイプに関係なく、すべての割り当てが32ビットワードでアライメントされている必要があることに注意してください。

私は、定数パラメータを使用してエラーを取得して共有メモリを割り当てようとしています。 私のカーネルは次のようになります:

__global__ void Kernel(const int count)
{
    __shared__ int a[count];
}

と私はエラーが発生していると言っている

エラー:式は定数値を持つ必要があります

カウントはconstです! なぜこのエラーが発生するのですか? そして、どうすればこの問題を回避できますか?


「CUDA Cプログラミングガイド」から:実行コンフィギュレーションは、次の形式の式を挿入することによって指定されます。

<<<Dg, Db, Ns, S>>>

ここで:

  • Dgdim3型で、グリッドのサイズとサイズを指定します...
  • Dbdim3型で、各ブロックのサイズとサイズを指定します...
  • Nssize_t型であり、静的に割り当てられたメモリに加えて、この呼び出しのブロックごとに動的に割り当てられる共有メモリのバイト数を指定します。 この動的に割り当てられたメモリは、 __shared__で述べたように外部配列として宣言された変数のいずれかによって使用されます。 Nsは省略可能な引数で、デフォルトは0です。
  • SはcudaStream_t型で、関連するストリームを指定しています...

したがって、動的パラメータNsを使用することで、このカーネルにいくつの共有変数が存在していても、ユーザは1つのカーネル機能が使用できる共有メモリの合計サイズを指定できます。


あなたはこのような共有変数を宣言することはできません..

__shared__ int a[count];

配列の最大サイズについて十分に分かっていれば、次のように直接宣言できます。

__shared__ int a[100];

この場合は、共有メモリをブロックに固定する(完全には利用されない)ため、プログラムにいくつのブロックがあるか心配する必要があります。グローバルメモリ(高いレイテンシ)でコンテキストを切り替えるため、パフォーマンスが低下します...

宣言するには、この問題に対する素晴らしい解決策があります

extern __shared__ int a[];

カーネルをメモリから呼び出しながらメモリを割り当てる

Kernel<<< gridDim, blockDim, a_size >>>(count)

カーネルで割り当てるよりも多くのメモリをブロック内で使用している場合は、予期しない結果が出ることになるので、ここでも気にする必要があります。


オプション1:共有メモリを定数値で宣言する( constと同じではない)

__global__ void Kernel(int count_a, int count_b)
{
    __shared__ int a[100];
    __shared__ int b[4];
}

オプション2:カーネル起動コンフィグレーションで共有メモリを動的に宣言する:

__global__ void Kernel(int count_a, int count_b)
{
    extern __shared__ int *shared;
    int *a = &shared[0]; //a is manually set at the beginning of shared
    int *b = &shared[count_a]; //b is manually set at the end of a
}

sharedMemory = count_a*size(int) + size_b*size(int);
Kernel <<<numBlocks, threadsPerBlock, sharedMemory>>> (count_a, count_b);

注:動的共有メモリへのポインタにはすべて同じアドレスが与えられます。 2つの共有メモリアレイを使用して、共有メモリに2つのアレイを手動で設定する方法を示します。





gpu-shared-memory