CUDA内核未访问数组的所有元素

Question

我已经编写了一个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等，但它没有打印出数组的所有索引。理想情况下，它应该打印任何0到n-1的排列，但它只打印少于(<n)的数字。

如果numBlocks和blockSize都为1，则它将访问所有元素。如果数组大小小于4096，则也会访问所有元素。

- kayush206

---

你是如何计算数字的？ - Ander Biguri

我将此程序的输出导入文件，然后运行 cat 文件名 | wc -l 命令。 - kayush206

如果有任何答案对您有帮助，请考虑接受它们作为有效答案。 - Ander Biguri

Answer 1

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

此外，请记住内核中的printf调用执行无序。此外，设备上的printf缓冲区存在限制，这些限制在文档中有解释。

- sgarizvi

---

printf 的顺序错了没关系，但它难道不应该打印出 n 个数字的任意排列吗？ - kayush206

我猜测缓冲区没有被完全刷新。如文档所述，请尝试在代码结尾处调用cudaDeviceReset()。 - sgarizvi

cudaDeviceReset() 没有帮助 :( 又出现了相同的问题。 - kayush206

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

Answer 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);
}

- Ander Biguri

---

@kayush206 printf 可以用来调试，但是一定不要使用 100000 个 printf 来调试！ - Ander Biguri