nbnxm_gpu_buffer_ops_internal_hip.cpp File Reference

#include "gmxpre.h"
#include "hip/hip_runtime.h"
#include "gromacs/gpu_utils/typecasts_cuda_hip.h"
#include "gromacs/gpu_utils/vectype_ops_hip.h"
#include "gromacs/math/functions.h"
#include "gromacs/nbnxm/grid.h"
#include "gromacs/nbnxm/hip/nbnxm_hip_types.h"
#include "gromacs/nbnxm/nbnxm_gpu_buffer_ops_internal.h"

Include dependency graph for nbnxm_gpu_buffer_ops_internal_hip.cpp:

Description

Define HIP kernel (and its wrapper) for transforming position coordinates from rvec to nbnxm layout.

Author

Alan Gray alang.nosp@m.@nvi.nosp@m.dia.c.nosp@m.om

Jon Vincent jvinc.nosp@m.ent@.nosp@m.nvidi.nosp@m.a.co.nosp@m.m

Szilard Pall pall..nosp@m.szil.nosp@m.ard@g.nosp@m.mail.nosp@m..com

Paul Bauer paul..nosp@m.baue.nosp@m.r.q@g.nosp@m.mail.nosp@m..com

Functions
static __global__ void	anonymous_namespace{nbnxm_gpu_buffer_ops_internal_hip.cpp}::nbnxmKernelTransformXToXq (int numColumns, float4 *__restrict__ gm_xq, const float3 *__restrict__ gm_x, const int *__restrict__ gm_atomIndex, const int *__restrict__ gm_numAtoms, const int *__restrict__ gm_cellIndex, int cellOffset, int numAtomsPerCell)
	HIP kernel for transforming position coordinates from rvec to nbnxm layout. More...
void	gmx::launchNbnxmKernelTransformXToXq (const Grid &grid, NbnxmGpu *nb, DeviceBuffer< Float3 > d_x, const DeviceStream &deviceStream, unsigned int numColumnsMax, int gridId)
	Launch coordinate layout conversion kernel. More...

Variables
static constexpr int	anonymous_namespace{nbnxm_gpu_buffer_ops_internal_hip.cpp}::c_bufOpsThreadsPerBlock = 64
	Number of HIP threads in a block.