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);というと、2 つのオブジェクトが暗黙のうちに構築され (コンストラクターが ではないためexplicit、混乱するかもしれません)、ホスト側でmemcpyデバイスに渡されます。プロセスに関与するコピー コンストラクターはありません。

これを説明しましょう:

    __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)

int代わりに sを取るようにカーネルを変更すると、次のようになり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 に答える