diff --git a/src/GPU/gpu_extra.h b/src/GPU/gpu_extra.h index 2f4cdcb2ca..9d982e3a8c 100644 --- a/src/GPU/gpu_extra.h +++ b/src/GPU/gpu_extra.h @@ -29,7 +29,8 @@ namespace GPU_EXTRA { MPI_Allreduce(&error_flag, &all_success, 1, MPI_INT, MPI_MIN, world); if (all_success != 0) { if (all_success == -1) - error->all(FLERR,"Accelerated style in input script but no fix gpu"); + error->all(FLERR, + "The package gpu command is required for gpu styles"); else if (all_success == -2) error->all(FLERR, "Could not find/initialize a specified accelerator device"); @@ -64,44 +65,43 @@ namespace GPU_EXTRA { /* ERROR/WARNING messages: -E: Accelerated style in input script but no fix gpu +E: The package gpu command is required for gpu styles -UNDOCUMENTED +Self-explanatory. E: Could not find/initialize a specified accelerator device -UNDOCUMENTED +Could not initialize at least one of the devices specified for the gpu package E: Insufficient memory on accelerator -UNDOCUMENTED +There is insufficient memory on one of the devices specified for the gpu +package E: GPU library not compiled for this accelerator -UNDOCUMENTED +Self-explanatory. E: Double precision is not supported on this accelerator -UNDOCUMENTED +Self-explanatory E: Unable to initialize accelerator for use -UNDOCUMENTED +There was a problem initializing an accelerator for the gpu package E: Accelerator sharing is not currently supported on system -UNDOCUMENTED +Multiple MPI processes cannot share the accelerator on your system. For NVIDIA +GPUs, see the nvidia-smi command to change this setting. E: GPU particle split must be set to 1 for this pair style. -UNDOCUMENTED +For this pair style, you cannot run part of the force calculation on the host. +See the package command. E: Unknown error in GPU library -UNDOCUMENTED - -E: The package gpu command is required for gpu styles - -UNDOCUMENTED +Self-explanatory. */ diff --git a/src/GPU/pair_buck_coul_cut_gpu.cpp b/src/GPU/pair_buck_coul_cut_gpu.cpp index 308a7dee3d..a3db1e0dd3 100644 --- a/src/GPU/pair_buck_coul_cut_gpu.cpp +++ b/src/GPU/pair_buck_coul_cut_gpu.cpp @@ -115,7 +115,7 @@ void PairBuckCoulCutGPU::compute(int eflag, int vflag) atom->nlocal, domain->boxlo, domain->prd); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startnlocal, domain->boxlo, domain->prd); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startnlocal, domain->boxlo, domain->prd); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); // communicate derivative of embedding function @@ -152,8 +152,6 @@ void PairEAMGPU::init_style() if (force->newton_pair) error->all(FLERR,"Cannot use newton pair with eam/gpu pair style"); - if (!allocated) error->all(FLERR,"Not allocate memory eam/gpu pair style"); - // convert read-in file(s) to arrays and spline them file2array(); diff --git a/src/GPU/pair_eam_gpu.h b/src/GPU/pair_eam_gpu.h index 686f276e8b..24c9e22f32 100644 --- a/src/GPU/pair_eam_gpu.h +++ b/src/GPU/pair_eam_gpu.h @@ -54,17 +54,13 @@ class PairEAMGPU : public PairEAM { /* ERROR/WARNING messages: -E: Out of memory on GPGPU +E: Insufficient memory on accelerator -GPU memory is limited. Reduce the size of the problem or increase the -number of GPUs. +There is insufficient memory on one of the devices specified for the gpu +package E: Cannot use newton pair with eam/gpu pair style Self-explanatory. -E: Not allocate memory eam/gpu pair style - -UNDOCUMENTED - */ diff --git a/src/GPU/pair_gayberne_gpu.cpp b/src/GPU/pair_gayberne_gpu.cpp index e0604bd5a1..f4a094b7ab 100644 --- a/src/GPU/pair_gayberne_gpu.cpp +++ b/src/GPU/pair_gayberne_gpu.cpp @@ -137,7 +137,7 @@ void PairGayBerneGPU::compute(int eflag, int vflag) cpu_time, success, quat); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_start < inum) { cpu_time = MPI_Wtime(); diff --git a/src/GPU/pair_gayberne_gpu.h b/src/GPU/pair_gayberne_gpu.h index 8a249c1da6..00e80a965b 100644 --- a/src/GPU/pair_gayberne_gpu.h +++ b/src/GPU/pair_gayberne_gpu.h @@ -53,10 +53,10 @@ E: Pair gayberne/gpu requires atom style ellipsoid Self-explanatory. -E: Out of memory on GPGPU +E: Insufficient memory on accelerator -GPU memory is limited. Reduce the size of the problem or increase the -number of GPUs. +There is insufficient memory on one of the devices specified for the gpu +package E: Cannot use newton pair with gayberne/gpu pair style diff --git a/src/GPU/pair_lj96_cut_gpu.cpp b/src/GPU/pair_lj96_cut_gpu.cpp index 76076a643a..b9c70b80af 100644 --- a/src/GPU/pair_lj96_cut_gpu.cpp +++ b/src/GPU/pair_lj96_cut_gpu.cpp @@ -107,7 +107,7 @@ void PairLJ96CutGPU::compute(int eflag, int vflag) vflag_atom, host_start, cpu_time, success); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startnlocal, domain->boxlo, domain->prd); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startnlocal, domain->boxlo, domain->prd); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startnlocal, domain->boxlo, domain->prd); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startnlocal, domain->boxlo, domain->prd); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startnlocal, domain->boxlo, domain->prd); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_start < inum) { cpu_time = MPI_Wtime(); diff --git a/src/GPU/pair_resquared_gpu.h b/src/GPU/pair_resquared_gpu.h index 3efacd97ea..6f95a8b17d 100644 --- a/src/GPU/pair_resquared_gpu.h +++ b/src/GPU/pair_resquared_gpu.h @@ -53,10 +53,10 @@ E: Pair resquared/gpu requires atom style ellipsoid Self-explanatory. -E: Out of memory on GPGPU +E: Insufficient memory on accelerator -GPU memory is limited. Reduce the size of the problem or increase the -number of GPUs. +There is insufficient memory on one of the devices specified for the gpu +package E: Cannot use newton pair with resquared/gpu pair style diff --git a/src/GPU/pair_table_gpu.cpp b/src/GPU/pair_table_gpu.cpp index a69fb1c6e0..fcd7e7edf8 100644 --- a/src/GPU/pair_table_gpu.cpp +++ b/src/GPU/pair_table_gpu.cpp @@ -114,7 +114,7 @@ void PairTableGPU::compute(int eflag, int vflag) vflag_atom, host_start, cpu_time, success); } if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startone(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (host_startq, domain->boxlo, delxinv, delyinv, delzinv); if (!success) - error->one(FLERR,"Out of memory on GPGPU"); + error->one(FLERR,"Insufficient memory on accelerator"); if (flag != 0) error->one(FLERR,"Out of range atoms - cannot compute PPPM"); - int i; + // If need per-atom energies/virials, also do particle map on host + // concurrently with GPU calculations + + if (evflag_atom) { + memory->destroy(part2grid); + nmax = atom->nmax; + memory->create(part2grid,nmax,3,"pppm:part2grid"); + particle_map(); + } + + int i,j; // convert atoms from box to lamda coords @@ -212,6 +222,29 @@ void PPPMGPU::compute(int eflag, int vflag) FFT_SCALAR qscale = force->qqrd2e * scale; PPPM_GPU_API(interp)(qscale); + // Compute per-atom energy/virial on host if requested + + if (evflag_atom) { + fillbrick_peratom(); + fieldforce_peratom(); + double *q = atom->q; + int nlocal = atom->nlocal; + + if (eflag_atom) { + for (i = 0; i < nlocal; i++) { + eatom[i] *= 0.5; + eatom[i] -= g_ewald*q[i]*q[i]/MY_PIS + MY_PI2*q[i]*qsum / + (g_ewald*g_ewald*volume); + eatom[i] *= qscale; + } + } + + if (vflag_atom) { + for (i = 0; i < nlocal; i++) + for (j = 0; j < 6; j++) vatom[i][j] *= 0.5*q[i]*qscale; + } + } + // sum energy across procs and add in volume-dependent term if (eflag_global) { @@ -745,6 +778,10 @@ void PPPMGPU::poisson() work1[n++] *= scaleinv * greensfn[i]; } + // extra FFTs for per-atom energy/virial + + if (evflag_atom) poisson_peratom(); + // compute gradients of V(r) in each of 3 dims by transformimg -ik*V(k) // FFT leaves data in 3d brick decomposition // copy it into inner portion of vdx,vdy,vdz arrays @@ -873,6 +910,12 @@ double PPPMGPU::memory_usage() bytes += nfft_both * sizeof(double); bytes += nfft_both*5 * sizeof(FFT_SCALAR); bytes += 2 * nbuf * sizeof(double); + + if (peratom_allocate_flag) { + bytes += 7 * nbrick * sizeof(FFT_SCALAR); + bytes += 2 * nbuf_peratom * sizeof(FFT_SCALAR); + } + return bytes + PPPM_GPU_API(bytes)(); } diff --git a/src/GPU/pppm_gpu.h b/src/GPU/pppm_gpu.h index d8bed15bc8..8fb8a5f5d2 100644 --- a/src/GPU/pppm_gpu.h +++ b/src/GPU/pppm_gpu.h @@ -69,10 +69,10 @@ E: Cannot use order greater than 8 with pppm/gpu. Self-explanatory. -E: Out of memory on GPGPU +E: Insufficient memory on accelerator -GPU memory is limited. Reduce the size of the problem or increase the -number of GPUs. +There is insufficient memory on one of the devices specified for the gpu +package E: Out of range atoms - cannot compute PPPM