essentially it seems that if I do __host__ __device__ then it passes the pointer to __host__ function while it is invoked within __device__ function. The runtime exception thrown is as follows: terminate called after throwing an instance of 'std::runtime_error' what(): cudaFuncGetAttributes(&attr_tmp, base_t::get_kernel_func()) error( cudaErrorIllegalAddress): an illegal memory access was encountered /home/nk/Install/kokkos-debug/include/Cuda/Kokkos_Cuda_KernelLaunch.hpp:654 Traceback functionality not available