CUDAヒストグラム2dが動作しない

Question

CUDAの2次元ヒストグラムがまったくないように見えるので（私が見つけることができる...ポインタが歓迎）、私はpyCUDAでそれを自分で実装しようとしています。ここで

は、ヒストグラムが（numpyのを使用して）どのように見えるかです：

は、ここで私はこれまで持っているものです：

code = ''' 
__global__ void histogram2d(const float *in_x, const float *in_y, const float *in_w, float *out) {{ 
    int start = blockIdx.x * blockDim.x + threadIdx.x; 

    float *block_out = &out[{xres} * {yres} * {num_chans} * blockIdx.x]; 

    for(int i = 0; i < {length}; i++) {{ 
     float x = in_x[start + i]; 
     float y = in_y[start + i]; 
     int w_idx = (start + i) * {num_chans}; 

     int xbin = (int) (((x - {xmin})/{xptp}) * {xres}); 
     int ybin = (int) (((y - {ymin})/{yptp}) * {yres}); 

     if (0 <= xbin && xbin < {xres} && 0 <= ybin && ybin < {yres}) {{ 
      for(int c = 0; c < {num_chans}; c++) {{ 
       atomicAdd(&block_out[(ybin * {xres} + xbin) * {num_chans} + c], in_w[w_idx + c]); 
      }} 
     }} 
    }} 
}} 
'''.format(**args) 

------ 

__global__ void histogram2d(const float *in_x, const float *in_y, const float *in_w, float *out) { 
    int start = blockIdx.x * blockDim.x + threadIdx.x; 

    float *block_out = &out[50 * 50 * 4 * blockIdx.x]; 

    for(int i = 0; i < 100; i++) { 
     float x = in_x[start + i]; 
     float y = in_y[start + i]; 
     int w_idx = (start + i) * 4; 

     int xbin = (int) (((x - -10.0)/20.0) * 50); 
     int ybin = (int) (((y - -10.0)/20.0) * 50); 

     if (0 <= xbin && xbin < 50 && 0 <= ybin && ybin < 50) { 
      for(int c = 0; c < 4; c++) { 
       atomicAdd(&block_out[(ybin * 50 + xbin) * 4 + c], in_w[w_idx + c]); 
      } 
     } 
    } 
} 

があるようです索引付けの問題ですが、以前はあまり純粋なCUDAを行っていないので、それが何であるかは分かりません。

def slow_hist(in_x, in_y, in_w, out, blockx, blockdimx, threadx): 
    start = blockx * blockdimx + threadx 

    block_out_addr = args['xres'] * args['yres'], args['num_chans'] * blockx 

    for i in range(args['length']): 
     x = in_x[start + i] 
     y = in_y[start + i] 
     w_idx = (start + i) * args['num_chans'] 

     xbin = int(((x - args['xmin'])/args['xptp']) * args['xres']) 
     ybin = int(((y - args['ymin'])/args['yptp']) * args['yres']) 

     if 0 <= xbin < args['xres'] and 0 <= ybin < args['yres']: 
      for c in range(args['num_chans']): 
       out[(ybin * args['xres'] + xbin) * args['num_chans'] + c] += in_w[w_idx + c] 

すべてのコードは、これらの画像、at the Github page of this notebook（このセルは、一番下にある）を含む、閲覧可能である：ここでは、私は同等のpythonがあることだと思うものです。

このCUDAコードで何が間違っていますか？私はたくさんの微調整を試しました（1、4、8、16のアトミックアドレスを積み重ねて、出力をトランスポーズするなど）、ポインターの算術演算の仕組みには微妙なものがないようです。どんな助けもありがとう。

Answer 1

CUDAセクションの出力配列に割り当てられている配列は、float32の代わりにNumpyのデフォルトのfloat64を使用していたため、メモリは予想より2倍大きかった。ここでは、新たなヒストグラム出力です：

私はまだ非常にこれらのヒストグラムは、すべて互いにそれほど異なっている理由を説明するのに役立つコメントや答えをいただければと思います。