2

我正在尝试将一些 POD 传递给一个内核，该内核具有一些非 POD 作为参数，并且具有非显式构造函数。其背后的想法是：在主机上分配一些内存，将内存传递给内核，并将内存封装在对象中，而无需用户显式执行该步骤。

构造函数被标记为 __device__ 代码，但是在传递参数时没有调用它们，我不知道为什么。

我的问题与我应该如何做这件事并没有真正的关系，而是试图了解幕后发生的事情。

这是一个示例（我使用的是具有 2.1 能力的 GPU 的 CUDA 5，因此是 printf）。

#include <stdio.h>

struct Test {
    __device__ Test() {
        printf("Default\n"),
        _n = 0;
    }
    __device__ Test(int n) {
        printf("Construct %d\n", n);
        _n = n;
    }
    __device__ Test(const Test &t) {
        printf("Copy constr %d\n", t._n);
        _n = t._n;
    }
    __device__ Test &operator=(const Test &t) {
        printf("Assignment %d\n", t._n);
        _n = t._n;
        return *this;
    }
    __device__ int calc() const {
        printf("Calculating %d\n", threadIdx.x + 10 * _n);
        return threadIdx.x + 10 * _n;
    }
    int _n;
};

__global__ void dosome(Test a, Test b) {
    printf("Kernel data %d %d\n", a._n, b._n);
    a.calc();
    b.calc();
}

int main(int argc, char **argv) {
    dosome<<<1, 2>>>(2, 3);
    cudaError_t cudaerr = cudaDeviceSynchronize();
    if (cudaerr != cudaSuccess)
        printf("kernel launch failed with error:\n\t%s\n",cudaGetErrorString(cudaerr));
    return 0;
}

编辑：忘了说，没有构造函数消息被打印，但 calc 和内核消息是。

EDIT2：是否保证 CUDA在将它复制到设备上之前会初始化一个 Test 对象？

4

2 回答 2

4

您必须像普通方法一样查看构造函数。如果你用 来限定它__host__，那么你就可以称它为主机端。如果你用 来限定它__device__，你就可以称之为设备端。如果你同时满足它，你就可以在两边都调用它。

当您这样做时，会发生dosome<<<1, 2>>>(2, 3);两个对象是隐式构造的（因为您的构造函数不是explicit，所以也许这也让您感到困惑）主机端，然后memcpy'd 到设备。该过程不涉及复制构造函数。

让我们来说明一下：

    __global__ void dosome(Test a, Test b) {
        a.calc();
        b.calc();
    }

    int main(int argc, char **argv) {
        dosome<<<1, 2>>>(2, 3); // Constructors must be at least __host__
        return 0;
    }

// Outputs:
Construct 2 (from the host side)
Construct 3 (from the host side)

现在，如果您将内核更改为ints 而不是Test：

__global__ void dosome(int arga, int argb) {
    // Constructors must be at least __device__
    Test a(arga);
    Test b(argb);
    a.calc();
    b.calc();
}

int main(int argc, char **argv) {
    dosome<<<1, 2>>>(2, 3);
    return 0;
}

// Outputs:
Construct 2 (from the device side)
Construct 3 (from the device side)
于 2013-02-17T13:45:13.583 回答
3

好的，如果我将 __host__ 和 __device__ 限定符添加到构造函数中，我发现它可以工作（调用构造函数）。对象的构造函数发生在主机端，然后它们被复制到设备（堆栈？）。这就是为什么没有调用构造函数的原因：它们是设备代码（但是在主机端调用了什么？！？）

在构造函数中同时使用 __host__ 和 __device__ 可以毫无问题地使用该类。

编辑：不过，我不确定构造是否总是在复制到设备之前发生。

于 2013-02-17T12:28:44.103 回答