cuda_xmipp_utils.cpp File Reference

#include "cuda_xmipp_utils.h"
#include "cuda_asserts.h"
#include <cuda_runtime.h>
#include <cufft.h>
#include <cufftXt.h>
#include <cuComplex.h>
#include <nvml.h>
#include <time.h>
#include <sys/time.h>

Include dependency graph for cuda_xmipp_utils.cpp:

This graph shows which files directly or indirectly include this file:

Go to the source code of this file.

Classes
struct	pointwiseMult

Functions
void	myStreamDestroy (void *ptr)
void	myStreamCreate (myStreamHandle &myStream)
void	mycufftDestroy (void *ptr)
void	calculateFFTPlanSize (mycufftHandle &myhandle)
void	createPlanFFT (int Xdim, int Ydim, int Ndim, int Zdim, bool forward, cufftHandle *plan)
void	createPlanFFTStream (int Xdim, int Ydim, int Ndim, int Zdim, bool forward, cufftHandle *plan, myStreamHandle &myStream)
void	gpuMalloc (void **d_data, size_t Nbytes)
void	gpuFree (void *d_data)
void	cpuMalloc (void **h_data, size_t Nbytes)
void	cpuFree (void *h_data)
void	initializeIdentity (float *d_data, float *h_data, int Ndim, myStreamHandle &myStream)
void	setTranslationMatrix (float *d_data, float *posX, float *posY, int Ndim, myStreamHandle &myStream)
void	setRotationMatrix (float *d_data, float *ang, int Ndim, myStreamHandle &myStream)
void	gpuCopyFromCPUToGPU (void *data, void *d_data, size_t Nbytes)
void	gpuCopyFromGPUToCPU (void *d_data, void *data, size_t Nbytes)
void	gpuCopyFromGPUToGPU (void *d_dataFrom, void *d_dataTo, size_t Nbytes)
void	gpuCopyFromCPUToGPUStream (void *data, void *d_data, size_t Nbytes, myStreamHandle &myStream)
void	gpuCopyFromGPUToCPUStream (void *d_data, void *data, size_t Nbytes, myStreamHandle &myStream)
void	gpuCopyFromGPUToGPUStream (void *d_dataFrom, void *d_dataTo, size_t Nbytes, myStreamHandle &myStream)
int	gridFromBlock (int tasks, int Nthreads)
void	cuda_check_gpu_memory (float *data)
void	cuda_check_gpu_properties (int *grid)
__device__ cufftComplex	CB_pointwiseMultiplicationComplexKernelLoad (void *dataIn, size_t offset, void *callerInfo, void *sharedPtr)
__device__ void	CB_pointwiseMultiplicationComplexKernelStore (void *dataOut, size_t offset, cufftComplex element, void *callerInfo, void *sharedPtr)
template float *	loadToGPU< float > (const float *data, size_t items)
template std::complex< float > *	loadToGPU< std::complex< float > > (const std::complex< float > *data, size_t items)
template<typename T >
T *	loadToGPU (const T *data, size_t items)

Variables
__device__ cufftCallbackLoadC	d_pointwiseMultiplicationComplexKernelLoad = CB_pointwiseMultiplicationComplexKernelLoad
__device__ cufftCallbackStoreC	d_pointwiseMultiplicationComplexKernelStore = CB_pointwiseMultiplicationComplexKernelStore

Function Documentation

calculateFFTPlanSize()

void calculateFFTPlanSize

(

mycufftHandle &

myhandle

)

Definition at line 79 of file cuda_xmipp_utils.cpp.

                                                  {
    printf("calculateFFTPlanSize  myhandle.ptr: %p\n",myhandle.ptr);
    size_t ws2;
    cufftHandle *planFptr=(cufftHandle *)myhandle.ptr;
    cufftGetSize(*planFptr, &ws2);
    printf("calculateFFTPlanSize  size %i\n", (int)ws2);
}

CB_pointwiseMultiplicationComplexKernelLoad()

__device__ cufftComplex CB_pointwiseMultiplicationComplexKernelLoad	(	void *	dataIn,
		size_t	offset,
		void *	callerInfo,
		void *	sharedPtr
	)

Definition at line 347 of file cuda_xmipp_utils.cpp.

{

    //printf("INSIDEEEE IFFT\n");
    pointwiseMult *myData = (pointwiseMult*)callerInfo;

    cufftComplex reference = ((cufftComplex*)dataIn)[offset];
    cufftComplex *mask = (cufftComplex*)myData->data;

    int normFactor = myData->normFactor;
    int indexM = offset%normFactor;

    float factor = 1.0f / normFactor;

    cufftComplex mulOut = cuCmulf((cuComplex)reference, (cuComplex)mask[indexM]);
    cufftComplex out;
    out.x = mulOut.x*factor;
    out.y = mulOut.y*factor;

    //if(offset>9000 && offset<9100)
    //  printf("offset %i, mask %f, data %f, mul %f, factor %f\n", offset, mask[indexM].x, reference.x, out.x, factor);

    return out;
}

