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

🐛 Fixed some bugs - ✨ neighborhood computation now seems to be quite fast

Browse Source
This commit is contained in:
Martin Bauernfeind 2022-06-26 20:19:59 +02:00
parent 16e8b76012
commit 4a32a62a98
1 changed files with 5 additions and 2 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

7
src/neighbor.cu
View File

@ -585,7 +585,7 @@ void buildNeighbor_cuda(Atom *atom, Neighbor *neighbor, Atom *c_atom, Neighbor *
while(resize) { while(resize) {
resize = 0; resize = 0;
checkCUDAError("c_new_maxneighs memset", cudaMemset(c_new_maxneighs, c_neighbor->maxneighs, sizeof(int) )); checkCUDAError("c_new_maxneighs memset", cudaMemset(c_new_maxneighs, 0, sizeof(int) ));
// TODO call compute_neigborhood kernel here // TODO call compute_neigborhood kernel here
const int num_blocks = ceil((float)atom->Nlocal / (float)num_threads_per_block); const int num_blocks = ceil((float)atom->Nlocal / (float)num_threads_per_block);
@ -598,10 +598,13 @@ void buildNeighbor_cuda(Atom *atom, Neighbor *neighbor, Atom *c_atom, Neighbor *
c_new_maxneighs, c_new_maxneighs,
cutneighsq); cutneighsq);
checkCUDAError( "PeekAtLastError ComputeNeighbor", cudaPeekAtLastError() );
checkCUDAError( "DeviceSync ComputeNeighbor", cudaDeviceSynchronize() );
// TODO copy the value of c_new_maxneighs back to host and check if it has been modified // TODO copy the value of c_new_maxneighs back to host and check if it has been modified
int new_maxneighs; int new_maxneighs;
checkCUDAError("c_new_maxneighs memcpy back", cudaMemcpy(&new_maxneighs, c_new_maxneighs, sizeof(int), cudaMemcpyDeviceToHost)); checkCUDAError("c_new_maxneighs memcpy back", cudaMemcpy(&new_maxneighs, c_new_maxneighs, sizeof(int), cudaMemcpyDeviceToHost));
if (new_maxneighs >= c_neighbor->maxneighs){ if (new_maxneighs > c_neighbor->maxneighs){
resize = 1; resize = 1;
} }