ฟังก์ชัน __device__ และ __host__

Kernels are entry points, but most code lives in helpers. CUDA gives two more qualifiers to say where a plain function may run. Let us meet them.

A __device__ function runs on the GPU and is callable only from other GPU code, like a kernel or another device function.

Think of __device__ functions as your GPU toolbox. They keep kernels tidy by holding logic the device threads reuse.

The __host__ qualifier means ordinary CPU code. It is the default, so plain C++ functions are already host functions.

Leave off all qualifiers and your function is host-only by default. That is why normal C++ still compiles inside a .cu file.

Combine the two and a function compiles for both sides. nvcc builds one version for the CPU and one for the GPU.

Marking a helper __host__ __device__ lets the same math run in CPU reference code and inside kernels, with zero duplication. ✨

Host code may not call a __device__ function, and device code may not call a host-only one. The compiler enforces the boundary.

A __device__ function is not launchable. Only __global__ kernels accept the triple-angle-bracket launch from the host.

The compiler often inlines small __device__ functions, so wrapping logic in helpers usually costs nothing at runtime.

Choose by where it runs: __device__ for GPU helpers, __host__ for CPU code, and both when the logic is shared.

Time to test the device and host qualifiers.

You learned __device__ runs on the GPU, __host__ runs on the CPU, the default is host, and combining both shares one function across worlds. 🎯