CUDA应用程序的自定义类int2_， float2_和double2_之间的重载操作符=

Overloading operator= between customized classes int2_, float2_ and double2_ for CUDA application

本文关键字：double2 之间 操作符 重载 自定义 应用程序 int2 CUDA float2      更新时间：2023-10-16

我定义了int2_, float2_和double2_类来处理c++和CUDA中的复杂算术。我想重载操作符=，用于上述类和int、float和double类型的对象的混合赋值。

我的实现如下:

class float2_;
class double2_;
class int2_ {
    public:
        int x;
        int y;
        int2_() : x(), y() {}
        __host__ __device__ inline const int2_& operator=(const int a)          { x = a;            y = 0.;         return *this; }
        __host__ __device__ inline const int2_& operator=(const float a)        { x = (int)a;       y = 0.;         return *this; }
        __host__ __device__ inline const int2_& operator=(const double a)       { x = (int)a;       y = 0.;         return *this; }
        __host__ __device__ inline const int2_& operator=(const int2_ a)        { x = a.x;          y = a.y;        return *this; }
        __host__ __device__ inline const int2_& operator=(const float2_ a);
        __host__ __device__ inline const int2_& operator=(const double2_ a);
};
class float2_ {
    public:
        float x;
        float y;
        float2_() : x(), y() {}
        __host__ __device__ inline const float2_& operator=(const int a)        { x = (float)a;     y = 0.;         return *this; }
        __host__ __device__ inline const float2_& operator=(const float a)      { x = a;            y = 0.;         return *this; }
        __host__ __device__ inline const float2_& operator=(const double a)     { x = (float)a;     y = 0.;         return *this; }
        __host__ __device__ inline const float2_& operator=(const int2_ a)      { x = (float)a.x;   y = (float)a.y; return *this; }
        __host__ __device__ inline const float2_& operator=(const float2_ a)    { x = a.x;          y = a.y;        return *this; }
        __host__ __device__ inline const float2_& operator=(const double2_ a);
};
class double2_ {
    public:
        double x;
        double y;
        double2_() : x(), y() {}
        __host__ __device__ inline const double2_& operator=(const int a)       { x = (double)a;    y = 0.;         return *this; }
        __host__ __device__ inline const double2_& operator=(const float a)     { x = (double)a;    y = 0.;         return *this; }
        __host__ __device__ inline const double2_& operator=(const double a)    { x = a;            y = 0.;         return *this; }
        __host__ __device__ inline const double2_& operator=(const int2_ a)     { x = (double)a.x;  y = (double)a.y;return *this; }
        __host__ __device__ inline const double2_& operator=(const float2_ a)   { x = (double)a.x;  y = (double)a.y;return *this; }
        __host__ __device__ inline const double2_& operator=(const double2_ a)  { x = a.x;          y = a.y;        return *this; }
};
__host__ __device__ inline const int2_& int2_::operator=(const float2_ a)           { x = (int)a.x;             y = (int)a.y;       return *this; }
__host__ __device__ inline const int2_& int2_::operator=(const double2_ a)      { x = (int)a.x;             y = (int)a.y;       return *this; }
__host__ __device__ inline const float2_& float2_::operator=(const double2_ a)  { x = (float)a.x;           y = (float)a.y;     return *this; }

但是，我在内核中收到一个编译错误

template <class A, class T1, class T2>
__global__ inline void evaluation_matrix(T1 *data_, const Expr<A,T2> e, int NumElements)
{
    const int i = blockDim.x * blockIdx.x + threadIdx.x;
    if(i < NumElements) data_[i] = e[i];
}

当e是表达式时。错误信息是

