14

CUDA 有没有办法在设备端函数中动态分配内存？我找不到任何这样做的例子。

从 CUDA C 编程手册：

B.15 动态全局内存分配

void* malloc(size_t size); 
void free(void* ptr);

从全局内存中的固定大小的堆中动态分配和释放内存。

CUDA 内核malloc()函数从设备堆中分配至少 size 个字节，并返回一个指向已分配内存的指针，如果没有足够的内存来满足请求，则返回 NULL。返回的指针保证与 16 字节边界对齐。

CUDA 内核free()函数释放 指向的内存ptr，该内存必须由先前的调用返回malloc()。如果ptr是NULL，则忽略对 free() 的调用。使用相同的 ptr 重复调用 free() 具有未定义的行为。

给定 CUDA 线程通过分配的内存malloc()在 CUDA 上下文的生命周期内保持分配，或者直到它通过调用显式释放free()。它可以被任何其他 CUDA 线程使用，即使在随后的内核启动时也是如此。任何 CUDA 线程都可以释放由另一个线程分配的内存，但应注意确保不会多次释放同一指针。

4

1 回答 1

21

根据http://developer.download.nvidia.com/compute/cuda/3_2_prod/toolkit/docs/CUDA_C_Programming_Guide.pdf您应该能够在设备功能中使用 malloc() 和 free() 。

第 122 页

B.15 动态全局内存分配 void* malloc(size_t size); 无效自由（无效* ptr）；从全局内存中的固定大小的堆中动态分配和释放内存。

手册中给出的示例。

__global__ void mallocTest()
{
    char* ptr = (char*)malloc(123);
    printf(“Thread %d got pointer: %p\n”, threadIdx.x, ptr);
    free(ptr);
}

void main()
{
    // Set a heap size of 128 megabytes. Note that this must
    // be done before any kernel is launched.
    cudaThreadSetLimit(cudaLimitMallocHeapSize, 128*1024*1024);
    mallocTest<<<1, 5>>>();
    cudaThreadSynchronize();
}

您需要编译器参数 -arch=sm_20 和支持 >2x 架构的卡。

于 2011-03-09T17:08:15.947 回答