Usando std :: vector en el código del dispositivo CUDA

Question

La pregunta es: ¿hay alguna manera de usar el "vector" de clase en los kernels Cuda? Cuando lo intento, aparece el siguiente error:

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

¿Hay una forma de usar un vector en la sección global? Recientemente he intentado lo siguiente:

1. crear un nuevo proyecto Cuda
2. ir a las características del proyecto
3. abierta Cuda C/C++
4. ir a dispositivo
5. cambiar el valor en "Código Generación "para establecerse en este valor: compute_20, sm_20

........ después de eso, pude usar el printf función de biblioteca estándar en mi kernel Cuda.

¿hay alguna manera de utilizar la clase de biblioteca estándar vector en la forma en que printf es compatible con el código kernel? Este es un ejemplo del uso de printf en el código del kernel:

// 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]); 
     } 
    } 
} 

Answer 1

No se puede utilizar la STL en CUDA, pero es posible que pueda utilizar el Thrust library a hacer lo que quiera. De lo contrario, solo copie el contenido del vector en el dispositivo y opere normalmente.

Answer 2

no se puede usar std::vector en el código del dispositivo, en su lugar debe usar una matriz.

Answer 3

En el empuje de la biblioteca cuda, puede usar thrust::device_vector<classT> para definir un vector en el dispositivo, y la transferencia de datos entre el vector STL del host y el vector del dispositivo es muy sencillo. Puede consultar este útil enlace: http://docs.nvidia.com/cuda/thrust/index.html para encontrar algunos ejemplos útiles.

Answer 4

Creo que puede implementar un vector de dispositivo por usted mismo, porque CUDA admite la asignación de memoria dinámica en los códigos del dispositivo. Operador nuevo/eliminar también son compatibles. Aquí hay un prototipo de vector de dispositivo extremadamente simple en CUDA, pero funciona. No ha sido probado lo suficiente.

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; 
    } 
};