4

我在一个小测试文件上使用 cuda-memcheck 的泄漏检查来测试我想在我正在处理的程序中实现的一些功能，我发现它没有报告全局内存上的一些非常明显的内存泄漏，当我同时调用cudaMalloc()（来自主机代码）和malloc()（来自设备代码）时。似乎对设备的调用malloc()正在破坏cuda-memcheck.
我在 Windows 10 上的 NVIDIA GeForce GTX 1050（计算能力 6.1）上运行它。我有 CUDA v10.2，使用 Visual Studio C++ 编译器（cl.exe）。我的一个朋友也在他的 Arch Linux 系统上运行了这个，使用 CUDA v9.1 和 NVIDIA GeForce MX150（计算能力 6.1），结果相同。这是我使用的代码：

#define gpuErrchk(ans){ gpuAssert((ans), __FILE__, __LINE__);}

__host__
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
{
    if (code != cudaSuccess)
    {
        printf("GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
        if (abort) exit(code);
    }
}

class intarr {
    public:
        static const int arr_size = 5;
        int arr[arr_size];
        __host__ __device__ void print() {
            printf("arr = {");
            for(int i = 0; i < arr_size; ++i) {
                printf("%d", arr[i]);
                if(i != arr_size-1){
                    printf(", ");
                }
            }
            printf("}\n");
        }
};

class Tester {
    __device__ void print_yay() {
        printf("yay = {");
        for (int i = 0; i < yay_size; ++i) {
            if (yay[i] == 'A' || yay[i] == 'a')    printf("%c", yay[i]);
            else    printf("-");
            if (i != yay_size - 1) {
                printf(", ");
            }
        }
        printf("}\n");
    }
    public:
        intarr * b0;
        char * yay;
        const int yay_size;
        __host__ Tester() : yay_size(2) {
            intarr b;
            yay = nullptr;
            for(int i = 0; i < intarr::arr_size; ++i)    b.arr[i] = i;
            gpuErrchk(cudaMalloc(&b0, sizeof(intarr)));
            gpuErrchk(cudaMemcpy(b0, &b, sizeof(intarr), cudaMemcpyDefault));
        }

        __device__ void lol() {
            yay = (char *)malloc(2*sizeof(char));
            new(yay) char[2]{'a', 'A'};
        }

        __device__ void print() {
            b0->print();
            print_yay();
        }

        __device__ void cleanup() {
            if (yay)    free(yay);
            yay = nullptr;
        }

        /*__host__ ~Tester() {
            if(b0)        cudaFree(b0);
            b0 = nullptr;
        }*/
};

__global__ void kernel(Tester * d_t) {
    d_t->lol();
    d_t->print();
    d_t->cleanup();
}

int main() {
    printf("Tester = %zu bytes, intarr = %zu bytes\n", sizeof(Tester), sizeof(intarr));
    Tester h_t;
    Tester * d_t;
    gpuErrchk(cudaMalloc(&d_t, sizeof(Tester)));
    gpuErrchk(cudaMemcpy(d_t, &h_t, sizeof(Tester), cudaMemcpyDefault));
    kernel<<<1, 1>>>(d_t);
    gpuErrchk(cudaDeviceSynchronize());
    gpuErrchk(cudaGetLastError());
    gpuErrchk(cudaDeviceReset());
}

这里很明显有 2 个内存泄漏，因为我没有释放d_t，以及成员指针b0，使用cudaFree(). 我使用 编译这个nvcc.exe -G -Xcompiler /Zi -o cuda cuda.cu，然后运行cuda-memcheck.exe --leak-check full cuda.exe​​. 输出：

========= CUDA-MEMCHECK
Tester = 24 bytes, intarr = 20 bytes
arr = {0, 1, 2, 3, 4}
yay = {a, A}
========= LEAK SUMMARY: 0 bytes leaked in 0 allocations
========= ERROR SUMMARY: 0 errors

d_t->cleanup()当我从内核中删除调用时，输出：

========= CUDA-MEMCHECK
Tester = 24 bytes, intarr = 20 bytes
arr = {0, 1, 2, 3, 4}
yay = {a, A}
========= Leaked 2 bytes at 0x7016fff24 on the device heap
=========
========= LEAK SUMMARY: 2 bytes leaked in 1 allocations
========= ERROR SUMMARY: 1 error

2 字节泄漏很可能是因为d_t->yay没有从设备堆中释放出来。（我没有明确检查这是否是确切的泄漏，我只是在猜测）
现在，当我还删除了对 的调用时d_t->lol()，以及print_yay()from （基本上使用设备和读取的代码从设备堆中Tester::print()删除分配内存的代码），所以现在内核看起来像：d_t->yaymalloc()d_t->yay

__global__ void kernel(Tester * d_t) {
    d_t->print();
}

输出：

========= CUDA-MEMCHECK
Tester = 24 bytes, intarr = 20 bytes
arr = {0, 1, 2, 3, 4}
========= Leaked 24 bytes at 0x501200200
=========     Saved host backtrace up to driver entry point at cudaMalloc time
=========     Host Frame:C:\WINDOWS\system32\nvcuda.dll (cuMemAlloc_v2 + 0x173) [0x19d7a3]
=========     Host Frame:D:\cudatest\cuda.exe (cudart::driverHelper::mallocPtr + 0x3e) [0x4017e]
=========     Host Frame:D:\cudatest\cuda.exe (cudart::cudaApiMalloc + 0x3e) [0x1ff1e]
=========     Host Frame:D:\cudatest\cuda.exe (cudaMalloc + 0xdd) [0xc31d]
=========     Host Frame:D:\cudatest\cuda.exe (cudaMalloc<Tester> + 0x1d) [0x48e2d]
=========     Host Frame:D:\cudatest\cuda.exe (main + 0x3b) [0x4894b]
=========     Host Frame:D:\cudatest\cuda.exe (__scrt_common_main_seh + 0x10c) [0x4a1b8]
=========     Host Frame:C:\WINDOWS\System32\KERNEL32.DLL (BaseThreadInitThunk + 0x14) [0x17bd4]
=========     Host Frame:C:\WINDOWS\SYSTEM32\ntdll.dll (RtlUserThreadStart + 0x21) [0x6ce51]
=========
========= Leaked 20 bytes at 0x501200000
=========     Saved host backtrace up to driver entry point at cudaMalloc time
=========     Host Frame:C:\WINDOWS\system32\nvcuda.dll (cuMemAlloc_v2 + 0x173) [0x19d7a3]
=========     Host Frame:D:\cudatest\cuda.exe (cudart::driverHelper::mallocPtr + 0x3e) [0x4017e]
=========     Host Frame:D:\cudatest\cuda.exe (cudart::cudaApiMalloc + 0x3e) [0x1ff1e]
=========     Host Frame:D:\cudatest\cuda.exe (cudaMalloc + 0xdd) [0xc31d]
=========     Host Frame:D:\cudatest\cuda.exe (cudaMalloc<intarr> + 0x1d) [0x48e5d]
=========     Host Frame:D:\cudatest\cuda.exe (Tester::Tester + 0x6d) [0x48edd]
=========     Host Frame:D:\cudatest\cuda.exe (main + 0x2b) [0x4893b]
=========     Host Frame:D:\cudatest\cuda.exe (__scrt_common_main_seh + 0x10c) [0x4a1b8]
=========     Host Frame:C:\WINDOWS\System32\KERNEL32.DLL (BaseThreadInitThunk + 0x14) [0x17bd4]
=========     Host Frame:C:\WINDOWS\SYSTEM32\ntdll.dll (RtlUserThreadStart + 0x21) [0x6ce51]
=========
========= LEAK SUMMARY: 44 bytes leaked in 2 allocations
========= ERROR SUMMARY: 2 errors

显然，它显示了正确的泄漏。
另外，我注意到还有一些奇怪的东西。当我的内核是：

__global__ void kernel(Tester * d_t) {
    d_t->print();
    d_t->cleanup();
}
// Also print_yay() is commented out in Tester::print(), to prevent cuda-memcheck
// from terminating prematurely due to an illegal memory access error

这与上面的基本相同，因为d_t->cleanup()无论如何都不会做任何事情，输出：

========= CUDA-MEMCHECK
Tester = 24 bytes, intarr = 20 bytes
arr = {0, 1, 2, 3, 4}
========= LEAK SUMMARY: 0 bytes leaked in 0 allocations
========= ERROR SUMMARY: 0 errors

它仍然停止显示泄漏！
这是 的问题cuda-memcheck，还是我的代码有问题？

4

0 回答 0