Funções __device__ e __host__

Os kernels são pontos de entrada, mas a maior parte do código fica nas funções auxiliares. O CUDA oferece mais dois qualificadores para indicar onde uma função comum pode ser executada. Vamos conhecê-los.

Uma função __device__ é executada na GPU e só pode ser chamada por outro código da GPU, como um kernel ou outra função do dispositivo.

Pense nas funções __device__ como sua caixa de ferramentas da GPU. Elas mantêm os kernels organizados ao armazenar a lógica reutilizada pelas threads do dispositivo.

O qualificador __host__ indica código comum da CPU. Ele é o padrão, portanto funções comuns de C++ já são funções do hospedeiro.

Se você não usar qualificadores, sua função será exclusiva do hospedeiro por padrão. É por isso que C++ normal continua compilando dentro de um arquivo .cu.

Combine os dois e a função será compilada para ambos os lados. O nvcc cria uma versão para a CPU e outra para a GPU.

Marcar uma função auxiliar como __host__ __device__ permite executar a mesma matemática no código de referência da CPU e dentro dos kernels, sem duplicação. ✨

O código do hospedeiro não pode chamar uma função __device__, e o código do dispositivo não pode chamar uma função exclusiva do hospedeiro. O compilador impõe essa separação.

Uma função __device__ não pode ser lançada. Somente kernels __global__ aceitam o lançamento com três sinais de menor feito pelo hospedeiro.

O compilador frequentemente faz expansão inline de funções __device__ pequenas, portanto envolver a lógica em funções auxiliares geralmente não custa nada durante a execução.

Escolha de acordo com o local de execução: __device__ para funções auxiliares da GPU, __host__ para código da CPU e ambos quando a lógica for compartilhada.

Você aprendeu que __device__ é executado na GPU, __host__ é executado na CPU, o padrão é o hospedeiro e combinar os dois compartilha uma função entre os mundos. 🎯