CUDA устройство на устройство передачи дорого - PullRequest
Новая
       14

CUDA устройство на устройство передачи дорого

5 голосов
/ 19 мая 2011

Я написал некоторый код, чтобы попытаться поменять квадранты 2D-матрицы для целей БПФ, которая хранится в плоском массиве. 

    int leftover = W-dcW;

    T *temp;
    T *topHalf;
cudaMalloc((void **)&temp, dcW * sizeof(T));

    //swap every row, left and right
    for(int i = 0; i < H; i++)
    {
        cudaMemcpy(temp, &data[i*W], dcW*sizeof(T),cudaMemcpyDeviceToDevice);
        cudaMemcpy(&data[i*W],&data[i*W+dcW], leftover*sizeof(T), cudaMemcpyDeviceToDevice);
        cudaMemcpy(&data[i*W+leftover], temp, dcW*sizeof(T), cudaMemcpyDeviceToDevice); 
    }

cudaMalloc((void **)&topHalf, dcH*W* sizeof(T));
    leftover = H-dcH;
    cudaMemcpy(topHalf, data, dcH*W*sizeof(T), cudaMemcpyDeviceToDevice);
    cudaMemcpy(data, &data[dcH*W], leftover*W*sizeof(T), cudaMemcpyDeviceToDevice);
    cudaMemcpy(&data[leftover*W], topHalf, dcH*W*sizeof(T), cudaMemcpyDeviceToDevice);

Обратите внимание, что этот код принимает указатели на устройства и передает DeviceToDevice.

Почему это работает так медленно? Можно ли это как-то оптимизировать? Я рассчитал это по сравнению с той же операцией на хосте, используя обычный memcpy, и это было примерно в 2 раза медленнее.

Есть идеи?

Ответы [ 2 ]

6 голосов
/ 27 мая 2011

Я закончил писать ядро, чтобы сделать перестановки. Это действительно было быстрее, чем операции memcpy между устройствами

3 голосов
/ 08 января 2013

Возможно, будет интересно следующее решение для выполнения 2d fftshift в CUDA: 

#define IDX2R(i,j,N) (((i)*(N))+(j))

__global__ void fftshift_2D(double2 *data, int N1, int N2)
{
    int i = threadIdx.y + blockDim.y * blockIdx.y;
    int j = threadIdx.x + blockDim.x * blockIdx.x;

    if (i < N1 && j < N2) {
        double a = pow(-1.0, (i+j)&1);

        data[IDX2R(i,j,N2)].x *= a;
        data[IDX2R(i,j,N2)].y *= a;
    }
}

Оно состоит в умножении матрицы, которая будет преобразована на шахматную доску 1 s и -1 s.что эквивалентно умножению на exp(-j*(n+m)*pi) и, следовательно, сдвигам в обоих направлениях в сопряженной области.

Вы должны вызывать это ядро ​​до и после применения CUFFT.

Одним из преимуществ является предотвращение движений / подкачки памяти.

УЛУЧШЕНИЕ В СКОРОСТИ

Следуя предложению, полученному на форуме NVIDIA , можно улучшить скорость, изменив инструкцию 

double a = pow(-1.0,(i+j)&1);

на 

double a = 1-2*((i+j)&1);

, чтобы избежать использования медленной рутины.

...