unique_ptr 和 cudaMalloc 的使用

Question

我一直在考虑在 CUDA 中使用 std::unique_ptr 和设备指针。我想知道的是当前的 c++11 unique_ptr 是否可以与 cudaMalloc 结合使用。我知道它可以与普通 malloc (Is it possible to use a C++ smart pointers together with C's malloc?) 一起使用，但 cudaMalloc 不会 return 函数的 return 语句中的指针。相反，它 return 是一个错误代码。指针在引用中被 returned。

本 blog post 推荐以下技术：

auto deleter=[&](float* ptr){ cudaFree(ptr); };
std::unique_ptr<float[], decltype(deleter)> d_in(new float[size], deleter);
cudaMalloc((void **) &d_in, size * sizeof(float));

问题： 但是，我担心这会创建永远不会被删除的主机内存（即 d_in(new float[size], deleter);）？除非 new float[size] 实际上不生成主机内存或被覆盖？如果以上实际上不起作用，可以定义我自己的 cudaMalloc 包装器吗？ - 将指针传递给 unique_ptr?

类似于：

void* myCudaMalloc(size_t mySize){ 
    void * p; 
    checkCUDAerrorMacro(cudaMalloc((void**) &p, size);) 
    return p;
}

...

auto deleter=[](float* ptr){ cudaFree(ptr); };
std::unique_ptr<float[], decltype(deleter)> d_in(myCudaMalloc(size_t mySize), deleter);

Answer 1

经过一些工作后，我想出了如何测试它的 3 个版本 - tl;dr 博客 post 的版本 (v1) 确实泄漏了，但可以进行调整以使其不泄漏 ( v2) 和改进 (v3):

常用代码：

template <typename Deleter>
using unique_p = std::unique_ptr<float[], Deleter>;

constexpr int length = 20;

v1:(博客里推荐的post)

void version1(){
    auto deleter = [](float* ptr) { cudaFree(ptr); std::cout<<"\nDeleted1\n"; };
    unique_p<decltype(deleter)> d_in(new float[length],deleter);
    cudaMalloc((void **) &d_in, length * sizeof(float));

    ...
}

v2：（与上面类似，但使用 nullptr 初始化 d_in）

void version2(){
    auto deleter = [](float* ptr) { cudaFree(ptr); std::cout<<"\nDeleted2\n"; };
    unique_p<decltype(deleter)> d_in(nullptr,deleter);
    cudaMalloc((void **) &d_in, length * sizeof(float));

    ...
}

v3: (d_in "adopts" 指针用 cudaMalloc 初始化)

void version3(){
    auto  myCudaMalloc = [](size_t mySize) { void* ptr; cudaMalloc((void**)&ptr, mySize); return ptr; };
    auto deleter = [](float* ptr) { cudaFree(ptr); std::cout<<"\nDeleted3\n"; };
    unique_p<decltype(deleter)> d_in((float*)myCudaMalloc(length*sizeof(float)),deleter);

    ...
}

所有 3 个都创建正确的设备指针。但是，版本 1 肯定会泄漏主机内存（使用 valgrind 测试并抑制了 cuda 警告：Valgrind and CUDA: Are reported leaks real?）。 v2 和 v3 都不会泄漏主机内存。 cuda-memcheck 还确认任何版本都没有设备端内存泄漏。

在第 2 版和第 3 版之间，我更喜欢第 3 版，因为它更清楚地表明 unique_ptr 拥有指针，并且它遵循 new 和 malloc 在 unique_ptr构造函数。您也只需定义一次构造 function/lambda ，然后就可以反复使用它，因此代码行数更少。

========================

完整测试代码（使用 nvcc -std=c++14 编译）：

#include <cuda_runtime.h>
#include <memory>
#include <iostream>

template <typename Deleter>
using unique_p = std::unique_ptr<float[], Deleter>;

__global__ void printArray(float * d_in, int num){
    for(int i = 0; i < num; i++){ printf("%f\t",d_in[i]); }
    printf("\n");

}

struct myDeleter{
    void operator()(float* ptr){ cudaFree(ptr); std::cout<<"\nDeleted\n"; } 
};

constexpr int length = 20;

void version1(){
    auto deleter = [](float* ptr) { cudaFree(ptr); std::cout<<"\nDeleted1\n"; };
    unique_p<decltype(deleter)> d_in(new float[length],deleter);
    cudaMalloc((void **) &d_in, length * sizeof(float));

    std::unique_ptr<float[]> h_out(new float[length]);

    for(int i = 0; i < length; i++){ h_out[i] = i; }

    cudaMemcpy(d_in.get(), h_out.get(),length*sizeof(float),cudaMemcpyHostToDevice);


    printArray<<<1,1>>>(d_in.get(),length);
}

void version2(){
    auto deleter = [](float* ptr) { cudaFree(ptr); std::cout<<"\nDeleted2\n"; };
    unique_p<decltype(deleter)> d_in(nullptr,deleter);
    cudaMalloc((void **) &d_in, length * sizeof(float));

    std::unique_ptr<float[]> h_out(new float[length]);

    for(int i = 0; i < length; i++){ h_out[i] = i; }

    cudaMemcpy(d_in.get(), h_out.get(),length*sizeof(float),cudaMemcpyHostToDevice);


    printArray<<<1,1>>>(d_in.get(),length);
}


void version3(){
    auto  myCudaMalloc = [](size_t mySize) { void* ptr; cudaMalloc((void**)&ptr, mySize); return ptr; };
    auto deleter = [](float* ptr) { cudaFree(ptr); std::cout<<"\nDeleted3\n"; };
    unique_p<decltype(deleter)> d_in((float*)myCudaMalloc(length*sizeof(float)),deleter);
    //unique_p<myDeleter> d_in((float*)myCudaMalloc(20*sizeof(float)));

    std::unique_ptr<float[]> h_out(new float[length]);
    for(int i = 0; i < length; i++){ h_out[i] = i; }

    cudaMemcpy(d_in.get(), h_out.get(),length*sizeof(float),cudaMemcpyHostToDevice);

    printArray<<<1,1>>>(d_in.get(),length);
}

int main(){

    version1();
    version2();
    version3();

    cudaDeviceReset();
    return 0;
}

Answer 2

这个模式对我很有效：

int main(){
float* deviceArray_raw;
gpuErrchk(cudaMalloc((void**)&deviceArray_raw, 100 * sizeof(float)));
auto deleter = [](float* ptr) { gpuErrchk(cudaFree(ptr)); };
std::unique_ptr<float[], decltype(deleter)> deviceArray(deviceArray_raw, deleter);
...
...
return 0;
}

除了主机内存泄漏，还需要注意设备内存泄漏。在 gpuErrchk 中结束 cuda API 调用有助于此。我能够使用它捕捉到一些奇怪的行为。

Answer 3

我推荐使用这个 git repo。 fun::cuda::shared_ptr 和 fun::cuda::unique_ptr 都已在此存储库中基于 std::shared_ptr 实现 Cuda device pointers。它还提供了许多其他函数来以直接的方式使用 Cuda 设备指针，包括：

bool create(size_t size);//! you can use create to allocate memory or simply upload your data without calling this function
void upload_async(const std::vector<T> &data_vec, cudaStream_t stream);
void upload(const std::vector<T> &data_vec);
void download_async(std::vector<T> &data_vec, cudaStream_t stream) const;
void download(std::vector<T> &data_vec) const;
void upload_async(const T* data_arr, int n_elements, cudaStream_t stream);
void upload(const T* data_arr, int n_elements);
void download_async(T* data_vec, cudaStream_t stream) const;
void download(T_ELEM* data_vec) const;

用法示例：

std::shared_ptr<T[]> data_host = std::shared_ptr<T[]>(new T[n]);
.
.
.

// In host code:
fun::cuda::shared_ptr<T> data_dev;
data_dev->upload(data_host.get(), n);
// In .cu file:
// data_dev.data() points to device memory which contains data_host;

这个存储库确实是一个单独的头文件 (cudasharedptr.h)，因此如果您的应用程序需要的话，它很容易操作。

unique_ptr 和 cudaMalloc 的使用

Use of unique_ptr and cudaMalloc

pointers

cuda

unique-ptr

c++11

用法示例：