CB_pointwiseMultiplicationComplexKernelStore()

__device__ void CB_pointwiseMultiplicationComplexKernelStore	(	void *	dataOut,
		size_t	offset,
		cufftComplex	element,
		void *	callerInfo,
		void *	sharedPtr
	)

Definition at line 376 of file cuda_xmipp_utils.cpp.

{

    pointwiseMult *myData = (pointwiseMult*)callerInfo;

    cufftComplex *mask = myData->data;
    int normFactor = myData->normFactor;
    int indexM = offset%normFactor;

    float factor = 1.0f / normFactor;

    cufftComplex mulOut = cuCmulf((cuComplex)element, (cuComplex)mask[indexM]);
    cufftComplex out;
    out.x = mulOut.x*factor;
    out.y = mulOut.y*factor;
    ((cufftComplex*)dataOut)[offset] = out;

}

cpuFree()

void cpuFree

(

void *

h_data

)

Definition at line 211 of file cuda_xmipp_utils.cpp.

{
    gpuErrchk(cudaFreeHost(h_data));
}

cpuMalloc()

void cpuMalloc	(	void **	h_data,
		size_t	Nbytes
	)

Definition at line 206 of file cuda_xmipp_utils.cpp.

{
    gpuErrchk(cudaMallocHost(h_data, Nbytes));
}

createPlanFFT()

void createPlanFFT	(	int	Xdim,
		int	Ydim,
		int	Ndim,
		int	Zdim,
		bool	forward,
		cufftHandle *	plan
	)

Definition at line 88 of file cuda_xmipp_utils.cpp.

                                                                                           {

    int Xfdim=(Xdim/2)+1;

    int nr1[] = {Xdim};   // --- Size of the image in real space
    int nr2[] = {Ydim, Xdim};   // --- Size of the image in real space
    int nr3[] = {Zdim, Ydim, Xdim};   // --- Size of the image in real space

    int nf1[] = {Xfdim};   // --- Size of the Fourier transform
    int nf2[] = {Ydim, Xfdim};   // --- Size of the Fourier transform
    int nf3[] = {Zdim, Ydim, Xfdim};   // --- Size of the Fourier transform
    int *nr=NULL, *nf=NULL;
    int NRANK; // 1D, 2D or 3D FFTs
    if (Ydim==1 && Zdim==1)
    {
        NRANK=1;
        nr=nr1;
        nf=nf1;
    }
    else if (Zdim==1)
    {
        NRANK=2;
        nr=nr2;
        nf=nf2;
    }
    else
    {
        NRANK=3;
        nr=nr3;
        nf=nf3;
    }

    int rstride = 1;                // --- Distance between two successive input/output elements
    int fstride = 1;
    int rdist = Xdim*Ydim*Zdim;     // --- Distance between batches
    int fdist = Xfdim*Ydim*Zdim;

    if(forward){
        gpuErrchkFFT(cufftPlanMany(plan, NRANK, nr, nr, rstride, rdist, nf, fstride, fdist, CUFFT_R2C, Ndim));
    }else{
        gpuErrchkFFT(cufftPlanMany(plan, NRANK, nr, nf, fstride, fdist, nr, rstride, rdist, CUFFT_C2R, Ndim));
    }

}

createPlanFFTStream()

void createPlanFFTStream	(	int	Xdim,
		int	Ydim,
		int	Ndim,
		int	Zdim,
		bool	forward,
		cufftHandle *	plan,
		myStreamHandle &	myStream
	)

