cuda-gdb doesn't demangle variable names in device code

[Saitama10000]

Cuda-gdb doesn't demangle variable names in device code. It only does this in host code. This is troublesome for writing custom pretty printers.

#include <iostream>
#include "cooperative_groups.h"
namespace cg = cooperative_groups;

template <typename _value_type, int _size>
class Array
{
public:
    using value_type = _value_type;
    __device__ __host__ constexpr Array(std::initializer_list<_value_type> l) { int i = 0; for (auto x : l) (*this)[i++] = x; }
    __device__ __host__ constexpr auto operator[](int i) const { return data[i]; }
    __device__ __host__ constexpr auto operator[](int i) -> auto& { return data[i]; }
    __device__ __host__ constexpr auto size() const { return _size; }
private:
    value_type data[_size]{};
};

template <typename T, int size>
__device__ __host__ void iota(Array<T, size>& arr, T start)
{
    for (int i = 0; i < arr.size(); i++)
        arr[i] = start++;
};

__global__ void kernel()
{
    Array<int, 8> arr{};
    iota(arr, 0);
}

int main()
{
    Array<int, 8> arr{};
    iota(arr, 0);

    kernel<<<1, 8>>>();
    cudaDeviceSynchronize();
}

If I remove @local and just do demangle _Z5ArrayIiLi8EE I get Array<int, 8> so cuda-gdb knows how to demangle, it just doesn't want to do it for some reason.

I don't know whether this is a bug or a known limitation or whether I'm doing something wrong on my end. I'm using cuda toolkit 12.0.1

[sanannavyaa]

Hi! Thanks for the feedback! I will make sure to report this!

For your future reference, cuda-gdb bugs are best reported on their github or on the NVIDIA developer forum

[Saitama10000]

Sorry about that, I didn’t realize this was not an nsight-vscode-extension issue. I somehow correlated the extension to cuda-gdb itself. My bad!

[sanannavyaa]

No problem at all! In many ways they are related and no problem if you want to report issues on here. It's just that you are likely to get a useful response on their forums faster and it's easier for tracking purposes. That said, really not a big deal!