Не сращенное ядро ​​float2 CUDA

У меня возникли проблемы с оптимизацией размеров сетки и блоков в приведенном ниже примере. Когда я делаю профилирование, кажется, что операция записи в память в коде ядра не объединена.

Я нашел некоторые решения в интернете, но они предложили мне изменить структуру c_image на [x1, x2, x3...] [y1, y2, y3...]

Однако мне нужна структура как [x1, y1] [x2, y2]... так как я использую библиотеку CUFFT в коде где-то еще, который требует этой точной формы.

Есть ли объединенный способ выполнения операций в структуре c_image?

Мой код:

void main()
{
    float2 *c_image;  // x1 y1 x2 y2 x3 y3 x4 y4 .. .. .. .. x2048 y2048
    cudamalloc(c_image, 2048*2048*8);

    //warp size = 32
    //max thread count = 1024
    dim3 blocksize(1024, 1);
    dim3 gridsize(2048, 2);
    test<<<gridsize, blocksize>>(C_image);  
}


__global__ void test(float2 *o) 
{
    int x = threadIdx.x + blockIdx.y*1024;
    int y = blockIdx.x;

    int index = x + 2048*y;

        o[index].x = x;
        o[index].y = y;
}

Большое спасибо!

PS: я пробовал это, но не повезло! CUDA float2 объединяется

Источник

1 ответ

Решение

Сокращение этого до одного присваивания с использованием временной переменной float2 должно привести к 64-битной записи.

_global__ void test(float2 *o) 
{
    int x = threadIdx.x + blockIdx.y * 1024;
    int y = blockIdx.x;
    int index = x + 2048 * y;
    float2 tmp = float2(x, y);
    o[index] = tmp;
}

Дополнительные подробности можно найти на

Источник
Другие вопросы по тегам