1

ねえ、私はCUDAとThrustライブラリを使用しています。ホストからObject*型のthrust::device_vector（ポインターのベクトル）がロードされたCUDAカーネル上のダブルポインターにアクセスしようとすると、問題が発生します。「nvcc-othrustmain.cpp cukernel.cu」を使用してコンパイルすると、「警告：グローバルメモリスペースを想定しているため、ポインタが何を指しているのかわかりません」という警告と、プログラムを実行しようとすると起動エラーが発生します。

Nvidiaフォーラムを読みましたが、解決策は「CUDAカーネルでダブルポインターを使用しないでください」のようです。カーネルに送信する前に、ダブルポインタを1Dポインタに折りたたむつもりはありません...この問題の解決策を見つけた人はいますか？必要なコードは以下のとおりです。よろしくお願いします。

--------------------------
        main.cpp
--------------------------

Sphere * parseSphere(int i)
{
  Sphere * s = new Sphere();
  s->a = 1+i;
  s->b = 2+i;
  s->c = 3+i;
  return s;
}

int main( int argc, char** argv ) {

  int i;
  thrust::host_vector<Sphere *> spheres_h;
  thrust::host_vector<Sphere> spheres_resh(NUM_OBJECTS);

  //initialize spheres_h
  for(i=0;i<NUM_OBJECTS;i++){
    Sphere * sphere = parseSphere(i);
    spheres_h.push_back(sphere);
  }

  //initialize spheres_resh
  for(i=0;i<NUM_OBJECTS;i++){
    spheres_resh[i].a = 1;
    spheres_resh[i].b = 1;
    spheres_resh[i].c = 1;
  }

  thrust::device_vector<Sphere *> spheres_dv = spheres_h;
  thrust::device_vector<Sphere> spheres_resv = spheres_resh;
  Sphere ** spheres_d = thrust::raw_pointer_cast(&spheres_dv[0]);
  Sphere * spheres_res = thrust::raw_pointer_cast(&spheres_resv[0]);

  kernelBegin(spheres_d,spheres_res,NUM_OBJECTS);

  thrust::copy(spheres_dv.begin(),spheres_dv.end(),spheres_h.begin());
  thrust::copy(spheres_resv.begin(),spheres_resv.end(),spheres_resh.begin());

  bool result = true;

  for(i=0;i<NUM_OBJECTS;i++){
    result &= (spheres_resh[i].a == i+1);
    result &= (spheres_resh[i].b == i+2);
    result &= (spheres_resh[i].c == i+3);
  }

  if(result)
  {
    cout << "Data GOOD!" << endl;
  }else{
    cout << "Data BAD!" << endl;
  }

  return 0;
}


--------------------------
        cukernel.cu
--------------------------
__global__ void deviceBegin(Sphere ** spheres_d, Sphere * spheres_res, float    
num_objects)
{
  int index = threadIdx.x + blockIdx.x*blockDim.x;

  spheres_res[index].a = (*(spheres_d+index))->a; //causes warning/launch error
  spheres_res[index].b = (*(spheres_d+index))->b; 
  spheres_res[index].c = (*(spheres_d+index))->c; 
}

void kernelBegin(Sphere ** spheres_d, Sphere * spheres_res, float num_objects)
{

 int threads = 512;//per block
 int grids = ((num_objects)/threads)+1;//blocks per grid

 deviceBegin<<<grids,threads>>>(spheres_d, spheres_res, num_objects);
}
4

1 に答える 1

3

ここでの基本的な問題は、デバイス ベクトルspheres_dvにホスト ポインターが含まれていることです。Thrust は、GPU とホスト CPU のアドレス空間の間で「ディープ コピー」またはポインター変換を実行できません。そのため、GPU メモリにコピーspheres_hすると、ホスト ポインターの GPU 配列が作成されます。GPU 上のホスト ポインターの間接化は違法です。それらは間違ったメモリ アドレス空間のポインターであるため、カーネル内でセグメンテーション違反に相当する GPU を取得しています。

解決策として、現在ホスト メモリに新しい各構造体を割り当てているparseSphereを使用するのではなく、関数を GPU でメモリ割り当てを実行するものに置き換える必要があります。parseSphereFermi GPU を使用していて (実際には使用していないようです)、CUDA 3.2 または 4.0 を使用している場合、1 つの方法はparseSphereカーネルに変換することです。C++new演算子はデバイス コードでサポートされているため、構造体の作成はデバイス メモリで行われます。このアプローチが機能するにSphereは、コンストラクターが関数として定義されるように、 の定義を変更する必要があります。__device__

別のアプローチでは、デバイス ポインターを保持するホスト配列を作成し、その配列をデバイス メモリにコピーします。この回答でその例を見ることができます。thrust::device_vector包含を宣言してもthrust::device_vector機能しない可能性があることに注意してください。そのため、基になる CUDA API 呼び出しを使用して、このデバイス ポインターの配列の構築を行う必要がある可能性があります。

逆コピー操作についても触れていないことに注意してください。逆コピー操作も同様に困難です。

要するに、推力 (および C++ STL コンテナー) は実際にはポインターを保持することを意図していないということです。それらは、値を保持し、ユーザーが見ることを想定していない反復子と基礎となるアルゴリズムを使用して、ポインターの間接化と直接メモリ アクセスを抽象化することを目的としています。さらに、「ディープ コピー」の問題は、NVIDIA フォーラムの賢明な人々が GPU コードの複数レベルのポインターに対して助言する主な理由です。コードが非常に複雑になり、GPU での実行も遅くなります。

于 2011-06-06T06:01:18.740 に答える