Как разделить код CUDA на несколько файлов

Я пытаюсь разделить программу CUDA на два отдельных файла.cu, чтобы приблизиться к написанию реального приложения на C++. У меня есть простая маленькая программа, которая:

Выделяет память на хосте и устройстве.
Инициализирует массив хостов серией чисел. Копирует массив хостов в массив устройств. Находит квадрат всех элементов массива с помощью ядра устройства. Копирует массив устройств обратно в массив хостов. Печатает результаты.

Это прекрасно работает, если я положу все это в один файл.cu и запусту его. Когда я делю его на два отдельных файла, у меня появляются ошибки связывания. Как и все мои недавние вопросы, я знаю, что это что-то маленькое, но что это?

KernelSupport.cu

#ifndef _KERNEL_SUPPORT_
#define _KERNEL_SUPPORT_

#include <iostream>
#include <MyKernel.cu>

int main( int argc, char** argv) 
{
    int* hostArray;
    int* deviceArray;
    const int arrayLength = 16;
    const unsigned int memSize = sizeof(int) * arrayLength;

    hostArray = (int*)malloc(memSize);
    cudaMalloc((void**) &deviceArray, memSize);

    std::cout << "Before device\n";
    for(int i=0;i<arrayLength;i++)
    {
        hostArray[i] = i+1;
        std::cout << hostArray[i] << "\n";
    }
    std::cout << "\n";

    cudaMemcpy(deviceArray, hostArray, memSize, cudaMemcpyHostToDevice);
    TestDevice <<< 4, 4 >>> (deviceArray);
    cudaMemcpy(hostArray, deviceArray, memSize, cudaMemcpyDeviceToHost);

    std::cout << "After device\n";
    for(int i=0;i<arrayLength;i++)
    {
        std::cout << hostArray[i] << "\n";
    }

    cudaFree(deviceArray);
    free(hostArray);

    std::cout << "Done\n";
}

#endif

MyKernel.cu

#ifndef _MY_KERNEL_
#define _MY_KERNEL_

__global__ void TestDevice(int *deviceArray)
{
    int idx = blockIdx.x*blockDim.x + threadIdx.x;
    deviceArray[idx] = deviceArray[idx]*deviceArray[idx];
}


#endif

Журнал сборки:

1>------ Build started: Project: CUDASandbox, Configuration: Debug x64 ------
1>Compiling with CUDA Build Rule...
1>"C:\CUDA\bin64\nvcc.exe"    -arch sm_10 -ccbin "C:\Program Files (x86)\Microsoft Visual Studio 9.0\VC\bin"    -Xcompiler "/EHsc /W3 /nologo /O2 /Zi   /MT  "  -maxrregcount=32  --compile -o "x64\Debug\KernelSupport.cu.obj" "d:\Stuff\Programming\Visual Studio 2008\Projects\CUDASandbox\CUDASandbox\KernelSupport.cu" 
1>KernelSupport.cu
1>tmpxft_000016f4_00000000-3_KernelSupport.cudafe1.gpu
1>tmpxft_000016f4_00000000-8_KernelSupport.cudafe2.gpu
1>tmpxft_000016f4_00000000-3_KernelSupport.cudafe1.cpp
1>tmpxft_000016f4_00000000-12_KernelSupport.ii
1>Linking...
1>KernelSupport.cu.obj : error LNK2005: __device_stub__Z10TestDevicePi already defined in MyKernel.cu.obj
1>KernelSupport.cu.obj : error LNK2005: "void __cdecl TestDevice__entry(int *)" (?TestDevice__entry@@YAXPEAH@Z) already defined in MyKernel.cu.obj
1>D:\Stuff\Programming\Visual Studio 2008\Projects\CUDASandbox\x64\Debug\CUDASandbox.exe : fatal error LNK1169: one or more multiply defined symbols found
1>Build log was saved at "file://d:\Stuff\Programming\Visual Studio 2008\Projects\CUDASandbox\CUDASandbox\x64\Debug\BuildLog.htm"
1>CUDASandbox - 3 error(s), 0 warning(s)
========== Build: 0 succeeded, 1 failed, 0 up-to-date, 0 skipped ==========

Я использую Visual Studio 2008 в Windows 7 64bit.

Редактировать:

Я думаю, что мне нужно немного подробнее остановиться на этом. Конечный результат, который я здесь ищу, - это создание нормального приложения на C++ с чем-то вроде Main.cpp с int main() событие, и вещи бегут оттуда. В определенный момент в моем коде.cpp я хочу иметь возможность ссылаться на биты CUDA. Поэтому я думаю (и поправьте меня, если здесь есть более стандартное соглашение), что я добавлю код ядра CUDA в их файлы.cu, а затем получу вспомогательный файл.cu, который позаботится о том, чтобы поговорить с устройством и вызвать функции ядра, а что нет.

