多个.cu、.h和.c文件之间的Nsight CUDA链接问题

Posted 2023-03-23

【中文标题】多个.cu、.h和.c文件之间的Nsight CUDA链接问题

【英文标题】：Nsight CUDA linkage problem between multiple .cu, .h and .c files

【发布时间】：2020-12-29 11:37:45

【问题描述】：

这是我第一次尝试在 NSight Ubuntu 中构建我的 CUDA 应用程序以从优化和分析中受益。此应用程序在 Ubuntu 20（或 18 ,16）中使用 nvcc（makefile）在终端上运行良好。我有多个 .cu、.c 和 .h 文件。所有文件都首先包含在 flags.h 文件中。我的代码以 main.cu （具有main() 函数）文件开头，该文件具有 #include "flags.h" 以确保包含所有文件以编译代码。 flags.h 有很多 #define 以及稍后用于不同的 .cu 和 .c 文件。

但是，在 NSight 内部，flags.h 中定义的 #define 参数在任何文件中都无法识别，并且出现错误。 以下是错误截图。 我附上了一个简单的 square_array 问题，拆分为 3 个文件（main.cu、flags.h 和 square_.cu）。

我无法在 NSight 中构建它。有人可以尝试构建它并请告诉我。 任何帮助或建议将不胜感激。

main.cu

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

#include "flags.h"


int main(void) 

    int i;
    int *a_h, *a_d;

    CUDA_CHECK_RETURN(cudaMalloc((void**) &a_d, sizeof(int) * WORK_SIZE));
    a_h = (int *)malloc(sizeof(int) * WORK_SIZE);        // Allocate array on host

    for (i = 0; i < WORK_SIZE; i++)
        a_h[i] = i+2.;

    int block_size = 4;
    int n_blocks = WORK_SIZE/block_size + (WORK_SIZE%block_size == 0 ? 0:1);

    sq_array<<<n_blocks, block_size>>>(a_d);

    CUDA_CHECK_RETURN(cudaGetLastError());
    CUDA_CHECK_RETURN(cudaMemcpy(a_h, a_d, sizeof(int) * WORK_SIZE, cudaMemcpyDeviceToHost));

    for (i = 0; i < WORK_SIZE; i++)
        printf("Input value: %d \n", a_h[i]  );

    CUDA_CHECK_RETURN(cudaFree((void*) a_d));
    CUDA_CHECK_RETURN(cudaDeviceReset());

    return 0;

标志.h

#ifndef FLAGS_H_
#define FLAGS_H_

#include "square_.cu"
#define CUDA_CHECK_RETURN(value)                                           \
    cudaError_t _m_cudaStat = value;                                        \
    if (_m_cudaStat != cudaSuccess)                                        \
        fprintf(stderr, "Error %s at line %d in file %s\n",                 \
                cudaGetErrorString(_m_cudaStat), __LINE__, __FILE__);       \
        exit(1);                                                            \
     


#define WORK_SIZE 29

#endif /* FLAGS_H_ */

square_.cu

__global__ void sq_array( int *a) 
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if (idx< WORK_SIZE) a[idx] = a[idx] * a[idx];

【问题讨论】：

你正在#include.cu 文件和编译.cu 文件，这意味着你将有sq_array 的两个定义。单独编译时，不指定定义WORK_SIZE的头文件。不要不包含.cu或.c文件——分别编译和链接。你需要安排一些事情，使sq_array和内核调用在同一个文件中编译一次。跨度>

@wcochran 我希望这应该是一个一键式构建过程，因为它可以在终端中使用一行 make 命令。这是解释我的问题的简短代码，但我的实际代码很长而且很复杂，有几十个内核，因此我不能像你建议的那样将内核调用和内核定义放在同一个文件中。另外，当我有几十个文件时，单独编译和链接会很繁琐。一些 youtube 视频会有所帮助。

它可以在命令行上运行，因为您没有显式编译 .cu 文件，但 IDE 会。你必须以不同的方式组织事物。我在使用 CUDA 的大量代码库上工作，我们通常的策略是创建封装 CUDA 内核代码和调用的“主机包装器”函数。使用#include "square_.cu" 是个坏主意——（你不会对.c 文件这样做）——包括flags.h 的每一段单独编译的代码都将定义sq_array 函数的另一个版本.

您能否在不将所有代码放在一个文件中的情况下发布解决方案？我在这里发布了这个简单的问题，以了解如何在 Nsight IDE 中“构建”大型代码（拆分为多个文件）。

【参考方案1】：

问题是您的 IDE 正在编译 square_.cu 和编译 main.cu，这也是再次编译 square_.cu，因为 flags.h 中的 #include "square_.cu" 为您提供了两个定义sq_array。当square_.cu 被编译时，WORK_SIZE 宏未定义给你一个编译时错误。当你在命令行编译时你没有编译square_.cu所以你避免了这个错误。

无论如何，#include .cu（或.c 文件）是个坏主意。这些应该单独编译，然后链接在一起。

您必须以不同的方式组织事物。 我不知道你的代码的细节，但你可以这样做：

square.cu:

#include "square.h"
  
__global__ void sq_array( int *a) 
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if (idx< WORK_SIZE) a[idx] = a[idx] * a[idx];


   
void host_sq_array(int *a_d) 
   int block_size = 4;
   int n_blocks = WORK_SIZE/block_size + (WORK_SIZE%block_size == 0 ? 0:1);   
   sq_array<<<n_blocks, block_size>>>(a_d);

square.h:

#ifndef SQUARE_H
#define SQUARE_H

#include "flags.h"  // REMOVE #include of .cu file!!!
void host_sq_array(int *a_d);

#endif

您可以放心地#include square.h，它只包括常量、类型定义和函数原型。