0

私はスラスト(cuda)が初めてで、いくつかの配列操作を行いたいのですが、インターネット上で同様の例が見つかりません。

次の2つの配列(2d)があります:

a = { {1, 2, 3}, {4} }
b = { {5}, {6, 7} }

私は推力がこの配列を計算したい:

c = { {1, 2, 3, 5}, {1, 2, 3, 6, 7}, {1, 2, 3, 5}, {1, 2, 3, 6, 7} }

私はそれがc/c++でどのように機能するかを知っていますが、それを行うために推力を言う方法は知りません.

これがおそらくどのように機能するかという私の考えです：

スレッド 1: a[0] を取得 -> b で展開します。c に書き込みます。

スレッド 2: a[1] を取得 -> b で展開します。c に書き込みます。

しかし、私はそれを行う方法がわかりません。次のように、配列 a と b を 1 次元配列に書き込むことができます。

thrust::device_vector<int> dev_a;
dev_a.push_back(3); // size of first array
dev_a.push_back(1);
dev_a.push_back(2);
dev_a.push_back(3);
dev_a.push_back(1); // size of secound array
dev_a.push_back(4);

thrust::device_vector<int> dev_b;
dev_b.push_back(1); // size of first array
dev_b.push_back(5);
dev_b.push_back(2); // size of secound array
dev_b.push_back(6);
dev_b.push_back(7);

そして疑似関数：

struct expand
{
  __host__ __device__
  ?? ?? (const array ai, const array *b) {
      for bi in b: // each array in the 2d array
      {
          c.push_back(bi[0] + ai[0]); // write down the array count

          for i in ai: // each element in the ai array
             c.push_back(i);

          for i in bi: // each element in the bi array
             c.push_back(i);
      }
  }
};

誰でもアイデアはありますか？

4

1 に答える 1

1

このような操作では、GPU で速度が向上しないと思います。これは、GPU での低速な操作である大量のメモリ アクセスが必要になるためです。

しかし、とにかくこれを実装したい場合：

  1. 前に書いた理由から、すぐに使えるアルゴリズムでは信頼は役に立たないでしょう。これは、独自のカーネルを作成する必要があることを意味しますが、メモリ管理はそのままにしておくことができます。

  2. CPU メモリに配列を作成し、準備ができたら配列全体を GPU にコピーする方が常に高速です。(CPU<->GPU コピーは、データの長い連続部分で高速です)

  3. GPU は何百ものスレッドを並行して実行することに注意してください。各スレッドは、何を読み取り、どこに書き込むかを知る必要があります。

  4. グローバル メモリ操作が遅い (300 ～ 400 クロック)。スレッドがグローバル メモリから配列全体を読み取って、最後の数バイトだけが必要であることを確認することは避けてください。

だから、あなたがプログラムしているのを見ることができます。

  1. CPU メモリ内の配列を 1D にすると、次のようになります。

    float array1[] = { 1, 2, 3, 4}; float array2[] = { 5, 6, 7}; int arr1offsets[] = {0, 2, 3, 1}; // 最初の要素の位置と部分配列のペアの長さ int arr2offsets[] = {0, 1, 1, 2};

  2. 配列とオフセットを GPU にコピーし、結果とそのオフセットにメモリを割り当てます。1つのジョイントサブアレイの最大長を数え、最悪の場合に備えてメモリを割り当てる必要があると思います。

  3. カーネルを実行します。

  4. 結果を収集する

カーネルは次のようになります (私があなたの考えを正しく理解していれば)

__global__ void kernel(float* arr1, int* arr1offset, 
                       float* arr2, int* arr2offset, 
                       float* result, int* resultoffset)
{
  int idx = threadIdx.x+ blockDim.x*blockIdx.x;
  int a1beg = arr1offset[Idx*2];
  int a2beg = arr2offset[Idx*2];
  int a1len = arr1offset[Idx*2+1];
  int a2len = arr2offset[Idx*2+1];
  resultoffset[idx*2] = idx*MAX_SUBARRAY_LEN;
  resultoffset[idx*2+1] = a1len+a2len;

  for (int k = 0; k < a1len; ++k) result[idx*MAX_SUBARRAY_LEN+k] = arr1[a1beg+k];
  for (int k = 0; k < a2len; ++k) result[idx*MAX_SUBARRAY_LEN+a1len+k] = arr2[a2beg+k];
}

このコードは完全ではありませんが、正しく動作するはずです。

于 2013-01-07T22:19:07.103 に答える