Definition at line 133 of file cuda_xmipp_utils.cpp.

                                                                  {

    int Xfdim=(Xdim/2)+1;

    int nr1[] = {Xdim};   // --- Size of the image in real space
    int nr2[] = {Ydim, Xdim};   // --- Size of the image in real space
    int nr3[] = {Zdim, Ydim, Xdim};   // --- Size of the image in real space

    int nf1[] = {Xfdim};   // --- Size of the Fourier transform
    int nf2[] = {Ydim, Xfdim};   // --- Size of the Fourier transform
    int nf3[] = {Zdim, Ydim, Xfdim};   // --- Size of the Fourier transform
    int *nr=NULL, *nf=NULL;
    int NRANK; // 1D, 2D or 3D FFTs
    if (Ydim==1 && Zdim==1)
    {
        NRANK=1;
        nr=nr1;
        nf=nf1;
    }
    else if (Zdim==1)
    {
        NRANK=2;
        nr=nr2;
        nf=nf2;
    }
    else
    {
        NRANK=3;
        nr=nr3;
        nf=nf3;
    }

    int rstride = 1;                // --- Distance between two successive input/output elements
    int fstride = 1;
    int rdist = Xdim*Ydim*Zdim;     // --- Distance between batches
    int fdist = Xfdim*Ydim*Zdim;

    cudaStream_t *stream = (cudaStream_t*) myStream.ptr;
    if(forward){
        gpuErrchkFFT(cufftPlanMany(plan, NRANK, nr, nr, rstride, rdist, nf, fstride, fdist, CUFFT_R2C, Ndim));
        gpuErrchkFFT(cufftSetStream(*plan, *stream));
    }else{
        gpuErrchkFFT(cufftPlanMany(plan, NRANK, nr, nf, fstride, fdist, nr, rstride, rdist, CUFFT_C2R, Ndim));
        gpuErrchkFFT(cufftSetStream(*plan, *stream));
    }

}

gpuCopyFromCPUToGPU()

void gpuCopyFromCPUToGPU	(	void *	data,
		void *	d_data,
		size_t	Nbytes
	)

Definition at line 278 of file cuda_xmipp_utils.cpp.

{
    gpuErrchk(cudaMemcpy(d_data, data, Nbytes, cudaMemcpyHostToDevice));
}

gpuCopyFromCPUToGPUStream()

void gpuCopyFromCPUToGPUStream	(	void *	data,
		void *	d_data,
		size_t	Nbytes,
		myStreamHandle &	myStream
	)

Definition at line 293 of file cuda_xmipp_utils.cpp.

{
    cudaStream_t *stream = (cudaStream_t*) myStream.ptr;
    gpuErrchk(cudaMemcpyAsync(d_data, data, Nbytes, cudaMemcpyHostToDevice, *stream));

    //gpuErrchk(cudaStreamSynchronize(*stream));
}

gpuCopyFromGPUToCPU()

void gpuCopyFromGPUToCPU	(	void *	d_data,
		void *	data,
		size_t	Nbytes
	)

Definition at line 283 of file cuda_xmipp_utils.cpp.

{
    gpuErrchk(cudaMemcpy(data, d_data, Nbytes, cudaMemcpyDeviceToHost));
}

gpuCopyFromGPUToCPUStream()

void gpuCopyFromGPUToCPUStream	(	void *	d_data,
		void *	data,
		size_t	Nbytes,
		myStreamHandle &	myStream
	)

Definition at line 301 of file cuda_xmipp_utils.cpp.

{
    cudaStream_t *stream = (cudaStream_t*) myStream.ptr;
    gpuErrchk(cudaMemcpyAsync(data, d_data, Nbytes, cudaMemcpyDeviceToHost, *stream));

    gpuErrchk(cudaStreamSynchronize(*stream));
    //cudaDeviceSynchronize();
}

gpuCopyFromGPUToGPU()

void gpuCopyFromGPUToGPU	(	void *	d_dataFrom,
		void *	d_dataTo,
		size_t	Nbytes
	)

Definition at line 288 of file cuda_xmipp_utils.cpp.

{
    gpuErrchk(cudaMemcpy(d_dataTo, d_dataFrom, Nbytes, cudaMemcpyDeviceToDevice));
}

gpuCopyFromGPUToGPUStream()

void gpuCopyFromGPUToGPUStream	(	void *	d_dataFrom,
		void *	d_dataTo,
		size_t	Nbytes,
		myStreamHandle &	myStream
	)

Definition at line 310 of file cuda_xmipp_utils.cpp.

{
    cudaStream_t *stream = (cudaStream_t*) myStream.ptr;
    gpuErrchk(cudaMemcpyAsync(d_dataTo, d_dataFrom, Nbytes, cudaMemcpyDeviceToDevice, *stream));
}

gpuFree()

void gpuFree

(

void *

d_data

)

Definition at line 201 of file cuda_xmipp_utils.cpp.

{
    gpuErrchk(cudaFree(d_data));
}

gpuMalloc()

void gpuMalloc	(	void **	d_data,
		size_t	Nbytes
	)

Definition at line 196 of file cuda_xmipp_utils.cpp.

{
    gpuErrchk(cudaMalloc(d_data, Nbytes));
}

gridFromBlock()

int gridFromBlock	(	int	tasks,
		int	Nthreads
	)

Definition at line 316 of file cuda_xmipp_utils.cpp.

{
    int numBlk = tasks/Nthreads;
    if(tasks%Nthreads>0)
        numBlk++;
    return numBlk;
}

initializeIdentity()

void initializeIdentity	(	float *	d_data,
		float *	h_data,
		int	Ndim,
		myStreamHandle &	myStream
	)

