【CUDA】将核函数编译成一个库，并通过main函数调用对程序功能进行实现

水瓶子里不装水

已于 2024-08-05 23:07:50 修改

阅读量958

点赞数 19

文章标签：开发语言

于 2024-08-05 22:58:52 首次发布

本文链接：https://blog.youkuaiyun.com/m0_74152439/article/details/140936398

版权

问题描述即程序功能实现

首先进行整体文件的架构

在CUDALib下，进行头文件以及核函数的内容进行编译并导出为动态库

在CUDApp下，编写main.cu文件内容，使其调用CUDALib内的动态库并最终对程序功能进行实现

问题描述即程序功能实现

问题：使用c++建立一个长度为2000的数组，赋值0到1999，使用cuda编程，将数组中每一个元素的值加1，即变成1到2000。注意一开始数组定义在cpu上，需要将其拷贝到GPU全局内存（核函数可读写），运算完成后，将数组由gpu内存拷贝到cpu内存，最后使用std::cout打印全部数组元素结果。其中要求把核函数编译成一个库，然后main函数调用它，得出最终结果。

首先进行整体文件的架构

CUDATEST
├─CUDAApp
│ ├─build
│ └─src
└─CUDALib
├─build
├─cmake
│ ├─include
│ └─lib
└─src

在CUDALib下，进行头文件以及核函数的内容进行编译并导出为动态库

这段代码是一个头文件的内容，它定义了 CUDATEST_API 宏，用于控制函数和变量的导入和导出属性。这个宏的使用依赖于预处理器变量 CUDATEST_EXPORTS 的定义，该变量通常在构建 DLL 时设置。AddKernelout 函数的声明指示这个函数是用来在 CUDA 内核和主机代码之间进行交互的。

而导出的函数一般需要定义为extern "c"，意思是该函数视为c语言进行编译。因为如果使用c++语法进行编译时，c++函数多态在不同的c++编译器下，会给函数重命名为不同的名字，你不能保证你的库的调用方和库的生成方使用的是一个c++编译器，就是别人去使用你的库，如果和你的编译器不同，可能用另外一个名字去找你dll里的函数，就可能找不到了。但是c语言编译器，对函数名有固定的要求，也就是说调用方按cudatest. h文件里的函数名，一定能找到dll里的同名函数。保证了你在调用库的时候能找到该库内包含的函数。

//cudatest.h
#pragma once

#include <cuda_runtime.h>

// 如果定义了 CUDATEST_EXPORTS，则使用 __declspec(dllexport) 来标记符号为导出，
// 这通常在编译 DLL 时发生。
// 否则，使用 __declspec(dllimport) 来标记符号为导入，这用于使用 DLL 的项目。
#ifdef CUDATEST_EXPORTS
#define CUDATEST_API __declspec(dllexport)
#else
#define CUDATEST_API __declspec(dllimport)
#endif

// CUDA Kernel 函数声明，添加了CUDA_API宏使其具有适当的导入/导出属性
extern "C" CUDATEST_API void AddKernelout(float *d_a, const int N);

在cudatest.cu中，由于cudatest.h文件中extern “c”表示该函数用c语言进行编译，而cuda函数无法用extern "C"修饰，所以必须使用host函数进行包裹一层才能对该函数进行定义最终调用。

//cudatest.cu
#include "cudatest.h"
#include <cuda_runtime.h>

// 使用 __global__ 关键字定义一个 CUDA 内核函数 AddKernel
// 这个函数将在 GPU 上并行执行，用于处理浮点型数组 d_a
__global__ void AddKernel(float *d_a, const int N) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < N) {
        d_a[idx] += 1.0f;
    }
}

// 使用之前在 "cudatest.h" 头文件中定义的 CUDATEST_API 宏来声明 AddKernelout 函数。
// 这个宏会根据编译环境决定函数是导出还是导入。
CUDATEST_API void AddKernelout(float *d_a, const int N){
    int blockSize = 256;
    int gridSize = (N + blockSize - 1) / blockSize;
// 调用 AddKernel 内核函数，并指定网格和块的大小。
    AddKernel<<<gridSize, blockSize>>>(d_a, N);
}

此CMake 代码段设置了一些变量，用于指定 CUDA 测试库的头文件目录、库文件目录和库文件名

cmake_minimum_required(VERSION 3.0)

# 设置 CUDATEST_INCLUDE_DIR 变量，指向包含 CUDA 头文件的目录。
set(CUDATEST_INCLUDE_DIR ${CMAKE_CURRENT_LIST_DIR}/include)

# 设置 CUDATEST_LIBRARY_DIR 变量，指向包含 CUDA 库文件的目录。
set(CUDATEST_LIBRARY_DIR ${CMAKE_CURRENT_LIST_DIR}/lib)

# 设置 CUDATEST_LIBRARIES 变量，定义了 CUDA 库文件的名称。
set(CUDATEST_LIBRARIES CUDAtest.dll)

该CMakeLists.txt 文件为构建 CUDA 动态库 CUDALibrary 进行了适当的设置。

cmake_minimum_required(VERSION 3.10)

project(CUDALibrary VERSION 1.0)

# 输出路径
set(EXECUTABLE_OUTPUT_PATH "${CMAKE_BINARY_DIR}/bin")
set(CMAKE_RUNTIME_OUTPUT_DIRECTORY ${EXECUTABLE_OUTPUT_PATH})
set(CMAKE_LIBRARY_OUTPUT_DIRECTORY ${EXECUTABLE_OUTPUT_PATH})
set(CMAKE_ARCHIVE_OUTPUT_DIRECTORY ${EXECUTABLE_OUTPUT_PATH})

