Introduction to Cuda C Programming-the impact of different grid and block sizes on Cuda built-in variables and how to determine the thread ID

Test Environment
The GPU platform tested is gtx660m and the computing power is 3.0.
First, we will introduce the functions provided by the GPU:

int printf(const char *format[, arg, ...]);

From the formatting of the kernel function to the host, only devices with computing power of 2. x or above are supported. Behavior is similar to Standard C. Here we use the output value of the built-in variable.
Core Function call Method

kernel<<<Dg, Db, Ns, S >>>

The DG type is dim3, indicating the size of each grid dimension. The DG. Z with the computing power of 1. X is equal to 1. Indicates that the number of loaded blocks is the product of each DG dimension.
DB is also of the dim3 type. It indicates the number of threads in each block under each dimension, and is equal to the product of each dimension of DB.
NS is the size of the shared memory of each block, in bytes. Is an optional parameter. The default value is 0.
The S type is cudastream_t, indicating the relevant stream. Is an optional parameter. The default value is 0.
Cuda built-in variables include:

• Griddim; Type: dim3; dim3 is initialized to 1 by default, indicating the size of each dimension of the grid
• Blockdim, type: dim3, indicating the size of each dimension of the block
• Blockidx of the uint3 type, which indicates the index of the block in the grid.
• Threadidx, of the uint3 type, indicates the thread index in the block.
• Warpsize, int type, indicating the Warp Size

Restrictions on built-in Variables
You can find the following limits from the hardware description of computing capacity 3.0:

• The grid dimension is three dimensions. The size of each dimension is limited to: X 2 ^ 31-1, and Y and Z are 65535.
• The block dimension is three dimensions. The size of each dimension is limited to: X and Y are 1024, and Z is 64.
• The maximum number of threads in each block is 1024, that is, the product of X, Y, and Z in blockdim cannot exceed 1024.
• The warp size is 32.

Next we will first figure out the impact of different grid and block sizes on the built-in Cuda variables.
The Code is as follows: grid is (2, 3, 1), and block size is (2, 2, 1 ):

#include "cuda_runtime.h"#include "device_launch_parameters.h"#include <stdio.h>__global__ void MyKernel(){    printf("(%d,%d,%d) (%d, %d, %d) (%d, %d, %d) (%d, %d, %d) %d\n",        gridDim.x, gridDim.y, gridDim.z,         blockDim.x, blockDim.y, blockDim.z,        blockIdx.x, blockIdx.y, blockIdx.z,        threadIdx.x, threadIdx.y, threadIdx.z,        warpSize);}int main(){    cudaError_t cudaStatus;    // Choose which GPU to run on, change this on a multi-GPU system.    cudaStatus = cudaSetDevice(0);    if (cudaStatus != cudaSuccess) {        fprintf(stderr, "cudaSetDevice failed! Do you have a CUDA-capable GPU installed?");    }    printf("gridDim | blockDim | blockIdx | threadIdx | warpSize\n");    dim3 gridSize(2, 3, 1);    dim3 blockSize(2, 2, 1);    // Launch a kernel on the GPU with one thread for each element.    MyKernel<<<gridSize, blockSize>>>();    // Check for any errors launching the kernel    cudaStatus = cudaGetLastError();    if (cudaStatus != cudaSuccess) {        fprintf(stderr, "addKernel launch failed: %s\n",         cudaGetErrorString(cudaStatus));    }    // cudaDeviceSynchronize waits for the kernel to finish, and returns    // any errors encountered during the launch.    cudaStatus = cudaDeviceSynchronize();    if (cudaStatus != cudaSuccess) {        fprintf(stderr, "cudaDeviceSynchronize returned error code %d after launching addKernel!\n", cudaStatus);    }    // cudaDeviceReset must be called before exiting in order for profiling and    // tracing tools such as Nsight and Visual Profiler to show complete traces.    cudaStatus = cudaDeviceReset();    if (cudaStatus != cudaSuccess) {        fprintf(stderr, "cudaDeviceReset failed!");        return 1;    }    return 0;}

Note: you need to set the options for generating Cuda code:

The execution result is as follows:

It indicates that the first parameter of the core function input is set to griddim, the second is set to blockdim, and then griddim affects the index value of blockidx. blockdim affects the threadidx index value, and the warpsize is 32.