Definition at line 216 of file cuda_xmipp_utils.cpp.

{
    //float *matrices = new float[Ndim*9];
    for(int i=0; i<Ndim; i++){
        float aux_matrix[9] = {1.0, 0.0, 0.0, 0.0, 1.0, 0.0, 0.0, 0.0, 1.0};
        int offset=i*9;
        for (int j=0; j<9; j++)
            h_data[offset+j] = aux_matrix[j];
    }
    cudaStream_t *stream = (cudaStream_t*) myStream.ptr;
    gpuErrchk(cudaMemcpyAsync((void*)d_data, h_data, Ndim*9*sizeof(float), cudaMemcpyHostToDevice, *stream));
    //delete []matrices;

}

loadToGPU< float >()

template float* loadToGPU< float >	(	const float *	data,
		size_t	items
	)

loadToGPU< std::complex< float > >()

template std::complex<float>* loadToGPU< std::complex< float > >	(	const std::complex< float > *	data,
		size_t	items
	)

mycufftDestroy()

void mycufftDestroy

(

void *

ptr

)

Definition at line 72 of file cuda_xmipp_utils.cpp.

{
    cufftHandle *planPtr = (cufftHandle *)ptr;
    cufftDestroy(*planPtr);
    delete planPtr;
}

myStreamCreate()

void myStreamCreate

(

myStreamHandle &

myStream

)

Definition at line 63 of file cuda_xmipp_utils.cpp.

{
    cudaStream_t *streamPtr = new cudaStream_t;
    gpuErrchk(cudaStreamCreate(streamPtr));
    myStream.ptr = (void*)streamPtr;
    //printf("ptr %p\n", myStream.ptr);
    //printf("streamPtr %p\n", streamPtr);
}

myStreamDestroy()

void myStreamDestroy

(

void *

ptr

)

Definition at line 57 of file cuda_xmipp_utils.cpp.

{
    cudaStream_t *streamPtr = (cudaStream_t *)ptr;
    cudaStreamDestroy(*streamPtr);
}

setRotationMatrix()

void setRotationMatrix	(	float *	d_data,
		float *	ang,
		int	Ndim,
		myStreamHandle &	myStream
	)

Definition at line 260 of file cuda_xmipp_utils.cpp.

{

    float *rad_vector;
    gpuErrchk(cudaMallocHost((void**)&rad_vector, sizeof(float)*Ndim*9));

    for(int i=0; i<Ndim; i++){
        float rad = (float)(-ang[i]*M_PI/180);
        float matrix[9] = {cosf(rad), -sinf(rad), 0, sinf(rad), cosf(rad), 0, 0, 0, 1};
        int offset=i*9;
        for (int j=0; j<9; j++)
            rad_vector[offset+j] = matrix[j];
    }
    cudaStream_t *stream = (cudaStream_t*) myStream.ptr;
    gpuErrchk(cudaMemcpyAsync((void*)d_data, rad_vector, Ndim*9*sizeof(float), cudaMemcpyHostToDevice, *stream));
    delete []rad_vector;
}

setTranslationMatrix()

void setTranslationMatrix	(	float *	d_data,
		float *	posX,
		float *	posY,
		int	Ndim,
		myStreamHandle &	myStream
	)

Definition at line 237 of file cuda_xmipp_utils.cpp.

{
    float *matrices;
    gpuErrchk(cudaMallocHost((void**)&matrices, sizeof(float)*Ndim*9));

    for(int i=0; i<Ndim; i++){
        float aux_matrix[9] = {1, 0, -posX[i], 0, 1, -posY[i], 0, 0, 1};
        int offset=i*9;
        //memcpy(&matrices[offset], &aux_matrix, 9*sizeof(float));
        for (int j=0; j<9; j++)
            matrices[offset+j] = aux_matrix[j];
    }
    cudaStream_t *stream = (cudaStream_t*) myStream.ptr;
    gpuErrchk(cudaMemcpyAsync((void*)d_data, matrices, Ndim*9*sizeof(float), cudaMemcpyHostToDevice, *stream));
    delete []matrices;
}

Variable Documentation

d_pointwiseMultiplicationComplexKernelLoad

__device__ cufftCallbackLoadC d_pointwiseMultiplicationComplexKernelLoad = CB_pointwiseMultiplicationComplexKernelLoad

Definition at line 372 of file cuda_xmipp_utils.cpp.

d_pointwiseMultiplicationComplexKernelStore

__device__ cufftCallbackStoreC d_pointwiseMultiplicationComplexKernelStore = CB_pointwiseMultiplicationComplexKernelStore

Definition at line 395 of file cuda_xmipp_utils.cpp.