Can string data types be used in C++ CUDA kernels?

Question

I am writing a CUDA kernel in which I'm using the string data type in C++. However, the compiler is throwing the following error :

error: calling a host function("std::basic_string<char, std::char_traits<char>, std::allocator<char> >::operator =") from a __device__/__global__ function("doDecompression") is not allowed

Are strings not allowed within a kernel? if not, what is the workaround to allocate space for a char array within a kernel?

Answer 1

You cannot use the C++ string type in a kernel because CUDA does not include a device version of the C++ String library that would be able run on the GPU. Even if it was possible to use string in a kernel, it's not something you would want to do because string handles memory dynamically, which would be likely to be slow.

Instead, create an array of fixed length strings and copy the strings to it. Then copy the array to the GPU. Pass the base address of the array of strings to your kernel and have each thread calculate the address to a given string by adding an offset based on its index to the base address.