主页 / coding / 问题 / 79493903
Accepted
PkDrew
Asked: 2025-03-08 13:00:22 +0800 CST

对 CUDA 全局函数中的 printf 缓冲规则感到困惑

  • 772

下面的代码总是在任何其他内容之前打印出“Hello from the start”，而在任何其他内容之后打印出“Hello from the end”，这是为什么呢？

代码：

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

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


__global__ void add1InGPU( int *devArr, int n, int nx )
{
    int ix = blockIdx.x * blockDim.x + threadIdx.x;
    int iy = blockIdx.y * blockDim.y + threadIdx.y;
    int id = nx * iy + ix;

    int whateverNum = 5;

    if (id == whateverNum) printf("Hello from the start\n");
    
    if ( id < n ) {
        devArr[id] += 1; 
        printf("blockIdx.x: %d, threadIdx.x: %d, blockIdx.y: %d, threadIdx.y: %d, processing index %d\n", blockIdx.x, threadIdx.x, blockIdx.y, threadIdx.y, id );
    }
    else {
        printf("blockIdx.x: %d, threadIdx.x: %d, blockIdx.y: %d, threadIdx.y: %d, skipped\n", blockIdx.x, threadIdx.x, blockIdx.y, threadIdx.y);
    }

    if (id == whateverNum) printf("Hello from the end\n");
}

int main(void)
{
    
    int* d_arr = NULL;
    cudaMalloc(&d_arr, 16 * sizeof(int));
    cudaMemset(d_arr, 0, 16 * sizeof(int));

    int nElem = 16;
    dim3 block( 2, 2 );
    dim3 grid( 2, 2 );
    // dim3 grid( ( nElem + block.x - 1 ) / block.x );
    
    add1InGPU<<<grid, block>>> ( d_arr, nElem, 4 );
    cudaDeviceSynchronize();
    int *h_arr = (int*)malloc(16 * sizeof(int));
    cudaMemcpy(h_arr, d_arr, 10 * sizeof(int), cudaMemcpyDeviceToHost);
    cudaFree(d_arr);

    volatile int _ = 0;
    for (int i = 0; i < 10; i++) {
        _ += h_arr[i];
    }

    cudaDeviceReset();
    
    return 0;
}

输出：

Hello from the start
blockIdx.x: 1, threadIdx.x: 0, blockIdx.y: 0, threadIdx.y: 0, processing index 2
blockIdx.x: 1, threadIdx.x: 1, blockIdx.y: 0, threadIdx.y: 0, processing index 3
blockIdx.x: 1, threadIdx.x: 0, blockIdx.y: 0, threadIdx.y: 1, processing index 6
blockIdx.x: 1, threadIdx.x: 1, blockIdx.y: 0, threadIdx.y: 1, processing index 7
blockIdx.x: 0, threadIdx.x: 0, blockIdx.y: 1, threadIdx.y: 0, processing index 8
blockIdx.x: 0, threadIdx.x: 1, blockIdx.y: 1, threadIdx.y: 0, processing index 9
blockIdx.x: 0, threadIdx.x: 0, blockIdx.y: 1, threadIdx.y: 1, processing index 12
blockIdx.x: 0, threadIdx.x: 1, blockIdx.y: 1, threadIdx.y: 1, processing index 13
blockIdx.x: 1, threadIdx.x: 0, blockIdx.y: 1, threadIdx.y: 0, processing index 10
blockIdx.x: 1, threadIdx.x: 1, blockIdx.y: 1, threadIdx.y: 0, processing index 11
blockIdx.x: 1, threadIdx.x: 0, blockIdx.y: 1, threadIdx.y: 1, processing index 14
blockIdx.x: 1, threadIdx.x: 1, blockIdx.y: 1, threadIdx.y: 1, processing index 15
blockIdx.x: 0, threadIdx.x: 0, blockIdx.y: 0, threadIdx.y: 0, processing index 0
blockIdx.x: 0, threadIdx.x: 1, blockIdx.y: 0, threadIdx.y: 0, processing index 1
blockIdx.x: 0, threadIdx.x: 0, blockIdx.y: 0, threadIdx.y: 1, processing index 4
blockIdx.x: 0, threadIdx.x: 1, blockIdx.y: 0, threadIdx.y: 1, processing index 5
Hello from the end
c++

1 个回答

  1. Best Answer

    对 CUDA 全局函数中的 printf 缓冲规则感到困惑

    您之所以感到困惑，是因为没有规则。块的调度和执行顺序以及构成这些块的线程的 warp（在某些架构上甚至是单个线程）都是未定义的。

    下面的代码总是先打印出“Hello from the start”

    它并不总是按该顺序打印。相反，您只观察到它按该顺序打印。如果您使用明显更大的网格大小和/或不同的硬件运行内核，您可能会看到不同的结果。这是未定义行为的本质。有时它看起来完全可以预测，即使事实并非如此。

    • 2

相关问题