calling a __host__ function("float2_::float2_") from a __global__  
function("evaluation_matrix<BinExpr<const float *, const float2_ *, CudaOpSum, float2_> 
, double2_, float2_> ") is not allowed

在本例中，data_是double2_对象，e是float2_表达式。

我在处理int, float, double, int2_, float2_或double2_类型或类别的data_方面没有任何问题。当e是int, float或double类型的表达式时，我甚至没有收到错误消息。只有当e属于int2_、float2_或double2_类时，才会出现问题。

帮忙吗?谢谢你。

class float2_;
class double2_;
class int2_ {
    public:
        int x;
        int y;
        __host__ __device__ int2_() : x(), y() {}
        __host__ __device__ inline const int2_& operator=(const int a)          { x = a;            y = 0.;         return *this; }
        __host__ __device__ inline const int2_& operator=(const float a)        { x = (int)a;       y = 0.;         return *this; }
        __host__ __device__ inline const int2_& operator=(const double a)       { x = (int)a;       y = 0.;         return *this; }
        __host__ __device__ inline const int2_& operator=(const int2_ a)        { x = a.x;          y = a.y;        return *this; }
        __host__ __device__ inline const int2_& operator=(const float2_ a);
        __host__ __device__ inline const int2_& operator=(const double2_ a);
};
class float2_ {
    public:
        float x;
        float y;
        __host__ __device__ float2_() : x(), y() {}
        __host__ __device__ inline const float2_& operator=(const int a)        { x = (float)a;     y = 0.;         return *this; }
        __host__ __device__ inline const float2_& operator=(const float a)      { x = a;            y = 0.;         return *this; }
        __host__ __device__ inline const float2_& operator=(const double a)     { x = (float)a;     y = 0.;         return *this; }
        __host__ __device__ inline const float2_& operator=(const int2_ a)      { x = (float)a.x;   y = (float)a.y; return *this; }
        __host__ __device__ inline const float2_& operator=(const float2_ a)    { x = a.x;          y = a.y;        return *this; }
        __host__ __device__ inline const float2_& operator=(const double2_ a);
};
class double2_ {
    public:
        double x;
        double y;
        __host__ __device__ double2_() : x(), y() {}
        __host__ __device__ inline const double2_& operator=(const int a)       { x = (double)a;    y = 0.;         return *this; }
        __host__ __device__ inline const double2_& operator=(const float a)     { x = (double)a;    y = 0.;         return *this; }
        __host__ __device__ inline const double2_& operator=(const double a)    { x = a;            y = 0.;         return *this; }
        __host__ __device__ inline const double2_& operator=(const int2_ a)     { x = (double)a.x;  y = (double)a.y;return *this; }
        __host__ __device__ inline const double2_& operator=(const float2_ a)   { x = (double)a.x;  y = (double)a.y;return *this; }
        __host__ __device__ inline const double2_& operator=(const double2_ a)  { x = a.x;          y = a.y;        return *this; }
};
__host__ __device__ inline const int2_& int2_::operator=(const float2_ a)           { x = (int)a.x;             y = (int)a.y;       return *this; }
__host__ __device__ inline const int2_& int2_::operator=(const double2_ a)      { x = (int)a.x;             y = (int)a.y;       return *this; }
__host__ __device__ inline const float2_& float2_::operator=(const double2_ a)  { x = (float)a.x;           y = (float)a.y;     return *this; }

这个错误说您不能从__global__函数调用__host__函数(在本例中是float2_类的构造函数)。乍一看，这与操作符无关，因为它们没有在错误消息中提到。但如果你仔细看，有data_[i] = e[i]。

注意:这里纯属猜测，因为您没有显示所有相关代码:

我猜e[i]给出了对表达式的一部分的引用，在这种情况下，类型为float2_。您将e[i]分配给double2_，相应的赋值运算符是double2_::__host__ __device__ inline const double2_& operator=(const float2_ a)，除了不必要和非常规地返回const引用外，还通过值获取float2_ ，因此编译器必须通过复制构造函数复制e[i]，该构造函数似乎是用__host__声明的。显然，从编译器消息来看，不允许从__global__函数调用__host__。

一种解决方案是声明构造函数__global__，或者让op=通过(const)引用接受其形参，这样就不需要调用复制构造函数了。但是，由于operator=本身被声明为__host__，因此您可能也会由于该调用而得到相同的错误。

我不知道cuda和不知道更多关于__host__和__global__比错误信息告诉我，但希望我能给你一个提示什么可能是错误的代码。

相关文章：