__host__ __device__ -- Generic programming in Cuda

TL;DR

This paper introduces programming patterns for Cuda/C++ that enable writing safe, generic code compatible with both host and device, addressing common compilation issues caused by dual instantiation of __host__ and __device__ functions.

Contribution

It presents novel patterns for writing templated Cuda/C++ functions that work seamlessly on CPU and GPU without compiler errors.

Findings

Patterns reduce compiler warnings/errors for dual __host__ and __device__ functions.

Enables safer, more portable generic programming in Cuda/C++.

Improves developer productivity and code maintainability.

Abstract

We present patterns for Cuda/C++ to write save generic code which works both on the host and device side. Writing templated functions in Cuda/C++ both for the CPU and the GPU bears the problem that in general both __host__ and __device__ functions are instantiated, which leads to lots of compiler warnings or errors.