CUDA: Llamar a una función __device__ desde un kernel

Question

Tengo un kernel que llama a dispositivo dentro de una sentencia if. El código es el siguiente:

__device__ void SetValues(int *ptr,int id) 
{ 
    if(ptr[threadIdx.x]==id) //question related to here 
      ptr[threadIdx.x]++; 
} 

__global__ void Kernel(int *ptr) 
{ 
    if(threadIdx.x<2) 
     SetValues(ptr,threadIdx.x); 
} 

En los hilos del núcleo 0-1 setValues ​​de llamadas al mismo tiempo. ¿Qué pasa después de eso? Quiero decir que ahora hay 2 llamadas concurrentes a SetValues. ¿Cada llamada de función se ejecuta en serie? ¿Entonces se comportan como 2 llamadas a la función kernel?

Answer 1

CUDA realmente incluye todas las funciones por defecto (aunque Fermi y las arquitecturas más recientes también admiten un ABI apropiado con punteros a funciones y llamadas a funciones reales). Por lo que su código de ejemplo se compila a algo como esto

__global__ void Kernel(int *ptr) 
{ 
    if(threadIdx.x<2) 
     if(ptr[threadIdx.x]==threadIdx.x) 
      ptr[threadIdx.x]++; 
} 

ejecución se produce en paralelo, al igual que el código normal. Si diseñas una carrera de memoria en una función, no hay un mecanismo de serialización que pueda salvarte.