Gromacs
2022.5
|
#include "gmxpre.h"
#include "leapfrog_gpu_internal.h"
#include "gromacs/gpu_utils/devicebuffer.h"
#include "gromacs/gpu_utils/gmxsycl.h"
#include "gromacs/math/vec.h"
#include "gromacs/mdlib/leapfrog_gpu.h"
#include "gromacs/mdtypes/group.h"
#include "gromacs/utility/arrayref.h"
#include "gromacs/utility/fatalerror.h"
#include "gromacs/utility/template_mp.h"
Implements Leap-Frog using SYCL.
This file contains SYCL implementation of back-end specific code for Leap-Frog.
Classes | |
class | LeapFrogKernel< numTempScaleValues, velocityScaling > |
Class name for leap-frog kernel. More... | |
Functions | |
template<NumTempScaleValues numTempScaleValues, VelocityScalingType velocityScaling> | |
auto | gmx::leapFrogKernel (sycl::handler &cgh, DeviceAccessor< Float3, mode::read_write > a_x, DeviceAccessor< Float3, mode::discard_write > a_xp, DeviceAccessor< Float3, mode::read_write > a_v, DeviceAccessor< Float3, mode::read > a_f, DeviceAccessor< float, mode::read > a_inverseMasses, float dt, OptionalAccessor< float, mode::read, numTempScaleValues!=NumTempScaleValues::None > a_lambdas, OptionalAccessor< unsigned short, mode::read, numTempScaleValues==NumTempScaleValues::Multiple > a_tempScaleGroups, Float3 prVelocityScalingMatrixDiagonal) |
Main kernel for the Leap-Frog integrator. More... | |
template<NumTempScaleValues numTempScaleValues, VelocityScalingType velocityScaling, class... Args> | |
static sycl::event | gmx::launchLeapFrogKernel (const DeviceStream &deviceStream, int numAtoms, Args &&...args) |
Leap Frog SYCL kernel launch code. | |
static NumTempScaleValues | gmx::getTempScalingType (bool doTemperatureScaling, int numTempScaleValues) |
Convert doTemperatureScaling and numTempScaleValues to NumTempScaleValues. | |
template<class... Args> | |
static sycl::event | gmx::launchLeapFrogKernel (NumTempScaleValues tempScalingType, VelocityScalingType prVelocityScalingType, Args &&...args) |
Select templated kernel and launch it. | |
void | gmx::launchLeapFrogKernel (int numAtoms, DeviceBuffer< Float3 > d_x, DeviceBuffer< Float3 > d_xp, DeviceBuffer< Float3 > d_v, DeviceBuffer< Float3 > d_f, DeviceBuffer< float > d_inverseMasses, float dt, bool doTemperatureScaling, int numTempScaleValues, DeviceBuffer< unsigned short > d_tempScaleGroups, DeviceBuffer< float > d_lambdas, VelocityScalingType prVelocityScalingType, Float3 prVelocityScalingMatrixDiagonal, const DeviceStream &deviceStream) |
Backend-specific function to launch GPU Leap Frog kernel. More... | |