#include "gmxpre.h"
#include "pme_solve_sycl.h"
#include <cassert>
#include "gromacs/gpu_utils/gmxsycl.h"
#include "gromacs/gpu_utils/sycl_kernel_utils.h"
#include "gromacs/math/units.h"
#include "pme_gpu_constants.h"
Implements PME GPU Fourier grid solving in SYCL.
- Author
- Mark Abraham mark..nosp@m.j.ab.nosp@m.raham.nosp@m.@gma.nosp@m.il.co.nosp@m.m
|
using | mode = sycl::access_mode |
|
|
template<GridOrdering gridOrdering, bool computeEnergyAndVirial, int subGroupSize> |
auto | makeSolveKernel (sycl::handler &cgh, DeviceAccessor< float, mode::read > a_splineModuli, SolveKernelParams solveKernelParams, OptionalAccessor< float, mode::read_write, computeEnergyAndVirial > a_virialAndEnergy, DeviceAccessor< float, mode::read_write > a_fourierGrid) |
| PME complex grid solver kernel function. More...
|
|
#define INSTANTIATE |
( |
|
subGroupSize | ) |
|
Value:
The kernel for PME solve.
Definition: pme_solve_sycl.h:75
Kernel class instantiations.
template<GridOrdering gridOrdering, bool computeEnergyAndVirial, int subGroupSize>
auto makeSolveKernel |
( |
sycl::handler & |
cgh, |
|
|
DeviceAccessor< float, mode::read > |
a_splineModuli, |
|
|
SolveKernelParams |
solveKernelParams, |
|
|
OptionalAccessor< float, mode::read_write, computeEnergyAndVirial > |
a_virialAndEnergy, |
|
|
DeviceAccessor< float, mode::read_write > |
a_fourierGrid |
|
) |
| |
PME complex grid solver kernel function.
- Template Parameters
-
gridOrdering | Specifies the dimension ordering of the complex grid. |
computeEnergyAndVirial | Tells if the reciprocal energy and virial should be computed. |
subGroupSize | Describes the width of a SYCL subgroup |