CUDA Appareil À transférer cher

J'ai écrit un code pour essayer de permuter les quadrants d'une matrice 2D FFT fins, qui sont stockées dans un tableau plat.

    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);

Notez que ce code prend dispositif de pointeurs, et ne DeviceToDevice transferts.

Pourquoi cela vous semble courir si lent? Cela peut-il être optimisé en quelque sorte? J'ai programmé ce, comparativement à la même opération sur l'hôte en utilisant un memcpy et il était environ 2x plus lent.

Des idées?

Le lancement de cudaMemcpy est coûteuse. Vous êtes mieux de l'écriture d'un noyau qui lit à partir de l'entrée, de swaps et de l'écrit à l'emplacement approprié que de mettre cudaMemcpy dans une boucle for.
hrmmm..la poisse. Que dire de la comparaison de procéder à un hôte memcpy, et de transférer à l'appareil?

OriginalL'auteur Derek | 2011-05-19

c++cuda fft

6

J'ai fini par écrire un noyau à faire de swaps. C'était en effet plus rapide que l'Appareil à memcpy opérations

OriginalL'auteur Derek
3

Peut-être la solution suivante pour effectuer la 2d fftshift dans CUDA serait d'intérêt:
```
#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;
    }
}
```
Il consiste à multiplier la matrice à être transformé par un échiquier de 1s et -1s qui est l'équivalent de la multiplication par exp(-j*(n+m)*pi) et donc à des changements dans les deux directions dans le conjugué de domaine.

Vous devez appeler ce noyau avant et après l'application de la CUFFT.

Un pro, c'est que la mémoire des mouvements/d'échange sont à éviter.

AMÉLIORATION DE LA VITESSE

À la suite de la suggestion à la NVIDIA Forum, l'amélioration de la vitesse peut être réalisé que par l'évolution de l'instruction
```
double a = pow(-1.0,(i+j)&1);
```
à
```
double a = 1-2*((i+j)&1);
```
d'éviter l'utilisation de la lenteur de la routine pow.

En effet, dans presque tous les filtrages des applications de cette étape peut être supprimée par le maintien de tous les filtres de la enveloppé fft espace.

OriginalL'auteur JackOLantern

Vous devez vous connecter pour publier un commentaire.