# 启用 C++11 支持
set(CMAKE_CXX_STANDARD 11)
set(CMAKE_CXX_STANDARD_REQUIRED True)

# 查找 CUDA
find_package(CUDA REQUIRED)

# 指定 CUDAtest 包的查找路径
set(CUDAtest_DIR "C:/Users/Admin/Desktop/zhangshuhan/LearnCUDA/CUDAtest2/CUDALib/cmake")
find_package(CUDAtest REQUIRED)

if(CUDA_FOUND)
    # 允许使用不受支持的编译器
    set(CUDA_NVCC_FLAGS ${CUDA_NVCC_FLAGS};-allow-unsupported-compiler)
endif()

# 包含 CUDA 头文件的路径
include_directories(${CUDA_INCLUDE_DIRS} ${CUDATEST_INCLUDE_DIR})

# 定义 CUDA 源文件
set(CUDA_SOURCES "src/cuda.cu")

# 定义导出符号
add_definitions(-DCUDATEST_EXPORTS)

# 创建动态库
cuda_add_library(CUDAtest SHARED ${CUDA_SOURCES})

# 消息输出
message(STATUS "CUDA INCLUDE DIRECTORIES: ${CUDA_INCLUDE_DIRS}")
message(STATUS "LIBRARY OUTPUT PATH: ${CMAKE_LIBRARY_OUTPUT_DIRECTORY}")

# 打印库文件的输出路径
message(STATUS "CMAKE_LIBRARY_OUTPUT_DIRECTORY: ${CMAKE_LIBRARY_OUTPUT_DIRECTORY}")

以上为对核函数进行编译并导出为库的文件内容，通过cmake命令进行编译执行后生成名为CUDAtest.dll的动态库

在CUDApp下，编写main.cu文件内容，使其调用CUDALib内的动态库并最终对程序功能进行实现

本main.cu 文件展示了一个典型的 CUDA 程序流程，包括主机端数组的初始化、设备端内存的分配、数据的传输、CUDA 内核的执行、结果的回传和内存的释放。

#include <cuda_runtime.h>
#include <iostream>
#include "cudatest.h"

int main()
{
    const int N = 2000;  // 数组长度
    const int nBytes = N * sizeof(float);  // 需要的内存大小

    // 在主机（CPU）端创建并初始化数组
    float *h_a = new float[N];
    for (int i = 0; i < N; i++)
    {
        h_a[i] = i;
    }
    // 在设备（GPU）端分配内存
    float *d_a;
    cudaMallocManaged((void **)&d_a, nBytes);
    // 将主机端的数组复制到设备端
    cudaMemcpy(d_a, h_a, nBytes, cudaMemcpyHostToDevice);
    // 调用 CUDA Kernel 函数
    AddKernelout(d_a, N);
    // 同步device 保证结果能正确访问
    cudaDeviceSynchronize();
    // 将结果从设备端复制回主机端
    cudaMemcpy(h_a, d_a, nBytes, cudaMemcpyDeviceToHost);
    // 打印结果
    for (int i = 0; i < N; i++)
    {
        std::cout << h_a[i] << ", ";
    }
    std::cout << std::endl;
    // 释放内存
    delete[] h_a;
    cudaFree(d_a);

    return 0;
}

CMakeLists.txt 文件为 CUDA 应用程序 CUDAApp 进行了适当的设置，包括项目配置、CUDA 包的查找、编译选项设置、源文件定义、链接库设置

# 指定最小 CMake 版本
cmake_minimum_required(VERSION 3.10)

# 项目名称和版本
project(CUDAApp VERSION 1.0)

# 设置 C++ 标准
set(CMAKE_CXX_STANDARD 11)
set(CMAKE_CXX_STANDARD_REQUIRED True)

# 查找 CUDA
find_package(CUDA REQUIRED)

# 指定 CUDAtest 包的查找路径
set(CUDAtest_DIR "C:/Users/Admin/Desktop/zhangshuhan/LearnCUDA/CUDAtest2/CUDALib/cmake")
find_package(CUDAtest REQUIRED)

if(CUDA_FOUND)
    # 允许使用不受支持的编译器
    set(CUDA_NVCC_FLAGS ${CUDA_NVCC_FLAGS};-allow-unsupported-compiler)
endif()

# 设置可执行文件的输出路径
set(EXECUTABLE_OUTPUT_PATH ${CMAKE_BINARY_DIR}/bin)

# 包含头文件路径
include_directories(${CUDA_INCLUDE_DIRS} ${CUDATEST_INCLUDE_DIR})

# 定义源文件
set(MAIN_SOURCE "src/main.cu")

# 指定动态库的路径，以便链接器可以找到它
link_directories(${CUDATEST_LIBRARY_DIR})

# 创建可执行文件
cuda_add_executable(CUDAApp ${MAIN_SOURCE})

# 链接动态库
target_link_libraries(CUDAApp ${CUDA_LIBRARIES} ${CUDATEST_LIBRARIES}) 

# 消息输出
message(STATUS "CUDA INCLUDE DIRECTORIES: ${CUDA_INCLUDE_DIRS}")
message(STATUS "EXECUTABLE OUTPUT PATH: ${EXECUTABLE_OUTPUT_PATH}")

以上为主函数主要内容以及对动态库进行调用的功能实现内容，通过cmake命令进行编译执行后实现程序功能运行。