Execution space qualifiers




__device__ is an execution space qualifier indicating that the corresponding function is to be called from a device and executed on a device. Functions with the __device__ execution space qualifier give access to gridDim, blockDim, blockIdx and threadIdx. Other execution space qualifiers are __global__ and __host__.




#include <stdio.h>
#include <stdlib.h>

 * @brief This function sets the int passed to 12345.
 * @details This function is only callable from a device because of the
 * __device__ function specifier.
__device__ void initialiseA(int* a)
    *a = 12345;

 * @brief This function does the interface between the host and the __device__
 * function because its __global__ function specifier means it will be called
 * from the host, and since this function is on the device, it can call
 * __device__ functions such as initialiseA.
__global__ void interface(int* a)

 * @brief Illustrates how to use a __device__ function.
 * @details This application consists in initialising a variable via a
 * __device__ function. To call such a function, a __global__ function serves
 * as the man-in-the-middle; both callable from host and able to call __device__
 * functions.
int main(int argc, char* argv[])
    // Declare variable on host.
    int a_host = 0;

    // Declare pointer that will point to the memory allocated on the device.
    int* a_device;

    // Allocate memory on the device
    cudaMalloc(&a_device, sizeof(int));

    // Launch the kernel on the device
    printf("Before launching the kernel, a = %d.\n", a_host);
    interface<<<1, 1>>>(a_device);

    // Copy the variable initialised back from the device to the host and print its value
    cudaMemcpy(&a_host, a_device, sizeof(int), cudaMemcpyDeviceToHost);
    printf("After launching the kernel, a = %d.\n", a_host);

    // Free resources

    return EXIT_SUCCESS;