9

いくつか質問があります。

最近、CUDAを使ってプログラムを作っています。

私のプログラムでは、std :: map(string、vector(int))でプログラムされたホストに1つのビッグデータがあります。

これらのデータを使用することにより、一部のvector(int)がGPUグローバルメモリにコピーされ、GPUで処理されます

処理後、いくつかの結果がGPUで生成され、これらの結果がCPUにコピーされます。

これらはすべて私のプログラムスケジュールです。

  1. cudaMemcpy(...、cudaMemcpyHostToDevice)
  2. カーネル関数(カーネル関数は、必要なデータがGPUグローバルメモリにコピーされている場合にのみ実行できます)
  3. cudaMemcpy(...、cudaMemcpyDeviceToHost)
  4. 1〜3ステップを1000回繰り返す(別のデータ(ベクトル)の場合)

でも処理時間を短縮したいです。

そこで、プログラムでcudaMemcpyAsync関数を使用することにしました。

いくつかのドキュメントやWebページを検索した後、GPUにコピーするデータを持つcudaMemcpyAsync関数のホストメモリを使用するには、グローバルメモリを固定メモリとして割り当てる必要があることに気付きました。

しかし、私のプログラムはstd :: mapを使用しているため、このstd::mapデータを固定メモリに作成できませんでした。

したがって、これを使用する代わりに、バッファ配列型の固定メモリを作成しました。このバッファは、ベクトルをコピーするすべてのケースを常に処理できます。

最後に、私のプログラムはこのように機能しました。

  1. Memcpy(データ全体がバッファにコピーされるまで、ループを使用してstd :: mapからバッファにデータをコピーします)
  2. cudaMemcpyAsync(...、cudaMemcpyHostToDevice)
  3. カーネル(カーネル関数は、データ全体がGPUグローバルメモリにコピーされた場合にのみ実行できます)
  4. cudaMemcpyAsync(...、cudaMemcpyDeviceToHost)
  5. 1〜4ステップを1000回繰り返す(別のデータ(ベクトル)の場合)

そして、私のプログラムは前のケースよりもはるかに速くなりました。

しかし、問題(私の好奇心)はこの時点にあります。

同様の方法で別のプログラムを作成しようとしました。

  1. Memcpy(std :: mapから1つのベクトルのバッファーにのみデータをコピーします)
  2. cudaMemcpyAsync(...、cudaMemcpyHostToDevice)
  3. データ全体がGPUグローバルメモリにコピーされるまでループ1〜2
  4. カーネル(カーネル関数は、必要なデータがGPUグローバルメモリにコピーされている場合にのみ実行できます)
  5. cudaMemcpyAsync(...、cudaMemcpyDeviceToHost)
  6. 1〜5ステップを1000回繰り返す(別のデータ(ベクトル)の場合)

この方法は、上記の方法よりも約10%高速であることがわかりました。

でも理由はわかりません。

cudaMemcpyAsyncはカーネル関数とのみオーバーラップできると思います。

しかし、私の場合はそうではないと思います。cudaMemcpyAsync関数間でオーバーラップできるように見えるのではなく。

長い質問で申し訳ありませんが、その理由を本当に知りたいです。

誰かが私に正確な機能「cudaMemcpyAsync」とは何か、そしてどの機能を「cudaMemcpyAsync」とオーバーラップできるかを教えたり説明したりできますか?

4

1 に答える 1

18

cudaMemcpyAsync のコピー アクティビティ (およびカーネル アクティビティ) は、任意のホスト コードとオーバーラップできます。さらに、デバイスとの間のデータ コピー (cudaMemcpyAsync 経由) は、カーネル アクティビティとオーバーラップする可能性があります。ホスト アクティビティ、データ コピー アクティビティ、およびカーネル アクティビティの 3 つのアクティビティはすべて、互いに非同期で実行でき、互いにオーバーラップすることもできます。

これまで見てきたように、ホスト アクティビティとデータ コピーまたはカーネル アクティビティは、比較的単純な方法で互いにオーバーラップできます。カーネルの起動は、cudaMemcpyAsync と同様に、ホストにすぐに戻ります。ただし、データ コピーとカーネル アクティビティの間で最適なオーバーラップの機会を得るには、いくつかの追加の概念を使用する必要があります。最適なオーバーラップの機会を得るには、次のものが必要です。

  1. cudaHostAlloc() などを介して固定されたホスト メモリ バッファ
  2. さまざまなタイプのアクティビティ (データ コピーとカーネル計算) を分離するための cuda ストリームの使用
  3. cudaMemcpyAsync の使用 (cudaMemcpy の代わりに)

当然のことながら、作業も分離可能な方法で分割する必要があります。これは通常、カーネルが特定の機能を実行している場合、このカーネルを複数回呼び出す必要がある可能性があることを意味します。これにより、各呼び出しが個別のデータで動作できるようになります。これにより、たとえば、最初のカーネル呼び出しがデータ ブロック A で動作している間に、データ ブロック B をデバイスにコピーできます。そうすることで、データ ブロック B のコピーをデータ ブロック A のカーネル処理とオーバーラップさせることができます。

(cudaMemcpy と比較して) cudaMemcpyAsync との主な違いは次のとおりです。

  1. 任意のストリームで発行できます (ストリーム パラメータを取ります)。
  2. 通常は、データのコピーが完了するのを待つのではなく、(カーネル呼び出しのように)すぐにホストに制御を返します。

項目 1 は、データ コピーをカーネル計算とオーバーラップできるようにするために必要な機能です。項目 2 は、データ コピーをホスト アクティビティとオーバーラップできるようにするために必要な機能です。

コピー/計算オーバーラップの概念は非常に単純ですが、実際には実装にはいくつかの作業が必要です。その他の参考資料については、次を参照してください。

  1. CUDA ベスト プラクティス ガイドの重複コピー/計算セクション。
  2. copy/compute overloadの基本的な実装を示すサンプル コード。
  3. 完全なマルチ/コンカレント カーネル コピー/コンピューティング オーバーラップ シナリオを示すサンプル コード。

上記の説明の一部は、コンピューティング機能 2.0 以上のデバイス (同時実行カーネルなど) を前提としていることに注意してください。また、異なるデバイスには 1 つまたは 2 つのコピー エンジンがある場合があります。つまり、デバイスへのコピーとデバイスからのコピーを同時に実行できるのは、特定のデバイスでのみです。

于 2012-12-06T15:00:54.983 に答える