c++ - 为什么以下 pop 函数不能在主机或设备 (cuda) 上运行？

Question

为什么代码的函数pop()在cuda编程中不能用global kernel或host thread运行，而其他函数没问题？调试了半天，也没发现什么问题。谁能告诉我为什么？

#define N 10
//define a vector for new or delete dynamically
template<typename T>
class LocalVector
{
private:
    T* m_begin;
    T* m_end;

    size_t capacity;
    size_t length;
    __host__ __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:
    __host__ __device__  explicit LocalVector() : length(0), capacity(15) {
        m_begin = new T[capacity];
        m_end = m_begin;
    }
    __host__ __device__ T& operator[] (unsigned int index) {
        return *(m_begin + index);
    }
    __host__ __device__ T* begin() {
        return m_begin;
    }
    __host__ __device__ T* end() {
        return m_end;
    }
    __host__ __device__ ~LocalVector()
    {
        delete[] m_begin;
        m_begin = nullptr;
    }
    __host__ __device__ void add(T t) {
        if ((m_end - m_begin) >= capacity) {
            expand();
        }

        new (m_end) T(t);
        m_end++;
        length++;
    }
    __host__ __device__ T pop() {
        T endElement = (*m_end);
        delete m_end;
        m_end--;
        return endElement;
    }
    __host__ __device__ size_t getSize() {
        return length;
    }
};
//cuda kernel function
__global__ void addKernel(LocalVector<int> *c)
{
    int i = threadIdx.x;
    c->pop();
}
//main function
int main() {
    cudaSetDevice(0);
    LocalVector<int> hlv;
    for (int i = 0; i < N; i++) {
        hlv.add(i*2);
    }
    hlv.pop();//can't work，and stop running
    printf("\n");
    for (int i = 0; i < hlv.getSize(); i++) {
        printf("hlv[%d]=%d,", i,hlv[i]);
    }

    return 0;
}

Answer 1

当你pop时，你不应该delete任何东西，只有当expand或delete时你才delete 析构时的整个内部数组。此更改将修复:

void add(T t) {
    if ((m_end - m_begin) >= capacity) {
        expand();
    }
    *(m_begin + length) = t;
    m_end++;
    length++;
}
 T pop() {
    T endElement = (*m_end);
    m_end--;
    return endElement;
}
 long long  getSize() {
    return length;
}

住在 Godbolt