He escrito algo de código para intentar cambiar los cuadrantes de la matriz 2D para FFT fines, que se almacena en un piso de la matriz.

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

Aviso que este código toma de dispositivos punteros, y no DeviceToDevice transferencias.

¿Por qué parece ser tan lento? Este puede ser optimizado de alguna manera? Yo conté esto, en comparación con la misma operación en el host utilizando regular memcpy y se acerca 2x más lento.

Alguna idea?

  • El lanzamiento de cudaMemcpy es una costosa. Es mejor escribir un kernel que lee de la entrada, swaps y se escribe en la ubicación apropiada de poner cudaMemcpy en un bucle for.
  • hrmmm..decepción. ¿Qué acerca de la comparación de de hacer un host memcpy, y transferir al dispositivo?
InformationsquelleAutor Derek | 2011-05-19

2 Comentarios

  1. 6

    Terminé de escribir un kernel para hacer el swap. Esto era de hecho más rápido que el Dispositivo a Dispositivo memcpy operaciones

  2. 3

    Quizás la siguiente solución para llevar a cabo el 2d fftshift en CUDA sería de interés:

    #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;
        }
    }

    Consiste en multiplicar la matriz a ser transformado por medio de un tablero de ajedrez de 1s y -1s que es equivalente a la multiplicación por exp(-j*(n+m)*pi) y por lo tanto a los cambios en ambas direcciones en el conjugado de dominio.

    Usted tiene que llamar a este núcleo antes y después de la aplicación de la CUFFT.

    Un pro es que la memoria de los movimientos/intercambio de evitar.

    MEJORA EN LA VELOCIDAD

    Siguiendo la sugerencia recibida en el NVIDIA Foro, la mejora de la velocidad se puede lograr cambiando la instrucción

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

    a

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

    para evitar el uso de la lentitud de la rutina de pow.

    • De hecho, en casi todas las aplicaciones de filtración de este paso puede ser removido por el mantenimiento de todos los filtros de la envuelta de la fft en el espacio.

Dejar respuesta

Please enter your comment!
Please enter your name here