【CUDA】认识CUDA

void CPUFunction()
{
  printf("This function is defined to run on the CPU.\n");
}
__global__ void GPUFunction()
{
  printf("This function is defined to run on the GPU.\n");
}

int main()
{
  CPUFunction();
  GPUFunction<<<1, 1>>>();
  cudaDeviceSynchronize();
  return 0;
}

__global__ void GPUFunction()

__global__ 关键字表明以下函数将在 GPU 上运行并可全局调用
将在 CPU 上执行的代码称为主机代码，而将在 GPU 上运行的代码称为设备代码
注意返回类型为 void，使用 __global__ 关键字定义的函数要求返回 void 类型

GPUFunction<<<1, 1>>>();

当调用要在 GPU 上运行的函数时，将此种函数称为已启动的核函数
启动核函数时，必须提供执行配置，即在向核函数传递任何预期参数之前使用 <<< … >>> 语法完成的配置。在宏观层面，程序员可通过执行配置为核函数启动指定线程层次结构，从而定义线程组（称为线程块）的数量，以及要在每个线程块中执行的线程数量

cudaDeviceSynchronize();

与许多 C/C++ 代码不同，核函数启动方式为异步：CPU 代码将继续执行而无需等待核函数完成启动。调用 CUDA 运行时提供的函数 cudaDeviceSynchronize 将导致主机 (CPU) 代码暂作等待，直至设备 (GPU) 代码执行完成，才能在 CPU 上恢复执行

三、CUDA关键字

_global__关键字

__global__执行空间说明符将函数声明为内核。其功能是：

在设备上执行
可从主机调用，可在计算能力为 3.2或更高的设备调用
__global__ 函数必须具有 void 返回类型，并且不能是类的成员函数
对 global 函数的任何调用都必须指定其执行配置
对 global 函数的调用是异步的，这意味着其在设备完成执行之前返回

__device__关键字

在设备上执行
只能从设备调用
__global__ 和 __device__ 执行空间说明符不能一起使用

__host__关键字

在主机上执行
只能从主机调用
__global__ 和 __host__ 执行空间说明符不能一起使用
__device__ 和 __host__ 执行空间说明符可以一起使用，此时该函数是为主机和设备编译的

四、device管理

4.1 初始化

当第一次调用任何CUDA运行时API（如cudaMalloc、cudaMemcpy等）时，CUDA Runtime会被初始化。这个初始化过程包括设置必要的内部数据结构、分配资源等，以便CUDA运行时能够管理后续的CUDA操作

每个CUDA设备都有一个与之关联的主上下文。主上下文是设备上的默认上下文，当没有显式创建任何上下文时，所有的CUDA运行时API调用都会在该主上下文中执行。主上下文包含了设备上的全局资源，如内存、纹理、表面等

开发者可以在程序启动时显式地指定哪个GPU成为"默认"设备。这个变化通常通过设置环境变量CUDA_VISIBLE_DEVICES或在程序中使用CUDA API（如cudaSetDevice）显式选择设备来实现。一旦选择了设备，随后的CUDA运行时初始化就会在这个指定的设备上创建主上下文

在没有显式指定设备的情况下，CUDA程序会默认在编号为0的设备（通常是第一个检测到的GPU）上执行操作

可以设置环境变量CUDA_VISIBLE_DEVICES-2来屏蔽其他GPU，这样只有GPU2能被使用。也可以使用CUDA_VISIBLE_DEVICES-2,3来设置多个GPU，其 device ID 分别为0和1

cudaDeviceReset

其作用是重置当前线程所关联的CUDA设备的状态，并释放该设备上所有已分配并未释放的资源

使用场景：

在程序结束时，调用该函数可以确保所有已分配的GPU资源都被正确释放，避免内存泄漏
若在程序的执行过程中遇到错误或需要中途退出，可释放已分配的资源，确保设备状态正确
在某些情况下，若设备状态出错（如由于之前的错误操作导致设备进入不可预测的状态），调用该函数可以尝试恢复设备到一个可用的状态

注意：

在调用该函数前，应确保所有已分配的设备内存和其他资源都已被正确地处理（如过cudaFree释放内存）。尽管其会释放这些资源，但最好还是在代码中显式地进行释放，以提高代码的可读性和可维护性
调用该函数后，当前线程与设备的关联关系可能会被重置。若需要继续使用设备，可能需要重新调用cudaSetDevice来设置当前线程要使用的设备

