Cuda проблемы с использованием общего буфера для симулированного распределения памяти

Question

Я преобразовываю программу из обычной программы на c в программу cuda и хотел реализовать простую оболочку для malloc, которая просто использует большой пул.

У меня 5000 тем. Мой размер блока 1024.

Вот структура буфера, которую я использую для отслеживания пула памяти каждого потока.

typedef struct buffer_t
{
    unsigned long size;
    char* current_index;
    char pool[];
} buffer_t;

Как вы можете себе представить, я использую:

cudaMalloc(&memptr, 262144*5000);

сделать распределение, где каждый поток должен создать буфер на его 262144 байта

Вот функции, которые я использую для распределения:

__device__ buffer_t* buffer_constructor(size_t size, void* memptr)
{
    buffer_t* buffer = (buffer_t*)memptr;
    buffer->size = size - sizeof(unsigned long) - sizeof(char*);
    buffer->current_index = buffer->pool;
    return buffer;
}
__device__ void* buffer_malloc(buffer_t* buffer, size_t size)
{
    if(size > buffer->size - (buffer->current_index - buffer->pool))
    {
        return NULL;
    }

    void* ptr = buffer->current_index;
    buffer->current_index += size;
    return ptr;
}

Каждый поток вызывает:

buffer_t* buffer = buffer_constructor(size, memptr+(tid * size));

Поэтому, когда я запускаю код, он просто возвращается из ядра в какой-то момент. Когда я запускаю отладчик, я получаю эту ошибку:

Program received signal CUDA_EXCEPTION_6, Warp Misaligned Address.
[Switching focus to CUDA kernel 0, grid 1, block (2,0,0), thread (768,0,0), device 0, sm 10, warp 24, lane 0]
0x0000000000b48428 in device_matrix_list_constructor (buffer=<optimized   out>, num=<optimized out>)
    at device_matrix_list.cu:8
8               return list;

Когда я запускаю memcheck, я получаю пару таких ошибок для пары блоков:

Invalid __global__ write of size 8
=========     at 0x00000258 in    /home/crafton.b/cuda_nn/device_matrix_list.cu:7:device_matrix_list_constructor(buffer_t*, unsigned int)
=========     by thread (897,0,0) in block (4,0,0)
=========     Address 0x235202a0fc is misaligned

Любая помощь очень ценится, я боролся с этим некоторое время