From 72e0e4807db04a61fab24588897e5ff4fffb620a Mon Sep 17 00:00:00 2001 From: "Ryan M. Bergmann" Date: Fri, 4 Mar 2016 13:12:21 +0100 Subject: [PATCH] checking popo routine... --- pop_fission.cu | 14 +++++++------- whistory.cpp | 4 +++- 2 files changed, 10 insertions(+), 8 deletions(-) diff --git a/pop_fission.cu b/pop_fission.cu index 18da278..fbb2c57 100644 --- a/pop_fission.cu +++ b/pop_fission.cu @@ -304,18 +304,18 @@ __global__ void pop_fission_kernel(unsigned N, cross_section_data* d_xsdata, par } // check temp array - if(fission_energy[ data_dex] != 0.0) {printf("NONZERO fission_energy[ data_dex] = % 6.4E\n",fission_energy[ data_dex] );} - if(fission_particles[data_dex].x != 0.0) {printf("NONZERO fission_particles[data_dex].x = % 6.4E \n",fission_particles[data_dex].x );} - if(fission_particles[data_dex].y != 0.0) {printf("NONZERO fission_particles[data_dex].y = % 6.4E \n",fission_particles[data_dex].y );} - if(fission_particles[data_dex].z != 0.0) {printf("NONZERO fission_particles[data_dex].z = % 6.4E \n",fission_particles[data_dex].z );} + if(fission_energy[ data_dex] != 0.0) {printf("NONZERO fission_energy[ data_dex] = % 6.4E \n",fission_energy[ data_dex] );} + if(fission_particles[data_dex].x != 0.0) {printf("NONZERO fission_particles[data_dex].x = % 6.4E \n",fission_particles[data_dex].x );} + if(fission_particles[data_dex].y != 0.0) {printf("NONZERO fission_particles[data_dex].y = % 6.4E \n",fission_particles[data_dex].y );} + if(fission_particles[data_dex].z != 0.0) {printf("NONZERO fission_particles[data_dex].z = % 6.4E \n",fission_particles[data_dex].z );} if(fission_particles[data_dex].xhat != 0.0) {printf("NONZERO fission_particles[data_dex].xhat = % 6.4E \n",fission_particles[data_dex].xhat );} if(fission_particles[data_dex].yhat != 0.0) {printf("NONZERO fission_particles[data_dex].yhat = % 6.4E \n",fission_particles[data_dex].yhat );} if(fission_particles[data_dex].zhat != 0.0) {printf("NONZERO fission_particles[data_dex].zhat = % 6.4E \n",fission_particles[data_dex].zhat );} if(fission_particles[data_dex].surf_dist != 10000.0){printf("NONZERO fission_particles[data_dex].surf_dist = % 6.4E \n",fission_particles[data_dex].surf_dist );} if(fission_particles[data_dex].enforce_BC != 0.0) {printf("NONZERO fission_particles[data_dex].enforce_BC = % 6.4E \n",fission_particles[data_dex].enforce_BC );} - if(fission_particles[data_dex].norm[0] != 1.0) {printf("NONZERO fission_particles[data_dex].norm[0] = % 6.4E \n",fission_particles[data_dex].norm[0] );} - if(fission_particles[data_dex].norm[1] != 0.0) {printf("NONZERO fission_particles[data_dex].norm[1] = % 6.4E \n",fission_particles[data_dex].norm[1] );} - if(fission_particles[data_dex].norm[2] != 0.0) {printf("NONZERO fission_particles[data_dex].norm[2] = % 6.4E \n",fission_particles[data_dex].norm[2] );} + if(fission_particles[data_dex].norm[0] != 1.0) {printf("NONZERO fission_particles[data_dex].norm[0] = % 6.4E \n",fission_particles[data_dex].norm[0] );} + if(fission_particles[data_dex].norm[1] != 0.0) {printf("NONZERO fission_particles[data_dex].norm[1] = % 6.4E \n",fission_particles[data_dex].norm[1] );} + if(fission_particles[data_dex].norm[2] != 0.0) {printf("NONZERO fission_particles[data_dex].norm[2] = % 6.4E \n",fission_particles[data_dex].norm[2] );} // set data in temp array since GRID-WISE threadsync cannot be done (easily?)! fission_energy[ data_dex ] = sampled_E; diff --git a/whistory.cpp b/whistory.cpp index 4c82ce6..d0e3c67 100644 --- a/whistory.cpp +++ b/whistory.cpp @@ -299,11 +299,13 @@ void whistory::init_device(){ // copy values from initialized host arrays check_cuda(cudaMemcpy( dh_particles.space , h_particles.space , Ndataset*sizeof(spatial_data) , cudaMemcpyHostToDevice )); + check_cuda(cudaMemcpy( d_fissile_points , h_particles.space , Ndataset*sizeof(spatial_data) , cudaMemcpyHostToDevice )); check_cuda(cudaMemcpy( dh_particles.cellnum , h_particles.cellnum , Ndataset*sizeof(unsigned) , cudaMemcpyHostToDevice )); check_cuda(cudaMemcpy( dh_particles.talnum , h_particles.talnum , Ndataset*sizeof(int) , cudaMemcpyHostToDevice )); check_cuda(cudaMemcpy( dh_particles.matnum , h_particles.matnum , Ndataset*sizeof(unsigned) , cudaMemcpyHostToDevice )); check_cuda(cudaMemcpy( dh_particles.rxn , h_particles.rxn , Ndataset*sizeof(unsigned) , cudaMemcpyHostToDevice )); check_cuda(cudaMemcpy( dh_particles.E , h_particles.E , Ndataset*sizeof(float) , cudaMemcpyHostToDevice )); + check_cuda(cudaMemcpy( d_fissile_energy , zeros , Ndataset*sizeof(float) , cudaMemcpyHostToDevice )); check_cuda(cudaMemcpy( dh_particles.Q , h_particles.Q , Ndataset*sizeof(float) , cudaMemcpyHostToDevice )); check_cuda(cudaMemcpy( dh_particles.rn_bank , h_particles.rn_bank , Ndataset*sizeof(float) , cudaMemcpyHostToDevice )); check_cuda(cudaMemcpy( dh_particles.isonum , h_particles.isonum , Ndataset*sizeof(unsigned) , cudaMemcpyHostToDevice )); @@ -1850,7 +1852,7 @@ void whistory::reset_cycle(float keff_cycle){ check_cuda(cudaPeekAtLastError()); //pop them in! should be the right size now due to keff rebasing - null_spatial(NUM_THREADS,Ndataset,dh_particles.space); + null_spatial(NUM_THREADS,Ndataset,d_fissile_points); check_cuda(cudaMemcpy( dh_particles.E, d_zeros, Ndataset*sizeof(float), cudaMemcpyDeviceToDevice )); check_cuda(cudaThreadSynchronize()); check_cuda(cudaDeviceSynchronize());