2

ローカルメモリについて言及しているCUDAのドキュメントを読みました。(これは主に初期のドキュメントです。)device-propertiesは、local-memサイズ(スレッドごと)を報告します。「ローカル」メモリとはどういう意味ですか?「ローカル」メモリとは何ですか?「ローカル」メモリはどこにありますか?「ローカル」メモリにアクセスするにはどうすればよいですか?__device__記憶ですよね?

device-propertiesは、グローバル、共有、および一定のmemサイズも報告します。これらのステートメントは正しいですか: グローバルメモリは__device__メモリです。グリッドスコープとグリッド(カーネル)の存続期間があります。 コンスタントメモリは__device__ __constant__メモリです。グリッドスコープとグリッド(カーネル)の有効期間があります。 共有メモリは__device__ __shared__メモリです。単一のブロックスコープと(スレッドの)そのブロックの存続期間があります。

共有メモリはSMメモリだと思います。つまり、その単一のSMのみが直接アクセスできるメモリ。かなり限られたリソース。SMは一度にたくさんのブロックを割り当てられていませんか?これは、SMが異なるブロックの実行をインターリーブできる(またはインターリーブできない)ことを意味しますか?つまり、ブロック* A *スレッドをストールするまで実行します。次に、ブロック* B *スレッドを停止するまで実行します。次に、再びブロック* A *スレッドにスワップバックします。またはSMは、ブロック* A *のスレッドのセットを、ストールするまで実行しますか。次に、別のブロック* A *スレッドのセットがスワップインされます。このスワップは、ブロック* A *が使い果たされるまで続きます。それからそしてその時だけ、作業はブロック* Bで始まります*。共有メモリのためにお願いします。1つのSMが2つの異なるブロックからコードをスワップインしている場合、SMはどのようにして共有メモリチャンクをすばやくスワップイン/スワップアウトしますか?(後のsenerioは真であり、共有メモリスペースのスワップイン/スワップアウトはないと思います。Block* A *は完了するまで実行され、次にblock * B *が実行を開始します。注:block * A *は異なる場合があります。ブロックよりカーネル* B *。)

4

1 に答える 1

5

CUDA Cプログラミングガイドのセクション5.3.2.2から、ローカルメモリがいくつかの状況で使用されていることがわかります。

  • 各スレッドにいくつかの配列があるが、コンパイル時にそれらのサイズがわからない場合(したがって、それらがレジスターに収まらない可能性があります)
  • コンパイル時に配列のサイズがわかっていて、このサイズがレジスタメモリに対して大きすぎる場合(これは大きな構造体でも発生する可能性があります)
  • カーネルがすでにすべてのレジスタメモリを使い果たしている場合(したがって、レジスタをn intsで埋めた場合、n+1thintはローカルメモリに入ります)-この最後のケースはレジスタのスピルであり、次の理由で回避する必要があります。

「ローカル」メモリは実際にはグローバルメモリ空間に存在します。つまり、メモリへの読み取りと書き込みは、レジスタや共有メモリに比べて比較的低速です。カーネルで、レジスタに収まらず、共有メモリではなく、グローバルメモリとして渡されなかった変数や配列などを使用するたびに、ローカルメモリにアクセスします。明示的に使用する必要はありません。実際、レジスタと共有メモリの方がはるかに高速であるため、使用を最小限に抑えるようにしてください。

編集:Re:共有メモリ。2つのブロックで共有メモリを交換したり、互いの共有メモリを確認したりすることはできません。ブロックの実行順序は保証されていないため、これを実行しようとすると、別のブロックが実行されるのを待つためにSMPを何時間も拘束する可能性があります。同様に、デバイス上で同時に実行されている2つのカーネルは、グローバルメモリでない限り、互いのメモリを認識できません。その場合でも、(競合状態の)火で遊んでいます。私の知る限り、ブロック/カーネルは実際には相互に「メッセージ」を送信することはできません。ブロックの実行順序は毎回異なるため、シナリオは実際には意味がありません。ブロックを停止して別のブロックを待機することはお勧めできません。

于 2012-08-02T20:30:40.313 に答える