Skip to content

Using cufft and streams #7

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Open
mnicely wants to merge 3 commits into
base: gpu_devel
Choose a base branch
from
Open

Using cufft and streams #7

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
3 commits
Select commit Hold shift + click to select a range
2fdf7aa
added streams and cufft
mnicely May 11, 2021
9a88c97
Move memory push
mnicely May 11, 2021
2d563a7
Optimized out duplicate FFT
mnicely May 12, 2021
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 1 addition & 1 deletion 2D_Program/Makefile
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -28,7 +28,7 @@ endif
ifneq ($(origin USE_CUFFTW), undefined)
ifneq ($(USE_CUFFTW), 0)
CFLAGS += -DUSE_CUFFTW
LIBS += -lcufft -lcufftw -lcudart
LIBS += -lcufft -lcudart
else
LIBS += -lfftw3_omp -lfftw3
endif
Expand Down
74 changes: 32 additions & 42 deletions 2D_Program/solver_functions/DST.c
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -8,74 +8,64 @@
#if USE_CUFFTW
#ifdef USE_COMBINE
void fullDST( const cudaStream_t *streams,
const cudaEvent_t * events,
const System sys,
const DSTN dst,
const fftw_plan plan,
const fftw_plan plan2,
const cufftHandle plan,
cuDoubleComplex * d_y,
double * in,
fftw_complex * out,
double * in2,
fftw_complex * out2 ) {
fftw_complex * out ) {

PUSH_RANGE( "forwardDST", 2 )
CUDA_RT_CALL( cudaStreamSynchronize( streams[0] ) );
load_1st_DST_wrapper( sys, dst, sys.rhs, in, in2 );
PUSH_RANGE( "1st DST", 2 )
CUDA_RT_CALL( cudaStreamWaitEvent( streams[0], events[1], cudaEventWaitDefault ) ); // Wait for sys.rhs

CUDA_RT_CALL( cudaStreamSynchronize( streams[1] ) );
fftw_execute( plan ); /********************* FFTW *********************/
fftw_execute( plan2 ); /********************* FFTW *********************/
load_1st_DST_wrapper( streams[0], sys, dst, sys.rhs, in );
CUDA_RT_CALL( cufftExecD2Z( plan, in, out ) ); // Running in streams[0]
POP_RANGE

PUSH_RANGE( "forwardDST", 3 )
CUDA_RT_CALL( cudaStreamSynchronize( streams[2] ) );
middle_stuff_ls_DST_wrapper( sys, dst, out, out2, in, in2, d_y );
PUSH_RANGE( "Trig Solver", 3 )
CUDA_RT_CALL( cudaStreamWaitEvent( streams[0], events[2], cudaEventWaitDefault ) ); // Wait forsys.U, sys.L, sys.Up

middle_stuff_ls_DST_wrapper( streams[0], sys, dst, out, in, d_y );
POP_RANGE

PUSH_RANGE( "forwardDST", 4 )
fftw_execute( plan ); /********************* FFTW *********************/
fftw_execute( plan2 ); /********************* FFTW *********************/
PUSH_RANGE( "2nd DST", 4 )
CUDA_RT_CALL( cufftExecD2Z( plan, in, out ) ); // Running in streams[0]

CUDA_RT_CALL( cudaStreamWaitEvent( streams[3], events[3], cudaEventWaitDefault ) ); // Wait for sys.sol
store_2st_DST_wrapper( streams[0], sys, dst, out, sys.sol );

CUDA_RT_CALL( cudaStreamSynchronize( streams[3] ) );
store_2st_DST_wrapper( sys, dst, out, out2, sys.sol );
CUDA_RT_CALL( cudaStreamSynchronize( streams[0] ) ); // Wait for store_2st_DST
POP_RANGE
}
#else
void fullDST( const cudaStream_t *streams,
const cudaEvent_t * events,
const System sys,
const DSTN dst,
const fftw_plan plan,
const fftw_plan plan2,
const cufftHandle plan,
cuDoubleComplex * d_rhat,
cuDoubleComplex * d_xhat,
cuDoubleComplex * d_y,
double * in,
fftw_complex * out,
double * in2,
fftw_complex * out2 ) {

PUSH_RANGE( "forwardDST", 2 )
CUDA_RT_CALL( cudaStreamSynchronize( streams[0] ) );
load_1st_DST_wrapper( sys, dst, sys.rhs, in, in2 );

CUDA_RT_CALL( cudaStreamSynchronize( streams[1] ) );
fftw_execute( plan ); /********************* FFTW *********************/
fftw_execute( plan2 ); /********************* FFTW *********************/
store_1st_DST_wrapper( sys, dst, out, out2, d_rhat );
fftw_complex * out ) {

PUSH_RANGE( "1st DST", 2 )
load_1st_DST_wrapper( NULL, sys, dst, sys.rhs, in );

CUDA_RT_CALL( cufftExecD2Z( plan, in, out ) ); // Running in streams[0]
store_1st_DST_wrapper( NULL, sys, dst, out, d_rhat );
POP_RANGE

PUSH_RANGE( "forwardDST", 3 )
CUDA_RT_CALL( cudaStreamSynchronize( streams[2] ) );
middle_stuff_DST_wrapper( sys, d_rhat, d_xhat, d_y );
PUSH_RANGE( "Trig Solver", 3 )
middle_stuff_DST_wrapper( NULL, sys, d_rhat, d_xhat, d_y );
POP_RANGE

PUSH_RANGE( "forwardDST", 4 )
CUDA_RT_CALL( cudaStreamSynchronize( streams[3] ) );
load_2st_DST_wrapper( sys, dst, d_xhat, in, in2 );
fftw_execute( plan ); /********************* FFTW *********************/
fftw_execute( plan2 ); /********************* FFTW *********************/
PUSH_RANGE( "2nd DST", 4 )
load_2st_DST_wrapper( NULL, sys, dst, d_xhat, in );

store_2st_DST_wrapper( sys, dst, out, out2, sys.sol );
CUDA_RT_CALL( cufftExecD2Z( plan, in, out ) ); // Running in streams[0]
store_2st_DST_wrapper( NULL, sys, dst, out, sys.sol );
POP_RANGE
}
#endif
Expand Down
Loading