CUDA выделяет память в функции __device__

Question

Есть ли способ в CUDA выделить память в функции __device__?Я не смог найти ни одного примера для этого.

Из руководства: B.15. Динамическое глобальное распределение памяти void * malloc (size_t size);void free (void * ptr);динамически выделять и освобождать память из кучи фиксированного размера в глобальной памяти.Встроенная в ядро ​​функция malloc () CUDA выделяет как минимум размер байтов из кучи устройства и возвращает указатель на выделенную память или NULL, если для выполнения запроса недостаточно памяти.Возвращаемый указатель гарантированно будет выровнен по 16-байтовой границе.Встроенная в ядро ​​функция free () CUDA освобождает память, на которую указывает ptr, которая должна была быть возвращена предыдущим вызовом malloc ().Если ptr равен NULL, вызов free () игнорируется.Повторные вызовы free () с одним и тем же ptr имеют неопределенное поведение.Память, выделенная заданным потоком CUDA через malloc (), остается выделенной на время существования контекста CUDA или до тех пор, пока она не будет явно освобождена вызовом free ().Он может использоваться любыми другими потоками CUDA даже после последующих запусков ядра.Любой поток CUDA может освободить память, выделенную другим потоком, но следует позаботиться о том, чтобы один и тот же указатель не был освобожден более одного раза.

Answer 1

Согласно http://developer.download.nvidia.com/compute/cuda/3_2_prod/toolkit/docs/CUDA_C_Programming_Guide.pdf вы должны иметь возможность использовать malloc () и free () в функции устройства.

Page 122

B.15 Динамическое глобальное распределение памяти void * malloc (size_t size);void free (void * 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 архитектуру.