从每个主机线程启动 CUDA 流答案

【问题标题】：Launching a CUDA stream from each host thread从每个主机线程启动 CUDA 流
【发布时间】：2011-07-19 17:34:58
【问题描述】：

我的意图是使用 n 个主机线程在 NVidia Tesla C2050 上同时创建 n 个流。内核是一个简单的向量乘法......我将数据平均分配给 n 个流，每个流将同时进行执行/数据传输。

数据是浮点数，有时 CPU/GPU 总和相等，有时它们相差很大...我想这可能是由于我的代码中丢失了同步结构，就我而言，但也我认为流之间不需要任何同步构造，因为我希望每个 CPU 都有一个唯一的流来控制，并且我不关心线程内的异步数据复制和内核执行。

以下是每个线程运行的代码：

//every thread would run this method in conjunction

static CUT_THREADPROC solverThread(TGPUplan *plan)
{

    //Allocate memory
    cutilSafeCall( cudaMalloc((void**)&plan->d_Data, plan->dataN * sizeof(float)) );

    //Copy input data from CPU
    cutilSafeCall( cudaMemcpyAsync((void *)plan->d_Data, (void *)plan->h_Data, plan->dataN * sizeof(float), cudaMemcpyHostToDevice, plan->stream) );
    //to make cudaMemcpyAsync blocking
    cudaStreamSynchronize( plan->stream );

    //launch
    launch_simpleKernel( plan->d_Data, BLOCK_N, THREAD_N, plan->stream);
    cutilCheckMsg("simpleKernel() execution failed.\n");

    cudaStreamSynchronize(plan->stream);

    //Read back GPU results
    cutilSafeCall( cudaMemcpyAsync(plan->h_Data, plan->d_Data, plan->dataN * sizeof(float), cudaMemcpyDeviceToHost, plan->stream) );
    //to make the cudaMemcpyAsync blocking...               
    cudaStreamSynchronize(plan->stream);

    cutilSafeCall( cudaFree(plan->d_Data) );

    CUT_THREADEND;
}

并创建多个线程并调用上述函数：

    for(i = 0; i < nkernels; i++)
            threadID[i] = cutStartThread((CUT_THREADROUTINE)solverThread, &plan[i]);

    printf("main(): waiting for GPU results...\n");
    cutWaitForThreads(threadID, nkernels);

我从一个 CUDA 代码 SDK 示例中采用了这个策略。正如我之前所说，这段代码有时会起作用，而有时它会给出任性的结果。我需要帮助来修复此代码...

【问题讨论】：

标签： cuda

【解决方案1】：

首先，从我的经验来看，我不是专家。

我不明白为什么这需要多个主机线程。似乎您正在管理一台设备并为其传递多个流。我看到的方式（伪代码）

{
create a handle

allocate an array of streams equal to the number of streams you want

for(int n=0;n<NUM_STREAMS;n++)
{
   cudaStreamCreate(&streamArray[n]);
}

}

从那里您可以将数组中的流传递给各种异步调用（cudaMemcpyAsync()、内核流等），其余的由设备管理。我在多个流中遇到了奇怪的可扩展性问题（不要尝试制作 10k 流，我在 GTX460 上遇到了 4-8 左右的问题），所以如果你遇到这些问题不要感到惊讶。祝你好运，

约翰

【讨论】：

我目前正在测试各种在 GPU 和多个 CPU 线程上编程的方法，每个控制一个流是测试用例之一。这只是一项观察行为的研究。对于上面的示例，我只在单个 GPU 上创建了 4 个流。顺便说一句，我确实有一个测试用例，其中单个主机线程控制 GPU 上的 4 个流（正如您所建议的那样），并且对于一定数量的数据来说效果很好。

【解决方案2】：

我敢打赌

块_N，线程_N

，不要涵盖您传递的数组的确切大小。请提供初始化流的代码和这些缓冲区的大小。

顺便说一句，流对于内存传输的重叠计算很有用。在每次异步调用之后同步流根本没有用。

【讨论】：

感谢您为我指明方向！我的块线程尺寸是 32X512，每个线程都在创建一个流，考虑在上面的solverThread 函数中第一行是cudaStreamCreate(&plan->stream)