Источник

4 ответа

Решение

Вы в том числе mykernel.cu в kernelsupport.cu, при попытке связать компилятор видит mykernel.cu дважды. Вам придется создать заголовок, определяющий TestDevice, и вместо этого включить его.

комментарий:

Как то так должно работать

// MyKernel.h
#ifndef mykernel_h
#define mykernel_h
__global__ void TestDevice(int* devicearray);
#endif

а затем измените включающий файл на

//KernelSupport.cu
#ifndef _KERNEL_SUPPORT_
#define _KERNEL_SUPPORT_

#include <iostream>
#include <MyKernel.h>
// ...

повторное редактирование

Пока заголовок, который вы используете в коде C++, не содержит каких-либо специфических для cuda вещей (__kernel__,__global__и т. д.) Вы должны хорошо связать C++ и cuda-код.

Источник

Если вы посмотрите на примеры кода CUDA SDK, у них есть extern C, который определяет функции ссылок, скомпилированные из файлов.cu. Таким образом, файлы.cu компилируются nvcc и связываются только с основной программой, тогда как файлы.cpp компилируются нормально.

Например, в marchingCubes_kernel.cu есть тело функции:

extern "C" void
launch_classifyVoxel( dim3 grid, dim3 threads, uint* voxelVerts, uint *voxelOccupied, uchar *volume,
                      uint3 gridSize, uint3 gridSizeShift, uint3 gridSizeMask, uint numVoxels,
                      float3 voxelSize, float isoValue)
{
    // calculate number of vertices need per voxel
    classifyVoxel<<<grid, threads>>>(voxelVerts, voxelOccupied, volume, 
                                     gridSize, gridSizeShift, gridSizeMask, 
                                     numVoxels, voxelSize, isoValue);
    cutilCheckMsg("classifyVoxel failed");
}

В то время как в marchingCubes.cpp (где находится main()) просто есть определение:

extern "C" void
launch_classifyVoxel( dim3 grid, dim3 threads, uint* voxelVerts, uint *voxelOccupied, uchar *volume,
                      uint3 gridSize, uint3 gridSizeShift, uint3 gridSizeMask, uint numVoxels,
                      float3 voxelSize, float isoValue);

Вы также можете поместить их в файл.h.

Источник

Получить разделение на самом деле довольно просто, пожалуйста, ознакомьтесь с этим ответом, чтобы узнать, как его настроить. Затем вы просто помещаете свой код хоста в файлы.cpp, а код устройства в файлы.cu, правила сборки сообщают Visual Studio, как связать их вместе в конечный исполняемый файл.

Непосредственная проблема в вашем коде, которую вы определяете __global__ TestDevice функционировать дважды, один раз, когда вы #include MyKernel.cu и один раз, когда вы самостоятельно компилируете MyKernel.cu.

Вам также нужно поместить оболочку в файл.cu - в тот момент, когда вы звоните TestDevice<<<>>> из вашей основной функции, но когда вы переместите это в.cpp файл, он будет скомпилирован с cl.exe, который не понимает <<<>>> синтаксис. Поэтому вы просто позвоните TestDeviceWrapper(griddim, blockdim, params) в файле.cpp и предоставьте эту функцию в вашем файле.cu.

Если вам нужен пример, образец SobolQRNG в SDK обеспечивает хорошее разделение, хотя он все еще использует cutil, и я всегда рекомендовал бы избегать cutil.

Источник

Простое решение - отключить сборку файла MyKernel.cu.

Свойства -> Общее -> Исключено из сборки

Лучшее решение imo - разделить ваше ядро ​​на файл cu и cuh, и включить его, например:

//kernel.cu
#include "kernel.cuh"
#include <cuda_runtime.h>

__global__ void increment_by_one_kernel(int* vals) {
  vals[threadIdx.x] += 1;
}

void increment_by_one(int* a) {
  int* a_d;

  cudaMalloc(&a_d, 1);
  cudaMemcpy(a_d, a, 1, cudaMemcpyHostToDevice);
  increment_by_one_kernel<<<1, 1>>>(a_d);
  cudaMemcpy(a, a_d, 1, cudaMemcpyDeviceToHost);

  cudaFree(a_d);
}


//kernel.cuh
#pragma once

void increment_by_one(int* a);


//main.cpp
#include "kernel.cuh"

int main() {
  int a[] = {1};

  increment_by_one(a);

  return 0;
}
Источник
Другие вопросы по тегам