cuda 内核没有访问数组的所有元素

Posted 2023-04-15

【中文标题】cuda 内核没有访问数组的所有元素

【英文标题】：cuda kernel is not accessing all the element of an array

【发布时间】：2019-03-14 03:19:17

【问题描述】：

我编写了一个 cuda 程序来对大型数组进行一些操作。但是当我将该数组传递给 cuda 内核时，它的所有元素都不会被线程访问。下面，有一个简单的程序解释我的用例：

#include <stdio.h>
#include <stdlib.h>

__global__
void kernel(int n)
        int s = threadIdx.x + blockIdx.x*blockDim.x;
        int t = blockDim.x*gridDim.x;
        for(int i=s;i<n;i+=t)
        printf("%d\n",i);  //printing index of array which is being accessed
        


int main(void)
        int i,n = 10000; //array_size
        int blockSize = 64;
        int numBlocks = (n + blockSize - 1) / blockSize;
        kernel<<<numBlocks, blockSize>>>(n);
        cudaDeviceSynchronize();

我试过不同的blockSize = 256, 128, 64, etc，它没有打印数组的所有索引。理想情况下，它应该打印0 to n-1 的任何排列，但是它打印的是较小的(<n) 数字。

如果numBlocks 和blockSize 都为1，那么它正在访问所有元素。如果数组大小小于 4096，那么它也正在访问所有元素。

【问题讨论】：

你是怎么数数的？

我将该程序的输出通过管道传输到文件然后cat filename | wc -l

如果有帮助，请考虑接受任何有效的答案。

【参考方案1】：

实际上，所有值都在当前情况下打印。但由于输出控制台的缓冲区限制，您可能无法看到所有这些。尝试增加输出控制台的缓冲区大小。

此外，请记住内核内部的printf 调用是乱序执行的。此外，设备上的printf 缓冲区存在限制，在documentation 中进行了说明。

【讨论】：

printf 出现故障，没关系。但它不应该打印n 数字的任何排列。

我猜，缓冲区没有被完全刷新。如文档中所述，尝试在代码末尾调用cudaDeviceReset()。

cudaDeviceReset() 没有帮助:(又是同样的问题。

增加输出控制台的缓冲区大小可以解决问题。 cudaDeviceSetLimit(cudaLimitPrintfFifoSize, size_t size) 是设置缓冲区大小的必要 API。缓冲区大小默认为1MB。

【参考方案2】：

使用更好的调试技术！您的代码功能正常

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>

#include <stdlib.h>

__global__
void kernel(int* in, int n)
    int s = threadIdx.x + blockIdx.x*blockDim.x;
    int t = blockDim.x*gridDim.x;
    for (int i = s; i<n; i += t)
        in[i] = 1;  //printing index of array which is being accessed
    


int main(void)
    int i, n = 10000; //array_size
    int blockSize = 64;
    int numBlocks = (n + blockSize - 1) / blockSize;
    int* d_res,*h_res;
    cudaMalloc(&d_res, n*sizeof(int));
    h_res = (int*)malloc(n*sizeof(int));

    kernel << <numBlocks, blockSize >> >(d_res, n);
    cudaDeviceSynchronize();
    cudaMemcpy(h_res, d_res, n*sizeof(int), cudaMemcpyDeviceToHost);

    int sum = 0;
    for (int i = 0; i < n; i++)
        sum += h_res[i];
    printf("%d", sum);

【讨论】：

@kayush206 printf 可以是一种调试方式，但肯定 100000 printf 不是一种调试方式！