moebiusband/MD-Bench
1
0
Fork 0
Code Issues Pull Requests Packages Projects Releases Wiki Activity

🐛 Fixed aforementioned correctness issue by deleting a superflous cudaMemcpy in computeForce() that was overwriting correct data with incorrect data

Browse Source
This commit is contained in:
Martin Bauernfeind 2022-07-05 00:54:11 +02:00
parent 4f0403d3ea
commit 71798f5ec5
3 changed files with 4 additions and 6 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

5
src/force.cu
View File

@ -193,11 +193,6 @@ double computeForce(
cudaProfilerStart(); cudaProfilerStart();
if(!reneighbourHappenend) {
checkCUDAError( "computeForce c_atom->x memcpy", cudaMemcpy(c_atom->x, atom->x, sizeof(MD_FLOAT) * atom->Nmax * 3, cudaMemcpyHostToDevice) );
}
const int num_blocks = ceil((float)Nlocal / (float)num_threads_per_block); const int num_blocks = ceil((float)Nlocal / (float)num_threads_per_block);
double S = getTimeStamp(); double S = getTimeStamp();

1
src/main.c
View File

@ -363,6 +363,7 @@ int main(int argc, char** argv)
cuda_final_integrate(doReneighbour, &param, &atom, &c_atom, num_threads_per_block); cuda_final_integrate(doReneighbour, &param, &atom, &c_atom, num_threads_per_block);
if(!((n + 1) % param.nstat) && (n+1) < param.ntimes) { if(!((n + 1) % param.nstat) && (n+1) < param.ntimes) {
checkCUDAError("computeThermo atom->x memcpy back", cudaMemcpy(atom.x, c_atom.x, atom.Nmax * sizeof(MD_FLOAT) * 3, cudaMemcpyDeviceToHost) );
computeThermo(n + 1, &param, &atom); computeThermo(n + 1, &param, &atom);
} }

2
src/pbc.cu
View File

@ -131,8 +131,10 @@ void updatePbc_cuda(Atom *atom, Parameter *param, Atom *c_atom, bool doReneighbo
* MD_FLOAT xprd, MD_FLOAT yprd, MD_FLOAT zprd) * MD_FLOAT xprd, MD_FLOAT yprd, MD_FLOAT zprd)
* */ * */
computePbcUpdate<<<num_blocks, num_threads_per_block>>>(*c_atom, c_PBCx, c_PBCy, c_PBCz, xprd, yprd, zprd); computePbcUpdate<<<num_blocks, num_threads_per_block>>>(*c_atom, c_PBCx, c_PBCy, c_PBCz, xprd, yprd, zprd);
if(doReneighbor){
checkCUDAError( "updatePbc atom->x memcpy back", cudaMemcpy(atom->x, c_atom->x, atom->Nmax * sizeof(MD_FLOAT) * 3, cudaMemcpyDeviceToHost) ); checkCUDAError( "updatePbc atom->x memcpy back", cudaMemcpy(atom->x, c_atom->x, atom->Nmax * sizeof(MD_FLOAT) * 3, cudaMemcpyDeviceToHost) );
} }
}
/* relocate atoms that have left domain according /* relocate atoms that have left domain according
* to periodic boundary conditions */ * to periodic boundary conditions */