c++ - CUDA: 構造体の共有データ メンバーとその構造体への参照のメンバーが異なるアドレス、値を持つ

Question

OK、問題は次のとおりです。CUDA 1.1 計算 GPU を使用して、スレッドごとに一連の (おそらくさまざまな数、ここでは 4 に固定されている) インデックスを維持しようとしています。私の問題は、メンバー配列にアクセスするときに構造体への参照を取得すると、結果が正しくないことです。メンバー配列の値を 0 で初期化し、元の構造体 var を使用して配列 vals を読み取ると、正しい値 (0) が得られます。 、しかし、構造体 var への参照を使用して読み取ると、ガベージ (-8193) が発生します。これは、構造体の代わりにクラスを使用している場合でも発生します。

なぜ tmp below != 0 ??

c++ は私の主要な言語ではないため、これは概念的な問題であるか、cuda での作業の癖である可能性があります。

struct DataIdx {
    int numFeats;
    int* featIdx;
};
extern __shared__ int sharedData[];

__global__  void myFn(){
    int tidx = blockIdx.x * blockDim.x + threadIdx.x;

    DataIdx myIdx;  //instantiate the struct var in the context of the current thread
    myIdx.numFeats = 4;
    size_t idxArraySize = sizeof(int)*4;
    //get a reference to my array for this thread. Parallel Nsight debugger shows myIdx.featIdx address = 0x0000000000000000e0
    myIdx.featIdx = (int*)(&sharedData[tidx*idxArraySize]);  

    myIdx.featIdx[0] = 0x0;  //set first value to 0 
    int tmp = myIdx.featIdx[0];  // tmp is correctly eq to 0 in Nsight debugger -- As Expected!!
    tmp = 2*tmp;    antIdx.featIdx[0] = tmp; //ensure compiler doesn't elide out tmp

    DataIdx *tmpIdx = &myIdx;  //create a reference to my struct var
    tmp = tmpIdx.featIdx[0];   // expected 0, but tmp = -8193 in debugger !! why?  debugger shows address of tmpIdx.featIdx = __devicea__ address=8
    tmpIdx.featIdx[0] = 0x0;
    tmp = tmpIdx.featIdx[0]; // tmp = -1; cant even read what we just set

    //forcing the same reference as myIdx.featIdx, still gives a problem! debugger shows address of tmpIdx.featIdx = __devicea__ address=8
    tmpIdx->featIdx =  (int*)(&sharedData[tidx*idxArraySize]); 
    tmp = tmpIdx.featIdx[0]; //tmp = -8193!! why != 0?

    DataIdx tmpIdxAlias = myIdx;
    tmp = tmpIdx.featIdx[0]; //aliasing the original var gives correct results, tmp=0


     myIdx.featIdx[0] = 0x0;
     mySubfn(&myIdx); //this is a problem because it happens when passing the struct by reference to subfns
     mySubfn2(myIdx);
}
__device__ mySubfn(struct DataIdx *myIdx){
  int tmp = myIdx->featIdx[0]; //tmp == -8193!! should be 0
}
__device__ mySubfn2(struct DataIdx &myIdx){
  int tmp = myIdx.featIdx[0]; //tmp == -8193!! should be 0
}