c - CUDA - nvcc -G - if not working properly -


i'm working on porting lava flow model in cuda (full code on github here: full source of cuda-sciara fv2 lava flow model .

edit

in order reproduce issue, use instruction on github readme in order dev copy of project sciara_fv2_cuda_multicells. compile withouth -g option , passing command line argument -c ../data/2006/parameters.cfg . (quoted code @ line 260 in file )

i'm facing problem if construct inside __device__ function.

based on current lava quantity , temperature computes new temperature , if lower constant parameter (the variable d_ptsol=1143.0) lava solified.

the problem in code below is works if compile -g options (for generation of device code debug infos) behave wrongly without.

double new_temp = d_computenewtemperature(sommah,sommath);         if(new_temp <= d_ptsol){             printf("solidified %.5f,%.5f\n",new_temp,d_ptsol);             double newquote = d_sbts_updated[d_getidx(row,col,altitude)]+d_sbts_current[d_getidx(row,col,thickness)];             //code lava solidification here     }else{            //there lava , not solidified -> activate cell!            adjustadaptivegrid(row,col);  } 

ouptutting @ point of simulation:

solidified 1344.68654 1143.00000 solidified 1343.99509 1143.00000 solidified 1320.50061 1143.00000 solidified 1325.53942 1143.00000 

to make things more subtle problem completly disappear if change the if condition strict inequality if(new_temp < d_ptsol).

compilation carried out following options , in separate compilation mode

-o3 -xcompiler -fpic -std=c++11 

and linking using

--cudart static --relocatable-device-code=true -gencode arch=compute_35,code=compute_35 -gencode arch=compute_35,code=sm_35 

has faced similar issue before? doing wrong?

update

the problem seem somehow related translation of if else construct <= condition. translating

if(new_temp <= d_ptsol) {         //solidification }else{         //something else } 

to

if(new_temp <= d_ptsol) {         //solidification } if(!(new_temp <= d_ptsol)){         //something else } 

makes code work perfectly.

i had similar issue. code worked -g option not without. needed add -fmad=false , -prec-div=false compiler options (for more information see: http://docs.nvidia.com/cuda/cuda-compiler-driver-nvcc/index.html#options-for-steering-gpu-code-generation).

by default values true , optimise single precision operations, @ cost of precision. if enable debug mode optimisation disabled. needed exact values, therefore code worked -g. maybe have similar issue. (more information cuda floating point precision: http://docs.nvidia.com/cuda/cuda-c-programming-guide/#mathematical-functions-appendix)


Comments