Improve docs and naming for MdModulesNotifiers

Remove explicit dependency of EnergyData on thermostat / barostats

Add unit test for the nonbonded fep kernel.

Workaround for compilation with hipSYCL 0.9.1

Expose init_interaction_const and move to correct header/source

Relax default tolerance on pressure comparisons

Fixes #3975

Replace defines with constexpr in ishift

Constructor for gmx_ekindata_t

Moved init_ekin_data from tgroup header/source to group and made
it the constructor for gmx_ekindata_t. Also made gmx_ekindata_t a
class since the number of threads is only needed by constructor
and destructor and can thus be private. Further refactoring of
members of gmx_ekindata_t would eliminate the need for the nthreads
data member entirely. Other members we not made private to avoid a
ripple of changes adding "()" to numerous lines in several files.

Prepare SelectionCollection for parallel analysis

Add a fourth coordinate to virial tests

Allow thermostat to scale velocities before and after step

Velocity scaling thermostats are currently only able to scale
velocities before the respective propagator step. This change allows
to also scale velocities at the end of the step. This functionality
is needed by the Nose-Hoover thermostat.

This requires an additional template parameter in the `updateVelocities`
function, and an additional callback in the
`PropagatorThermostatConnection` struct to allow temperature coupling
algorithms to get a view to the end velocity scaling.

This also sorts the forward declarations in modularsimulatorinterfaces.h
alphabetically to simplify follow-up changes.

Refs #3423

Add a note on early return in device buffer operations

Functions that (re)allocate, clear and copy to or from
device buffer have an early return in case of zero operation
size. This is a safety mechnism that allows calling these
functions even if the buffers are not intialized, if the
size of the respective buffer is set to zero. A note is
added to the devicebuffer.h that explains why these early
returns are needed.

Clean some headers in NBNXM GPU

1. Remove declarations of non-existing functins.
2. Remove declarations of internal helper functions,
   move some of the helper functions up in the
   respective sources so that their declaration can
   be removed from the headers.
3. Remove unnecessary includes.

Refs #2608

Propagate from current rather than previous position

The current implementation was using the previous position as a starting
point of position propagation. This is conceptually wrong: We want to
propagate starting from the current position.

This had no influence currently, as none of the algorithms present has
multiple position propagations per step. Moving forward, this does however
pose the risk of bugs.

Add GpuEventSynchronizer::isMarked to SYCL verion

