CUDA多设备问题，推力:: system_error

Question

我开发了一个算法，使用thrust。我的办公室的电脑有一个支持CUDA卡架构：

--- General information about Device 0 Name: Quadro 2000 Compute Capability: 2.1 Clock Rate: 1251000 kHz Device Overlap: Enabled Kernel Execution Timeout: Disabled

在这个机器上，我的算法没有任何错误运行。然而，当试图生成device_vector时，在实验室机器上的干净构建引发讨厌的thrust::system::system_error。两台机器都运行RedHat 6并配置相同，但多个图形卡除外。这个实验机包含三个具有以下架构支持CUDA卡：

--- General information about Device 0 Name: Tesla C2050 Compute Capability: 2.0 Clock Rate: 1147000 kHz Device Overlap: Enabled Kernel Execution Timeout: Disabled

--- General information about Device 1 Name: Quadro 2000 Compute Capability: 2.1 Clock Rate: 1251000 kHz Device Overlap: Enabled Kernel Execution Timeout: Disabled

--- General information about Device 2 Name: Quadro 2000 Compute Capability: 2.1 Clock Rate: 1251000 kHz Device Overlap: Enabled Kernel Execution Timeout: Enabled`

我知道thrust必须以对工作目标架构编译。因此，我将CUDA设备设置为1。但是，错误仍然存​​在。

作为一项调试措施，我在device_vector分配之前立即调用了一个cudaGetDevice()调用。该设备正确说明为1。

int device; 
CUDA_CHECK_RETURN(cudaGetDevice(&device), __FILE__, __LINE__); 
std::cout << "Operating on device " << device << std::endl; // <-- device 1 

// copy the turns to the runtime 
thrust::device_vector<MalfunctionTurn> d_turns = turns; // <-- error here 

我在我的智慧结束试图调试这一点。有没有人看到过这样的错误？更值得注意的是，在我不知道的cudaSetDevice()中有没有限制？我很担心，因为不同机器上的两张相同的卡不能运行相同的代码。

在此先感谢。

---

编辑

编译命令行：nvcc -rdc=true -arch=sm_21 -O3 file

这里是再现误差小例子：由于靠近我可以告诉

#define DEVICE __device__ 
#define HOST __host__ 

#include <thrust/host_vector.h> 
#include <thrust/device_vector.h> 

template <typename T, std::size_t N> 
class Container { 
public: 

    DEVICE HOST 
    Container() { 

    } 

private: 
    T data[N]; 
}; 

typedef Container<double, 7> double7; 

template <std::size_t N = 10 > 
class History { 
public: 

    DEVICE HOST 
    History() { 

    } 

    DEVICE HOST 
    virtual ~History() { 

    } 

private: 

    double7 history[N]; 
}; 

int main() { 

    try { 

     thrust::host_vector<History<> > histories(1); 
     thrust::device_vector<History<> > d_histories = histories; 
    } catch (const thrust::system_error &) { 
     std::cerr << "boo boo" << std::endl; 
    } 

    return 0; 
} 

Answer 1

，你的代码是（引擎盖下）违反了对usage of classes with virtual functions的CUDA限制：

It is not allowed to pass as an argument to a __global__ function an object of a class with virtual functions.

如果我把你的代码并将其删除：

#include <JARSS.h> 

，取而代之的是：

#define HOST __host__ 
#define DEVICE __device__ 

我可以编译它。通过启动一个内核作为参数的主机上的对象被复制到设备

thrust::device_vector<History<> > d_histories = histories; 

：然而，在引擎盖下，推力执行这条线。 （例如，您可以使用nvprof验证内核启动。）这是推力的常见行为。问题在于这些具有虚拟析构函数的对象不能以这种方式复制。

与您的问题状态相反，此代码不应在任何CUDA计算机上正常运行。

你可以通过注释虚拟析构函数来“修复”你的代码。如果您需要多态行为，那么对实际代码的修复可能会比这更复杂。应用仍然可以使用带推力的多态对象行为，请参阅this answer以获得一个工作示例。