CUDA学习笔记1——核函数与线程

阿卡蒂奥

已于 2023-11-28 20:35:55 修改

阅读量283

点赞数

分类专栏： CUDA 文章标签：学习 CUDA

于 2023-10-08 11:21:38 首次发布

本文链接：https://blog.csdn.net/akadiao/article/details/133314071

版权

CUDA 专栏收录该内容

10 篇文章 1 订阅

订阅专栏

CUDA安装并配置环境后，使用VS新建工程时会出现对应选项：
在这里插入图片描述

CUDA核函数调用

CUDA核函数必须用限定词 global 修饰，返回类型必须为void，二者次序随意。
函数名与()之间用三括号<<<网格大小，线程块大小>>>指明核函数中线程数目与排列情况；
cudaDeviceSynchronize()函数用于同步主机与设备；

#include <stdio.h>
#include "cuda_runtime.h"

//如何调用cuda核函数
__global__ void first_kernel()
{
	printf("Hellow World GPU...\n");
}

int main()
{
	printf("Hellow World CPU  \n");

	first_kernel<<<2, 3>>>();
	cudaDeviceSynchronize();
	return 0;
}

在这里插入图片描述

CUDA核函数线程索引

每个核函数允许指派多个线程，线程的组织结构为：<<<grid_size, block_size>>>，乘积为线程总数

对于大小为 $D_x , D_y)$ 二维网络，线程的索引为（x，y），则这个线程的ID是 $x + y * D_x)$
对于大小为 $D_x , D_y, D_z)$ 三维网络，线程的索引为（x，y，z），则这个线程的ID是 $x + y * D_x + z * D_x * D_y)$

一维网络

网格、线程块为一维

gridDim.x：数值为grid_size的数值
blockDim.x：数值为block_size的数值
blockIdx.x：线程在网格中的线程块ID，取值范围 0 ~ gridDim.x - 1；
threadIdx.x：线程在线程块中的线程ID，取值范围 0 ~ blockDim.x - 1；

例：

dim3 grid_size(4);//定义grid与block尺寸
dim3 block_size(8);

kernel_fun<<<grid_size,block_size>>>(...);//调用核函数

//计算线程索引
int id = blockIdx.x*blockDim.x+threadIdx.x;//线程块索引*线程块中线程的个数+线程在当前线程块中的索引

在这里插入图片描述

多维网络

网格、线程块为多维
dim3 grid_size(Gx,Gy,Gz);
dim3 block_size(Bx,By,Bz);

blockIdx.x：取值范围 0 ~ gridDim.x - 1；
blockIdx.y：取值范围 0 ~ gridDim.y - 1；
blockIdx.z：取值范围 0 ~ gridDim.z - 1；
threadIdx.x：取值范围 0 ~ blockDim.x - 1；
threadIdx.y：取值范围 0 ~ blockDim.y - 1；
threadIdx.z：取值范围 0 ~ blockDim.z - 1；

网格大小限制：x、y、z方向最大值为2^31-1、65535、65535

线程块大小限制：x、y、z方向最大值为1024、1024、64
其中，x维度是最内层（变化最快），z维度是最外层（变化最慢）

在这里插入图片描述

例：

dim3 grid_size(3,2);// 列优先
dim3 block_size(3,4);// 定义grid和block尺寸

kernel_fun<<< grid_size, block_size>>>(...);//调用核函数

//计算线程索引
int blockId = blockId.x + blockId.y*gridDim.x;//3=0+1*3
int threadId = threadIdx.y*blockDim.x+threadIdx.x;//10=3*3 + 1
//线程在网格中的唯一标识
int id = blockId*(blockDim.x*blockDim.y)+threadId;//46=3*(3*4) + 10

在这里插入图片描述

warpSize

warpSize为内建变量，表示线程束大小，对目前所以GPU架构warpSize=32；即一个线程束就是连续的32个线程。

#include <stdio.h>
#include "cuda_runtime.h"

//cuda核函数线程索引
__global__ void first_kernel()
{
	int tidx = threadIdx.x;
	int tidy = threadIdx.y;
	int tidz = threadIdx.z;
	int bidx = blockIdx.x;
	int bidy = blockIdx.y;
	int bidz = blockIdx.z;
	printf("GPU Kernel...thread index: (%d,%d), block index: (%d,%d) \n",  tidy, tidx, bidy, bidx);
}


int main()
{

	first_kernel<<<2, 3>>>();
	cudaDeviceSynchronize();
	return 0;
}