This function is needed to work around #3988 when implementing GPU
update (#3932).

template: fix build

Move ScalingMatrix to common GPU update code.

Also, add a constructor to initialize it from an old-style matrix.

Currently, this struct is used only in CUDA, but would be used in SYCL
soon.

Refs #3932.

Add early returns in clearDeviceBuffer(...) in OpenCL and CUDA

Clearing of empty buffer can cause the OpenCL/CUDA API failure.
Early return when requested size to clear is zero allows to
avoid this failure not calling the API for empty buffers.

This also makes the definition of the clearDeviceBuffer(...) in
OpenCL, CUDA similar to its definition in SYCL, where the early
return is already present.

Fixes #4002.

Pass ref_t to initialize_lambdas

Fix several errors when compiling with MSVC

Minor forcerec cleanup

* gmx_bool -> bool
* rename a variable
* rvec -> RVec
* more array and vector
* now has default destructor

Clean up leaks

Rename NBAtomData into NBAtomDataGpu

NBAtomData can only be present in GPU code, hence it should has
Gpu prefix in its name.

Replace gmx_mtop_ilistloop_all

Move domdec/partition into gmx namespace; minor cleanup

* Move functions in partition header into gmx namespace
* Localize variables
* Remove an in-out parameter that is never used

Ensure coordinates are copied for dipole moment calculation

With GPU update, we should wait for the coordinates to be copied to host
before using them for calculating the dipole moment.

This did not seem to cause any issues with CUDA but was causing a unit
test failure with SYCL (with !1329):

    SYCL_BE=PI_OPENCL GMX_USE_GPU_BUFFER_OPS=1 GMX_FORCE_UPDATE_DEFAULT_GPU=1 ./bin/mdrun-test --gtest_filter=EwaldSurfaceTerm/EwaldSurfaceTermTest.WithinTolerances/0

Refs #3930, #3932

Make FileOptionTypes enum class

Changed to enum class and made mapping to legacy types use
EnumerationArray.

Removed Unknown type as it wasn't used except for error checking. Used
Count type instead in its place.

Clarify connection of temperature / pressure coupling and propagators

Temperature and pressure coupling and propagators are linked, as most
thermostats and barostats apply some type of scaling to positions and
/ or velocities. The temperature and pressure coupling objects must
therefore be linked to the propagators to properly function.

The current implementation did not allow for much flexibility, as it
only allowed propagators to decide whether to connect to a thermostat
or barostat or not. To allow for more complex algorithms, more
flexibility is required. To that end, this change introduces a
PropagatorTag object, which tags the propagators, and allows the
temperature and pressure coupling objects to pick which propagator to
act on. Additionally, this moves the responsibility of connecting to
the coupling algorithm rather than the propagator, which makes more
sense conceptually.

Refs #3423

Add unit test for 1-4 interactions.

Remove dead code in sasa module

If needed, the code can be found again in git.

Make EnumerationArray::size() not static

It's formally correct that the method can be static, but this makes
code more complex at the point of use for the doubtful benefit of
making clear to the human reader that the value returned by size() is
a compile-time constant. The compiler can always see it.

ArrayRefs and const refs in global_stat

Use const refs where possible in global_stat. Use ArrayRef for
signalBuffer and return an int instead of passing pointer to int.
Also movemost variable definitions to closer to their first use.

Use EnumerationArray for pull isAngleType

Remove PME coordinate send code duplication

Unifies codepath for PME coordinate send for default and GPU direct
codepaths. The only effect this has on ordering of activities is for a
run config with GPU buffer ops, without GPU direct communication and
without GPU update, where the (CPU-side) PME-PP coordinate transfer
will now be slightly delayed to allow for an H2D launch of coordinate
data.

Partly addresses #3914

Move cluster methods to its own file

Breaks out some of the methods in gmx cluster into a separate file
before future refactoring.

Refs #4000

Fix doxygen for updateMDLeapfrogGeneral

Fixes an issue created in !1350.

Use more ArrayRefs in signatures of update impl class

Part of ongoing work to make refactoring mdatoms easier. This only
changes the impl class of update. A followup change can propagate
using ArrayRef to the update class itself, which will make it easier
to call update in both tests and in nblib.

Use using for ArrayRef and RVec in pull code

Unify gpu_free(...) function in OpenCL, CUDA and SYCL versions of NBNXM

Refs #2608

Fix conditional on when DtoH forces copy occur

d2d4a50b4c636c203028c5bff311924ec15e7825 introduced performance
regression with forces copied from device to host on each step.
This fixes the issue by reinstantiating proper condition on the
copy call.

Fixes #4001
Refs #2608

Move dynamic pull groups and pull nthreads

Moved dynamic pull groups into their respective coordinates.
Set the number of threads per pull group instead of globally.
Now all thread parallel regions run on 1 thread with less than 100
local atoms.

Use default constructor for the device buffers that can be uninitialized

Some buffers in NBNXM are initialized conditionally, which posseses
a problem for the OpenCL kernel jit compilation, where these buffers
are used as arguments.

Refs #2608

Fix some include guards in gmx ana

Clang tidy would complain about the defines starting with _.

Allow the use of only one backend at a time with SYCL-DPCPP

This is not a hard requirement; multiple backends can be used together
just fine.

But in DPCPP, the same physical device can appear as different virtual
devices provided by different backends (e.g., the same GPU can be accessible
via both OpenCL and L0).
Thus, using devices from two backends is more likely to be a user error than
the desired behavior. In this function, we choose the backend with the most compatible
devices. In case of a tie, we choose OpenCL (if present), or some arbitrary backend
among those with the most devices.

In hipSYCL, the problem of using the same device twice is unlikely to manifest.
It has (as of 2021-03-03) another issues: D2D copy between different backends
is not allowed. We don't use D2D in SYCL yet. Additionally, hipSYCL does not implement
the `sycl::platform::get_backend()` function.

Thus, we only do the backend filtering with DPCPP to keep the code
simple.

If needed, the `GMX_GPU_DISABLE_COMPATIBILITY_CHECK` might be used to
disable this limitation, or a backend-specific device filtering (e.g.,
`SYCL_BE` and `SYCL_DEVICE_FILTER`) can be used to only
provide the desired backend.

Turn t_forcerec.shift_vec into an std::vector of gmx::RVec

Change function signatures so that call sites have minimal changes

Add coordIndex to t_pull_coord

The pull coordinate index is an integral part of a pull coordinate
and can not be changed. Thus it can be part of t_pull_coord which
allows avoiding index passing in several functions.

Move responsibility for checking bondeds in reverse topology

The reverse topology is a data structure distributed across all
domains that should refer exactly once to all bonded interactions
present in the global topology. We check that this is true at run
time, but need a complex implementation to avoid doing extra
collective communication that supports running what is effectively an
assertion that the reverse topology construction is correct.

Past attempts to improve this (e.g. 8b6c99a1) left the code in an
intermediate state where the responsibility for this check was mostly
implemented in do_md rather than compute_globals. That gave the
impression of compute_globals() returning an integer
parameter. However as above it's really an implementation detail of a
check of the reverse topology, and so the logic and memory for that
check should be in the domain decomposition module and object.

This change move the details of that check to the rest of the code
that handles the reverse topology. It renames several variables and
adds documentation for them.

This simplifies parts of the implementation of the modular simulator,
and prepares for implementing the proposal for #3887 by making all
outputs from global reduction local to the module that consumes them.

Refs #3421
Refs #3887

Add tests for calcvir

Fix webpage build

MR !1235 (dff1d7e309eb751b6b7585853501c544919db59f) had a missing empty
line, which upset Sphinx.

Avoid unnecessary CPU force buffer clearing

Only clear CPU force buffers on steps that actually use the buffers.

Partly addresses #3970

Move nsgrid header/source from mdlib to domdec

The function get_nsgrid_boundaries is only called in domdec module
so it should live there. Removed an unused constant and moved another
to the source which is the only consumer. Also removed a todo which
no longer makes sense.

Remove unused LSAN suppressions

Some unknown refactoring has left these unneeded

Update developer guide on error handling

Deferred to C++ core guidelines on most content, plus
some GROMACS-specific clarification.

Avoid enqueueing the same event multiple times in nbnxn_atomdata_x_to_nbat_x_gpu

With several non-local grids, we were calling
nbnxnInsertNonlocalGpuDependency(nb, NonLocal) in nbnxn_gpu_x_to_nbat_x
after enqueueing the transformation kernel for each one.

Now, we only call it once after submitting all kernels for
given locality (marking the event for Local, enqueueing the wait for it
for NonLocal).

Refs #3988.

Change logic for checksumming availability

Changed order of checks to see if we can calculate the checksum or not.

Use ArrayRefs instead of mdatoms in gmx_nb_free_energy_kernel signature

In addition to helping facilitate the coming mdatoms refactoring,
this is needed to facilitate calling gmx_nb_free_energy_kernel from
nblib without constructing mdatoms.

Allow setting nbnxm cluster size for SYCL

Rename GMX_OPENCL_NB_CLUSTER_SIZE to GMX_GPU_NB_CLUSTER_SIZE and use its
value in SYCL builds too. The former is still taken into account at
cmake-time, but users are advised to use the latter.

Refs #3847 #3933 #3935

Localize variables in do_lbfgs()

Also replaced pointers by std::vector.

Use more vectors for EnergyOutput private members

Unify gpu_upload_shiftvec(...) function in NBNXM

Refs. #2608

Constructor for t_tabledata

Added default and specialized constructors for t_tabledata. With
one additional minor change this means that smalloc header is no
longer needed in forcetable.cpp.

ArrayRef and const ref in gmx_nb_free_energy_kernel

Use const ref where possible in nonbonded free energy kernels. Also
use ArrayRef for passing coordinates.

Use ArrayRef in special forces

Pass ArrayRef<RVec> instead of rvec* in edsam, pull_rotation, and
imd. Also removed unneeded header basedefinitions.

A followup change can refactor communicate_group_positions to take
ArrayRefs.

Fix some doxygen errors

Fixed unused variable warning

Use ArrayRef in signatures of constraints and some pulling

In order to use ArrayRef in constr, settle, shake, and lincs, some
pull code also needed to be refactored. Part of preparation for
refactoring of t_mdatoms.

Inline wallcycle counting functions

Inlines start and stop functions for wallcycle counting.

Unify nbnxn_gpu_init_x_to_nbat_x

Previously, we had this function only for CUDA, but we will need it for
SYCL (#3932). OpenCL implementation is unlikely to be needed, but should
not hurt either.

Refs. #2608

Add missing GMX_ASSERT(haveCopiedXFromGpu)

We make this assertion every time we call
stateGpu->waitCoordinatesReadyOnHost(AtomLocality::Local) in the
do_force function.

This does not fix any bugs, just makes the code a bit more consistent in
how it checks self-consistency.

Refs #3988.

Remove nb_kernel_data_t

All data was copied from gmx_enerdata_t. Now use ArrayRefs to pass
the parameters to the kernels.

Remove unneeded members from nb_kernel_data_t

Several unused members are removed from nb_kernel_data_t. Additionally,
some members that can be passed directly from the gmx_enerdata_t
struct are removed, and instead added to the call signature of
gmx_nb_free_energy_kernel. This expansion of the call signature
is the only way to get rid of the real* members of nb_kernel_data_t.
Also, function with no implementation is removed.

Use TextWriter in msd tests

Removes raw file pointer for mdp input file writing.

Fixes #3986

Add comparison operator to BasicVector

The default comparison seemed to compare the memory addresses
and not the elements.

Fix deviceIdsAssigned() to return a unique list of IDs

The method originally returned the rank's task-to-device mapping
rather than the unique list of devices that have been assigned tasks.
When #ranks > #devices this list contains duplicate device IDs.

Due to the changes made in e2a2fe80 the vector returned by
deviceIdsAssigned() was passed to setupGpuDevicePeerAccess()
which results in redundant attempts to repeatedly enable peer access for
the same device pairs.

Refs #3980

Forward declare ArrayRef more and inlcude basedefinitions where needed

Remove many includes of ArrayRef that can be forward declared. Also
remove includes of basedefinitions in many headers. In some cases
removing basedefinitinos meant also changing gmx_bool to bool.
Note that no change to source files is needed, apart from possibly
adding the basedefintions include, because gmx_bool uses true and
false under the hood.

Modernize t_nrnb

* Use array instead of pointer.
* Remove unused functions.
* Add doxygen to inc_nrnb.
* Use inline function instead of macro in all places where possible.
* Remove unused headers.

Use ArrayRef in calc_mu

More const ref in shellfc

Pass a const ref to t_mdatoms in shellfc functions. Part of work
to make refactoring t_mdatoms easier.

Detect CUDA-aware support in underlying MPI implementation

Checks CUDA-aware MPI support at cmake configuration time and at runtime. Print CUDA-aware MPI support status with gmx -version.

Use ArrayRef in do_ewald

Also removed a function with no implementation.

Use StringToEnumValueConverter and enumValueToString more

Follow up from MR 1213

Fix compiler and linter warnings

- Unused variables in pme_pp.cpp
- Insufficient NOLINTing in update.cpp

Remove incorrect assertion for GPU timing in SYCL

- The expression was backwards: it was failing when timing was disabled
  instead of failing when it is enabled.
- The expression was run on a freshly-initialized data structure, and
  the value of bDoTime was always overridden later, so it was not
  actually checking anything.

Since we anyway never set bDoTime to true in SYCL build, we might as
well remove this assertion here altogether.

Introduced in !1296 (c4a672b93b655ff674b671fd10ebe728b2d19ec8).

Clean up leaks in t_inputrec

Refs #3984

Remove a misplaced update cycle closing

This was removed in b80b7a327b1 but got added back later presumably
during a merge.

Refs #3764

Rename sendFToPpCudaDirect

sendFSynchronizerToPpCudaDirect expresses what the method actually does.

Move the atomToInteractionLocality(...) into locality.h

The atomToInteractionLocality(...) does not depend on NBNXM
and can naturally reside alongside the definitions of localitiy
enumerations. This allows to combine haveGpuShortRangeWork(...)
functions in NBNXM, avoding confuciton of having two very similar
functions with different signatures.

Store buffer without offset in GpuForceReduction::Impl::baseForce_

- Store the pointer to the beginning of the array and apply offset to
it when invoking the kernel. Offsets to other arrays are applied at
the invocation time, so it makes sense to do the same with baseForce_.
- This also avoid storing pointers to the middle of the device buffer,
which is not a good fit for OpenCL and SYCL.

This change is preparation for adding a SYCL implementation of GPU
update (Issue #3932).

Unify init_gpu function in NBNXM

Refs. #2608

Use DeviceBuffer<RVec> in GPU force reduction and PME code

... instead of raw device pointers.

Preparation for #3932. PME change is incidental, the main focus is
GpuForceReduction.

Enable compiling LeapFrog with hipSYCL

No testing yet. Trivial change after !1268.

Closes #3941.

Bulk change to const ref for mtop

Change function signatures in mtop_lookup.h and mtop_util.h to be const
ref only for mtop. Also change signatures in other files as I went
along.

This change is only refactoring.

Make wallcycle comments proper doxygen

Should reduce noise in other refactoring commits.

Remove const cast in update code

Use more arrayref in listed forces sigatures

Pass members of t_mdatoms directly in many places in listed_forces.
Part of work on making refactor of t_mdatoms easier.

Remove mdatoms from forceprovider call signature

This makes it possible to test the forceprovider without having to build
a partial mdatoms datastructure first. Needed for future refactoring of
mdatoms.

Use ArrayRef in ewald_LRcorrection

This will make refactoring t_mdatoms easier. Also changed
signatures of calculateLongRangeNonbondeds to use more
ArrayRef.

Make non bonded energy terms enum class

Also renamed some of the enum values.

Fix a warning from CUDA 10.1

Discovered during nightly build:

warning: static_assert with no message is a C++17 extension [-Wc++17-extensions]

Use ArrayRef in gmx_pme_send_parameters

This will make refactoring t_mdatoms easier.