4.2 Runtime API查询GPU信息

cudaError_t cudaGetDeviceProperties(cudaDeviceProp *prop, int device);

GPU的信息被存放在cudaDeviceProp结构体中

#include <cuda_runtime_api.h>
#include <iostream>
#include <cmath>
using namespace std;

int main()
{
    // 获取GPU数量
    int deviceCount = 0;
    cudaError_t errorId = cudaGetDeviceCount(&deviceCount);
    if (errorId != cudaSuccess) {
        printf("cudaGetDeviceCount returned %d\n-> %s\n", static_cast<int>(errorId), cudaGetErrorString(errorId));
        printf("Result = FAIL\n");
        exit(EXIT_FAILURE);
    }
    if (deviceCount == 0) {
        printf("There are no available device(s) that support CUDA\n");
    } else {
        printf("Detected %d CUDA Capable device(s)\n", deviceCount);
    }

    // 指定第一个GPU
    int device = 0;
    cudaSetDevice(device);
    // 获取GPU信息
    cudaDeviceProp deviceProp;
    cudaGetDeviceProperties(&deviceProp, device);
    int driverVersion = 0, runtimeVersion = 0;
    cudaDriverGetVersion(&driverVersion);
    cudaRuntimeGetVersion(&runtimeVersion);
    // 打印信息
    printf(" Device %d: \"%s\"\n", device, deviceProp.name);
    printf(" CUDA Driver Version / Runtime Version %d.%d / %d.%d\n", driverVersion/1000, (driverVersion%100)/10,runtimeVersion/1000, (runtimeVersion%100) / 10);
    printf(" CUDA Capability Major/Minor version number: %d.%d\n", deviceProp.major, deviceProp.minor);
    printf(" 全局内存总量: %.2f MBytes (%llu bytes)\n", (float)deviceProp.totalGlobalMem/(pow(1024.0,3)), static_cast<unsigned long long>(deviceProp.totalGlobalMem));
    printf(" GPU Clock rate: %.0f MHz (%0.2f GHz)\n", deviceProp.clockRate * 1e-3f, deviceProp.clockRate * 1e-6f);
    printf(" Memory Clock rate: %.0f Mhz\n", deviceProp.memoryClockRate * 1e-3f);
    printf(" Memory Bus Width: %d-bit\n", deviceProp.memoryBusWidth);

    if (deviceProp.l2CacheSize) {
        printf(" L2 Cache Size: %d bytes\n",
        deviceProp.l2CacheSize);
    }

    printf(" Max Texture Dimension Size (x,y,z) 1D=(%d), 2D=(%d,%d), 3D=(%d,%d,%d)\n",
    deviceProp.maxTexture1D , deviceProp.maxTexture2D[0],
    deviceProp.maxTexture2D[1],
    deviceProp.maxTexture3D[0], deviceProp.maxTexture3D[1],
    deviceProp.maxTexture3D[2]);

    printf(" Max Layered Texture Size (dim) x layers 1D=(%d) x %d, 2D=(%d,%d) x %d\n",
    deviceProp.maxTexture1DLayered[0], deviceProp.maxTexture1DLayered[1],
    deviceProp.maxTexture2DLayered[0], deviceProp.maxTexture2DLayered[1],
    deviceProp.maxTexture2DLayered[2]);

    printf(" 常量内存总量: %lu bytes\n",deviceProp.totalConstMem);

    printf(" 每个块的共享内存总量: %lu bytes\n",deviceProp.sharedMemPerBlock);

    printf(" 每个块可用的寄存器总数: %d\n",deviceProp.regsPerBlock);

    printf(" Warp size: %d\n", deviceProp.warpSize);

    printf(" 每个多处理器的最大线程数: %d\n",deviceProp.maxThreadsPerMultiProcessor);

    printf(" 每个块的最大线程数: %d\n",deviceProp.maxThreadsPerBlock);

    printf(" 块各维度的最大尺寸: %d x %d x %d\n", deviceProp.maxThreadsDim[0], deviceProp.maxThreadsDim[1], deviceProp.maxThreadsDim[2]);

    printf(" 网格每个维度的最大尺寸: %d x %d x %d\n", deviceProp.maxGridSize[0], deviceProp.maxGridSize[1], deviceProp.maxGridSize[2]);

    printf(" Maximum memory pitch: %lu bytes\n", deviceProp.memPitch);

    return 0;
}

