CUDAカーネル内のメモリ割り当て

Question

私は以下の（スニペットの）カーネルを持っています。

__global__ void plain(int* geneVec, float* probs, int* nComponents, float* randomNumbers,int *nGenes) 
{ 

    int xid = threadIdx.x + (blockDim.x * blockIdx.x); 

    float* currentProbs= (float*)malloc(sizeof(float)*tmp); 

     ..... 
     ..... 

    currentProbs[0] = probs[start]; 
    for (k=1;k<nComponents[0]; k++) 
    { 
     currentProbs[k] = currentProbs[k-1] + prob; 
    } 

     ... 
     ... 
     free(currentProbs); 

} 

それは静的です（でも同じサイズ）、それは非常に高速ですが、CurrentProbsを動的に（上記のように）割り当てられた場合、パフォーマンスがひどいです。

この質問は、私は、カーネル内部でこれを行うことができると言いました。ここでCUDA allocate memory in __device__ function

は、関連する質問です：Efficiency of Malloc function in CUDA

任意の他の方法は、論文で提案したものよりも、この他を解決した場合、私は不思議でしたか？ この種の罰金がないと、カーネル内でmalloc/freeを実行できないというのはばかげているようです。

Answer 1

malloc（）を導入する理由は、コードが減速する理由は、グローバルメモリにメモリを割り当てることだと思います。固定サイズの配列を使用すると、コンパイラはレジスタファイルに格納する可能性が高くなります。これははるかに高速です。

カーネル内でmallocを実行すると、1つのカーネルで多すぎる作業をしようとしている可能性があります。各スレッドが異なる量のメモリを割り当てると、各スレッドはforループで異なる回数実行され、ワー​​プの分岐が多く発生します。

ワープ内の各スレッドが同じ回数だけループを実行する場合は、前に配置します。それらが異なる回数実行されても、一定のサイズを使用できます。しかし、代わりに、カーネルからそのループを完全に削除するためにコードをリファクタリングする方法を見てください。