Asignación de memoria dentro de un kernel CUDA

Question

Tengo el siguiente (fragmento) de un kernel.

__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); 

} 

Cuando es (incluso los mismos tamaños) estáticas es muy rápido, pero cuando CurrentProbs se asigna dinámicamente (como antes) el rendimiento es horrible.

Esta pregunta me dijeron que podía hacer esto dentro de un núcleo: CUDA allocate memory in __device__ function

Aquí es una pregunta relacionada: Efficiency of Malloc function in CUDA

Me preguntaba si otros métodos han resuelto este no sea el que se propone en el documento? Parece ridículo que uno no pueda malloc/free dentro de un kernel sin este tipo de penalización.

Answer 1

Creo que el motivo por el que se introduce malloc() ralentiza el código es porque asigna memoria a la memoria global. Cuando utiliza una matriz de tamaño fijo, es probable que el compilador la coloque en el archivo de registro, que es mucho más rápido.

Tener que hacer un malloc dentro de su kernel puede significar que está tratando de hacer demasiado trabajo con un solo kernel. Si cada subproceso asigna una cantidad diferente de memoria, cada subproceso se ejecuta un número diferente de veces en el ciclo for, y obtienes mucha divergencia de distorsión.

Si cada hilo en una urdimbre ejecuta repeticiones la misma cantidad de veces, simplemente asigne la parte delantera. Incluso si ejecutan un número diferente de veces, puede usar un tamaño constante. Pero, en cambio, creo que debería ver cómo puede refactorizar su código para eliminar por completo ese bucle de su kernel.