4.3 决定最佳GPU

对于支持多GPU的系统，需从中选择一个来作为device，抉择出最佳计算性能GPU的一种方法就是由其拥有的处理器数量决定

int main()
{
    int numDevices = 0;
    cudaGetDeviceCount(&numDevices);
    if (numDevices > 1) 
    {
        int maxMultiprocessors = 0, maxDevice = 0;
        for (int device=0; device < numDevices; ++device) 
        {
            cudaDeviceProp props;
            cudaGetDeviceProperties(&props, device);
            if (maxMultiprocessors < props.multiProcessorCount) {
                maxMultiprocessors = props.multiProcessorCount;
                maxDevice = device;
            }
        }
        cudaSetDevice(maxDevice);
    }  
    return 0;
}

原文地址：https://blog.csdn.net/GG_Bruse/article/details/143628110

免责声明：本站文章内容转载自网络资源，如本站内容侵犯了原著者的合法权益，可联系本站删除。更多内容请关注自学内容网（zxcms.com）！

上一篇：redis bind 127.0.0.1和bind 10.34.56.78的区别
下一篇：EasyExcel的AbstractColumnWidthStyleStrategy注入CellStyle不生效

量子力学（1）如何理解海森堡的不确定性原理
海森堡不确定性原理
阅读更多2024-11-09
5G-A如何与AI融合发展？华为MBBF2024给出解答
2024年，对于整个ICT产业来说，是极为重要的一年。一方面，随着3GPP R18标准的正式冻结，我们全面进入了5G-A时代，迎来了这项全新技术标准的规模商用。另一方面，在经历了“百模大战”的喧嚣之后
阅读更多2024-11-09
C++ Double Dispatch，即双重调度
对于初次接触或曾有所闻但未曾深入了解的开发人员而言，调度（Dispatching）这一概念值得我们深入探讨。调度，从字面意义上理解，即将某物发送至特定目的地。在C++编程中，调度同样指将控制权从一处转
阅读更多2024-11-09
编译器、IDE对C/C++新标准的支持
很遗憾的是，微软VC++对C99的支持直到VS2015才基本完全支持，在VS2012之前，仅仅支持C89. 也正是2015年之前，这么多人骂微软并纷纷转移到其它C语言编译器的原因。VS2019开始才正
阅读更多2024-11-09
css:基础
陶喆 {font-family: '宋体'.周杰伦 {可以看见部分字体改变了我们还可以设置多个字体，可以方便用户在查看网页的时候，如果没安装第一种字体，你们就执行第二种字体.eat {一般都使用系统默
阅读更多2024-11-09
FPGA课程内容升级【纯逻辑和图像开发】
手写axi lite控制器（master和slave）mipi-datamover-hdmi应用实例。mipi-axi-dma-hdmi应用实例。datamover读写 PL DDR。mipi-vdm
阅读更多2024-11-09
前端页面性能优化的常见问题与解决方案
在当今互联网高速发展的时代，前端页面的性能对于用户体验至关重要。一个加载缓慢、交互卡顿的页面很可能会导致用户流失。本文将深入探讨前端页面性能优化中常见的问题以及相应的解决方案。
阅读更多2024-11-09
linux常见命令
linux常见命令
阅读更多2024-11-09
Java后端请求想接收多个对象入参的数据方法
在Java后端开发中，如果我们希望接收多个对象作为HTTP请求的入参，可以使用Spring Boot框架来简化这一过程。Spring Boot提供了强大的RESTful API支持，能够方便地处理各种
阅读更多2024-11-09
使用 PageHelper 在 Spring Boot 项目中实现分页查询
在 Web 应用开发中，分页查询是非常常见的需求，特别是在涉及大量数据的应用场景中，通过分页可以减少数据加载压力，提升系统性能。然而，手动实现分页查询需要编写繁琐的 SQL 语句和逻辑代码，容易出现错
阅读更多2024-11-09

【CUDA】认识CUDA

一、CUDA编程

二、第一个CUDA程序

三、CUDA关键字

四、device管理

4.1 初始化

4.2 Runtime API查询GPU信息

4.3 决定最佳GPU

相关文章