Как мне скомпилировать код CUDA с Qt в среде Windows / MSVC?

Question

После долгого устранения неполадок мне удалось заставить мою небольшую тестовую программу работать с Qt Creator (из того, что я видел, у некоторых людей возникли проблемы из-за этого).Я делюсь здесь решением (см. Мой ответ), не стесняйтесь комментировать или исправлять вещи, которые можно улучшить, особенно если у кого-то есть решение проблем, упомянутых ниже .

Двапроблемы, с которыми я столкнулся:

• Возможно, есть способ заставить все компилировать и ссылаться одновременно, но когда я пробую это, я всегда получаю странную ошибку, утверждающую, что main.cpp не может бытьнайдено, тогда как все пути в MakeFile верны.

• Кроме того, я не знаю точно, почему, но использование опции -dlink или -dc для включения перемещаемого кода дает внешний символ _cudaRegisterLinkedBinaryошибка.Может ли перемещаемый код работать с отдельной компиляцией?Это проблема из-за выделения времени компиляции?

Answer 1

Вот возможное решение.

Основная идея заключалась в кодировании процедуры умножения матриц.Для этого я просто использовал файл matMul.cu, содержащий функцию-обертку, которая вызывает функцию ядра CUDA.Затем этот файл был скомпилирован с использованием nvcc с помощью следующей команды:

nvcc.exe -lib -o lib_cuda / matMul.lib -c matMul.cu

Имея файл .lib, я могдобавьте библиотеку в Qt, используя инструмент «Добавить библиотеку» со статическим связыванием, который автоматически добавляет последние 8 строк в файл .pro.

Вот файлы проекта:

.pro файл:

QT -= gui

CONFIG += c++11 console
CONFIG -= app_bundle

SOURCES +=  main.cpp

OTHER_FILES =+ matMul.cu

# The following library conflicts with something in Cuda
QMAKE_LFLAGS_RELEASE = /NODEFAULTLIB:msvcrt.lib
QMAKE_LFLAGS_DEBUG   = /NODEFAULTLIB:msvcrtd.lib
QMAKE_LFLAGS_DEBUG   = /NODEFAULTLIB:libcmt.lib


# Used to avoid conflicting flags between CUDA and MSVC files, should make everything static
QMAKE_CFLAGS_DEBUG      += /MTd
QMAKE_CFLAGS_RELEASE    += /MT
QMAKE_CXXFLAGS_DEBUG    += /MTd
QMAKE_CXXFLAGS_RELEASE  += /MT

# CUDA settings <-- may change depending on your system
CUDA_DIR        = "C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v9.1"            # Path to cuda toolkit install
SYSTEM_NAME     = x64

# include paths
CUDA_INC += $$CUDA_DIR\include
INCLUDEPATH += $$CUDA_INC

# library directories
QMAKE_LIBDIR += $$CUDA_DIR/lib/$$SYSTEM_NAME \

# Add the necessary CUDA libraries
LIBS += -lcuda -lcudart

# Add project related libraries containing kernels
win32: LIBS += -L$$PWD/lib_cuda/ -lmatMul_d

INCLUDEPATH += $$PWD/lib_cuda
DEPENDPATH += $$PWD/lib_cuda

win32:!win32-g++: PRE_TARGETDEPS += $$PWD/lib_cuda/matMul_d.lib
else:win32-g++: PRE_TARGETDEPS += $$PWD/lib_cuda/libmatMul_d.a

main.cpp:

#include <cmath>
#include <chrono>
#include <iostream>

#include <cuda.h>
#include <cuda_runtime.h>

typedef struct
{
    int width;
    int height;
    int stride;

    float *elements;
} Matrix;


void matMul_wrapper(Matrix &C, const Matrix &A, const Matrix &B, cudaDeviceProp devProp);

int main()
{
    int devCount;
    cudaGetDeviceCount(&devCount);

    cudaDeviceProp devProp;
    for(int i=0; i < devCount; ++i)
    {
     cudaGetDeviceProperties(&devProp, i);
     std::cout << "\nDevice: "                   << devProp.name << "\n";
     std::cout << "  Compute capability:     " << devProp.major << "\n";
     std::cout << "  Max threads per block:  " << devProp.maxThreadsPerBlock << "\n";
     std::cout << "  Warp size:              " << devProp.warpSize << "\n\n";
    }


    Matrix A {1000, 1000, 1, new float[1000*1000]};
    Matrix B {1000, 1000, 1, new float[1000*1000]};
    Matrix C {B.width, A.height, 1, new float[1000*1000]};


    for(int row=0; row < A.height; ++row)
    {
     for(int col=0; col < A.width; ++col)
         A.elements[row*A.width + col] = (float)(row*A.width + col) / (float)100000;
    }

    for(int row=0; row < B.height; ++row)
    {
     for(int col=0; col < B.width; ++col)
         B.elements[row*B.width + col] = (float)(row*B.width + col) / (float)100000;
    }

    std::cout << A.elements[20000] << '\n';

    matMul_wrapper(C, A, B, devProp);

    std::cout << A.elements[20000] << '\n';

    delete[] A.elements;
    delete[] B.elements;
    delete[] C.elements;

    return 0;
}

matMul.cu:

#include <cuda.h>
#include <cuda_runtime.h>

#define BLOCK_SIZE 16

typedef struct
{
    int width;
    int height;
    int stride;

    float *elements;
} Matrix;

__global__
void matMulKernel(Matrix C, const Matrix A, const Matrix B)
{
    int col = blockIdx.x * blockDim.x + threadIdx.x;
    int row = blockIdx.y * blockDim.y + threadIdx.y;
    int idx = row*C.width + col;

    float out = 0;
    if(idx < C.width * C.height)
    {
        for(int j=0; j < A.width; ++j)
            out += A.elements[row*A.width + j] * B.elements[j*B.width + col];
    }

    C.elements[idx] = out;

}


void matMul_wrapper(Matrix &C, const Matrix &A, const Matrix &B, cudaDeviceProp devProp)
{
    dim3 block(BLOCK_SIZE, BLOCK_SIZE, 1);
    dim3 grid(  (C.width + block.x - 1) / block.x,
                (C.height + block.y - 1) / block.y,
                1);

    Matrix d_A {A.width, A.height, A.stride};
    size_t size = A.height * A.width * sizeof(float);
    cudaMallocManaged(&d_A.elements, size);
    cudaMemcpy(d_A.elements, A.elements, size, cudaMemcpyHostToDevice);

    Matrix d_B {B.width, B.height, B.stride};
    size = B.height * B.width * sizeof(float);
    cudaMallocManaged(&d_B.elements, size);
    cudaMemcpy(d_B.elements, B.elements, size, cudaMemcpyHostToDevice);

    Matrix d_C {C.width, C.height, C.stride};
    size = C.height * C.width * sizeof(float);
    cudaMallocManaged(&d_C.elements, size);
    cudaMemcpy(d_C.elements, C.elements, size, cudaMemcpyHostToDevice);

    matMulKernel<<<grid, block>>>(d_C, d_A, d_B);

    cudaDeviceSynchronize();

    cudaMemcpy(C.elements, d_C.elements, size, cudaMemcpyDeviceToHost);

    cudaFree(d_A.elements);
    cudaFree(d_B.elements);
    cudaFree(d_C.elements);
}

Надеюсь, это поможет.