在 CUDA 设备代码中使用 std::vector答案

【问题标题】：Using std::vector in CUDA device code在 CUDA 设备代码中使用 std::vector
【发布时间】：2015-07-12 11:49:25
【问题描述】：

问题是：有没有办法在 Cuda 内核中使用“向量”类？当我尝试时，我收到以下错误：

error : calling a host function("std::vector<int, std::allocator<int> > ::push_back") from a __device__/__global__ function not allowed

那么有办法在全局部分使用向量吗？我最近尝试了以下方法：

创建一个新的 Cuda 项目
转到项目的属性
打开 Cuda C/C++
转到设备
将“代码生成”中的值更改为以下值：计算_20,sm_20

........之后我就可以在我的 Cuda 内核中使用 printf 标准库函数了。

有没有办法以内核代码支持 printf 的方式使用标准库类vector？这是在内核代码中使用 printf 的示例：

// this code only to count the 3s in an array using Cuda
//private_count is an array to hold every thread's result separately 

__global__ void countKernel(int *a, int length, int* private_count) 
{
    printf("%d\n",threadIdx.x);  //it's print the thread id and it's working

    // vector<int> y;
    //y.push_back(0); is there a possibility to do this?

    unsigned int offset  = threadIdx.x * length;
    int i = offset;
    for( ; i < offset + length; i++)
    {
        if(a[i] == 3)
        {
            private_count[threadIdx.x]++;
            printf("%d ",a[i]);
        }
    }   
}

【问题讨论】：

+1 完全合法的问题（不知道为什么它被否决。不幸的是，目前的答案是否定的。
抱歉发布了 necroposting。只是想知道现在是否有任何答案

标签： cuda

【解决方案1】：

我认为你可以自己实现一个设备向量，因为 CUDA 支持设备代码中的动态内存分配。还支持运算符 new/delete。这是 CUDA 中一个非常简单的设备向量原型，但它确实有效。它还没有经过充分的测试。

template<typename T>
class LocalVector
{
private:
    T* m_begin;
    T* m_end;

    size_t capacity;
    size_t length;
    __device__ void expand() {
        capacity *= 2;
        size_t tempLength = (m_end - m_begin);
        T* tempBegin = new T[capacity];

        memcpy(tempBegin, m_begin, tempLength * sizeof(T));
        delete[] m_begin;
        m_begin = tempBegin;
        m_end = m_begin + tempLength;
        length = static_cast<size_t>(m_end - m_begin);
    }
public:
    __device__  explicit LocalVector() : length(0), capacity(16) {
        m_begin = new T[capacity];
        m_end = m_begin;
    }
    __device__ T& operator[] (unsigned int index) {
        return *(m_begin + index);//*(begin+index)
    }
    __device__ T* begin() {
        return m_begin;
    }
    __device__ T* end() {
        return m_end;
    }
    __device__ ~LocalVector()
    {
        delete[] m_begin;
        m_begin = nullptr;
    }

    __device__ void add(T t) {

        if ((m_end - m_begin) >= capacity) {
            expand();
        }

        new (m_end) T(t);
        m_end++;
        length++;
    }
    __device__ T pop() {
        T endElement = (*m_end);
        delete m_end;
        m_end--;
        return endElement;
    }

    __device__ size_t getSize() {
        return length;
    }
};

【讨论】：

即使是正确的，也有可能因为expand()函数内部的内存分配而变慢。虽然很努力。

【解决方案2】：

在cuda库推力中，可以使用thrust::device_vector<classT>在设备上定义一个向量，主机STL向量和设备向量之间的数据传输非常简单。你可以参考这个有用的链接：http://docs.nvidia.com/cuda/thrust/index.html 找到一些有用的例子。

【讨论】：

【解决方案3】：

你不能在设备代码中使用std::vector，你应该使用数组来代替。

【讨论】：

【解决方案4】：

您不能在 CUDA 中使用 STL，但您可以使用 Thrust library 来做您想做的事。否则只需将vector的内容复制到设备上，正常操作即可。

【讨论】：

我不明白这应该有什么帮助，因为 thrust::device_vector 也不能在内核中使用。