c++ - nvidia cuda カーネルで配列を作成する

Question

こんにちは、nvidia cudaカーネル内で次のことが可能かどうか知りたいだけです

__global__ void compute(long *c1, long size, ...)
{
  ...
  long d[1000];
  ...
}

または次の

__global__ void compute(long *c1, long size, ...)
{
  ...
  long d[size];
  ...
}

Answer 1

最初の例は実行できますが、2 番目の例は試していません。

ただし、それを助けることができる場合は、これを行わないようにプログラムを再設計することをお勧めします。カーネルに 4000 バイトのメモリを割り当てたくありません。すべてをレジスタに収めることができないため、これにより CUDA ローカルメモリが大量に使用されます。CUDA ローカル メモリは低速です (400 サイクルのメモリ レイテンシ)。

Answer 2

＃1を実行できますが、これはすべてのスレッドで実行されることに注意してください。

カーネルランタイムでの動的メモリ割り当てがサポートされていないため、2番目のスニペットは機能しません。

Answer 3

カーネルの起動時に、共有メモリを動的に割り当てることができます。

__global__ void compute(long *c1, long size, ...)
 {
  ...
   extern __shared__ float shared[];
  ...
 }

compute <<< dimGrid, dimBlock, sharedMemSize >>>( blah blah );

CUDA プログラミング ガイド:

配列のサイズは起動時に決定されます (セクション 4.2.3 を参照)。

Answer 4

カーネル ランタイムでの動的メモリ割り当てがサポートされています。sdk の例、新しい削除を確認してください。