1

ブロックごとの共有メモリ使用量に基づいて、 CUDA共有メモリとブロック実行で実行状態をクリアしたいと考えています。

州

ブロックごとに 48KB の共有メモリと 15 のストリーミング マルチプロセッサを備えた GTX480 nvidia カードをターゲットにしています。したがって、15 ブロックでカーネルを宣言すると、それぞれが 48KB の共有メモリを使用し、他の制限 (レジスタ、ブロックあたりの最大スレッド数など) に達しません。この場合、同じブロックのワープ間のスケジューリングのみが必要です。

質問

したがって、私の誤解のシナリオは次のとおり
です。各 SM に 2 つのブロックが存在するように、30 ブロックのカーネルを呼び出します。各 SM のスケジューラは、異なるブロックからのワープを処理する必要があります。ただし、共有メモリの全量 (SM あたり 48KB) を使用するため、一方のブロックの実行が終了した場合にのみ、他方のブロックのワープが SM 上で実行されます。これが起こらず、異なるブロックのワープが同じ SM での実行をスケジュールしている場合、1 つのブロックが共有メモリ内の別のブロックからロードされた値を読み取ることができるため、結果が間違っている可能性があります。私は正しいですか？

4

1 に答える 1

2

これについて心配する必要はありません。あなたが正しく言ったように、使用される共有メモリの量が原因で SM ごとに 1 つのブロックしか収まらない場合、一度に 1 つのブロックのみがスケジュールされます。そのため、共有メモリのオーバーコミットによってメモリが破損する可能性はありません。

ところで、パフォーマンス上の理由から、通常は SM ごとに少なくとも 2 つのブロックを実行することをお勧めします。

  • __syncthreads() の間、ブロックからのワープがまだ実行可能である可能性があるため、SM は不必要にアイドル状態になる可能性があります。
  • 同じブロックのワープは密結合で実行される傾向があるため、すべてのワープがメモリを待機する場合と、すべてのワープが計算を実行する場合があります。ブロックが増えると、これはさらに良くなり、全体的なリソース使用率が向上します。

もちろん、SM ごとに複数のブロックを実行するよりも、ブロックごとの共有メモリが多いほど高速になる理由があるかもしれません。

于 2012-09-29T11:04:52.867 に答える