在cuda中并行执行内核

Question

可以说我有三个全局数组，这些全局数组已使用cudaMemcpy复制到了gpu中，但是c中的这些gloabl数组尚未使用cudaHostAlloc进行分配，因此无法分配页面锁定的内存，而是简单的gloabl分配。

int a[100],b [100],c[100];
cudaMemcpy(d_a,a,100*sizeof(int),cudaMemcpyHostToDevice);
cudaMemcpy(d_b,b,100*sizeof(int),cudaMemcpyHostToDevice);
cudaMemcpy(d_c,c,100*sizeof(int),cudaMemcpyHostToDevice);

现在我有10个内核，它们在单独的流中启动，以便可以并行运行，其中一些正在使用在gpu中复制的全局数组。现在这些内核正在运行1000次迭代。他们不需要在迭代过程中将任何内容复制回主机。

但是问题在于它们不是并行执行，而是以串行方式执行。

  cudaStream_t stream[3];

   for(int i=0;i<3;i++)cudaStreamCreate (&stream[i]);

   for(int i=0;i<100;i++){
       kernel1<<<blocks,threads,0,stream[0]>>>(d_a,d_b);
       kernel2<<<blocks,threads,0,strea[1]>>(d_b,d_c);
      kernal3<<<blocks,threads,0,stream[2]>>>(d_c,d_a);
    cudaDeviceSynchronize();
  }

我不明白为什么？

Answer 1

内核以这种方式发布：

   for(int i=0;i<100;i++){
    kernel1<<<blocks,threads>>>(d_a,d_b);
    kernel2<<<blocks,threads>>>(d_b,d_c);
    kernal3<<<blocks,threads>>>(d_c,d_a);
    cudaDeviceSynchronize();
  }

将始终串行运行。为了使内核能够同时运行，必须将它们发出以单独的CUDA流。并且还有其他要求。阅读文档。

您需要创建一些CUDA流，然后像这样启动内核：

   cudaStream_t stream1, stream2, stream3;
   cudaStreamCreate(&stream1); cudaStreamCreate(&stream2); cudaStreamCreate(&stream3);

   for(int i=0;i<100;i++){
    kernel1<<<blocks,threads,0,stream1>>>(d_a,d_b);
    kernel2<<<blocks,threads,0,stream2>>>(d_b,d_c);
    kernal3<<<blocks,threads,0,stream3>>>(d_c,d_a);
    cudaDeviceSynchronize();
  }

实际上，见证并发内核执行通常也将需要资源利用率有限的内核。如果给定的内核由于大量的块，每个块的线程，共享内存的使用或其他资源的使用而“填充”了计算机，那么您实际上将不会看到并发性。机器上没有空了。

您还可以查看一些CUDA示例代码，如simpleStreams和concurrentKernels。