ICODE-ADC/lammps
4
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

class n func rename

Browse Source
This commit is contained in:
Eddy BARRAUD
2024-06-17 10:03:06 +02:00
parent a4ac48addf
commit 4be1b41aef
8 changed files with 74 additions and 74 deletions
Download Patch File Download Diff File Expand all files Collapse all files

26
lib/gpu/lal_dpd_coul_slater_long.cpp
View File

@ -24,26 +24,26 @@ const char *dpd_coul_slater_long=0;
#include "lal_dpd_coul_slater_long.h" #include "lal_dpd_coul_slater_long.h"
#include <cassert> #include <cassert>
namespace LAMMPS_AL { namespace LAMMPS_AL {
#define DPDChargedT DPDCharged<numtyp, acctyp> #define DPDCoulSlaterLongT DPDCoulSlaterLong<numtyp, acctyp>
extern Device<PRECISION,ACC_PRECISION> device; extern Device<PRECISION,ACC_PRECISION> device;
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
DPDChargedT::DPDCharged() : BaseDPD<numtyp,acctyp>(), _allocated(false) { DPDCoulSlaterLongT::DPDCoulSlaterLong() : BaseDPD<numtyp,acctyp>(), _allocated(false) {
} }
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
DPDChargedT::~DPDCharged() { DPDCoulSlaterLongT::~DPDCoulSlaterLong() {
clear(); clear();
} }
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
int DPDChargedT::bytes_per_atom(const int max_nbors) const { int DPDCoulSlaterLongT::bytes_per_atom(const int max_nbors) const {
return this->bytes_per_atom_atomic(max_nbors); return this->bytes_per_atom_atomic(max_nbors);
} }
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
int DPDChargedT::init(const int ntypes, int DPDCoulSlaterLongT::init(const int ntypes,
double **host_cutsq, double **host_a0, double **host_cutsq, double **host_a0,
double **host_gamma, double **host_sigma, double **host_gamma, double **host_sigma,
double **host_cut_dpd, double **host_cut_dpdsq, double **host_cut_dpd, double **host_cut_dpdsq,
@ -75,7 +75,7 @@ int DPDChargedT::init(const int ntypes,
int success; int success;
bool need_charges = true; bool need_charges = true;
success=this->init_atomic(nlocal,nall,max_nbors,maxspecial,cell_size, success=this->init_atomic(nlocal,nall,max_nbors,maxspecial,cell_size,
gpu_split,_screen,dpd_coul_slater_long,"k_dpd_charged",onetype, extra_fields, need_charges); gpu_split,_screen,dpd_coul_slater_long,"k_dpd_coul_slater_long",onetype, extra_fields, need_charges);
if (success!=0) if (success!=0)
return success; return success;
@ -140,7 +140,7 @@ int DPDChargedT::init(const int ntypes,
} }
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
void DPDChargedT::clear() { void DPDCoulSlaterLongT::clear() {
if (!_allocated) if (!_allocated)
return; return;
_allocated=false; _allocated=false;
@ -153,15 +153,15 @@ void DPDChargedT::clear() {
} }
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
double DPDChargedT::host_memory_usage() const { double DPDCoulSlaterLongT::host_memory_usage() const {
return this->host_memory_usage_atomic()+sizeof(DPDCharged<numtyp,acctyp>); return this->host_memory_usage_atomic()+sizeof(DPDCoulSlaterLong<numtyp,acctyp>);
} }
// --------------------------------------------------------------------------- // ---------------------------------------------------------------------------
// Calculate energies, forces, and torques // Calculate energies, forces, and torques
// --------------------------------------------------------------------------- // ---------------------------------------------------------------------------
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
int DPDChargedT::loop(const int eflag, const int vflag) { int DPDCoulSlaterLongT::loop(const int eflag, const int vflag) {
int nall = this->atom->nall(); int nall = this->atom->nall();
// signal that we need to transfer extra data from the host // signal that we need to transfer extra data from the host
@ -215,7 +215,7 @@ int DPDChargedT::loop(const int eflag, const int vflag) {
} }
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
void DPDChargedT::update_coeff(int ntypes, double **host_a0, double **host_gamma, void DPDCoulSlaterLongT::update_coeff(int ntypes, double **host_a0, double **host_gamma,
double **host_sigma, double **host_cut_dpd) double **host_sigma, double **host_cut_dpd)
{ {
UCL_H_Vec<numtyp> host_write(_lj_types*_lj_types*32,*(this->ucl_device), UCL_H_Vec<numtyp> host_write(_lj_types*_lj_types*32,*(this->ucl_device),
@ -229,11 +229,11 @@ void DPDChargedT::update_coeff(int ntypes, double **host_a0, double **host_gamma
// --------------------------------------------------------------------------- // ---------------------------------------------------------------------------
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
void DPDChargedT::get_extra_data(double *host_q) { void DPDCoulSlaterLongT::get_extra_data(double *host_q) {
q = host_q; q = host_q;
} }
template class DPDCharged<PRECISION,ACC_PRECISION>; template class DPDCoulSlaterLong<PRECISION,ACC_PRECISION>;
} }

8
lib/gpu/lal_dpd_coul_slater_long.cu
View File

@ -10,8 +10,8 @@
// This file is part of the LAMMPS Accelerator Library (LAMMPS_AL) // This file is part of the LAMMPS Accelerator Library (LAMMPS_AL)
// __________________________________________________________________________ // __________________________________________________________________________
// //
// begin : Jan 15, 2014 // begin : May 28, 2024
// email : nguyentd@ornl.gov // email : eddy.barraud@outlook.fr
// *************************************************************************** // ***************************************************************************
#if defined(NV_KERNEL) || defined(USE_HIP) #if defined(NV_KERNEL) || defined(USE_HIP)
@ -162,7 +162,7 @@ _texture_2d( vel_tex,int4);
} }
#endif #endif
__kernel void k_dpd_charged(const __global numtyp4 *restrict x_, __kernel void k_dpd_coul_slater_long(const __global numtyp4 *restrict x_,
const __global numtyp4 *restrict extra, const __global numtyp4 *restrict extra,
const __global numtyp4 *restrict coeff, const __global numtyp4 *restrict coeff,
const int lj_types, const int lj_types,
@ -336,7 +336,7 @@ __kernel void k_dpd_charged(const __global numtyp4 *restrict x_,
ans,engv); ans,engv);
} }
__kernel void k_dpd_charged_fast(const __global numtyp4 *restrict x_, __kernel void k_dpd_coul_slater_long_fast(const __global numtyp4 *restrict x_,
const __global numtyp4 *restrict extra, const __global numtyp4 *restrict extra,
const __global numtyp4 *restrict coeff_in, const __global numtyp4 *restrict coeff_in,
const __global numtyp *restrict sp_lj_in, const __global numtyp *restrict sp_lj_in,

6
lib/gpu/lal_dpd_coul_slater_long.h
View File

@ -21,10 +21,10 @@
namespace LAMMPS_AL { namespace LAMMPS_AL {
template <class numtyp, class acctyp> template <class numtyp, class acctyp>
class DPDCharged : public BaseDPD<numtyp, acctyp> { class DPDCoulSlaterLong : public BaseDPD<numtyp, acctyp> {
public: public:
DPDCharged(); DPDCoulSlaterLong();
~DPDCharged(); ~DPDCoulSlaterLong();
/// Clear any previous data and set up for a new LAMMPS run /// Clear any previous data and set up for a new LAMMPS run
/** \param max_nbors initial number of rows in the neighbor matrix /** \param max_nbors initial number of rows in the neighbor matrix

16
lib/gpu/lal_dpd_coul_slater_long_ext.cpp
View File

@ -22,12 +22,12 @@
using namespace std; using namespace std;
using namespace LAMMPS_AL; using namespace LAMMPS_AL;
static DPDCharged<PRECISION,ACC_PRECISION> DPDCMF; static DPDCoulSlaterLong<PRECISION,ACC_PRECISION> DPDCMF;
// --------------------------------------------------------------------------- // ---------------------------------------------------------------------------
// Allocate memory on host and device and copy constants to device // Allocate memory on host and device and copy constants to device
// --------------------------------------------------------------------------- // ---------------------------------------------------------------------------
int dpd_charged_gpu_init(const int ntypes, double **host_cutsq, double **host_a0, double **host_gamma, int dpd_coul_slater_long_gpu_init(const int ntypes, double **host_cutsq, double **host_a0, double **host_gamma,
double **host_sigma, double **host_cut_dpd, double **host_cut_dpdsq, double **host_cut_slatersq, double **host_sigma, double **host_cut_dpd, double **host_cut_dpdsq, double **host_cut_slatersq,
double **host_scale, double *special_lj, const int inum, double **host_scale, double *special_lj, const int inum,
const int nall, const int max_nbors, const int maxspecial, const double cell_size, const int nall, const int max_nbors, const int maxspecial, const double cell_size,
@ -93,11 +93,11 @@ int dpd_charged_gpu_init(const int ntypes, double **host_cutsq, double **host_a0
return init_ok; return init_ok;
} }
void dpd_charged_gpu_clear() { void dpd_coul_slater_long_gpu_clear() {
DPDCMF.clear(); DPDCMF.clear();
} }
int ** dpd_charged_gpu_compute_n(const int ago, const int inum_full, const int nall, int ** dpd_coul_slater_long_gpu_compute_n(const int ago, const int inum_full, const int nall,
double **host_x, int *host_type, double *sublo, double **host_x, int *host_type, double *sublo,
double *subhi, tagint *tag, int **nspecial, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, tagint **special, const bool eflag, const bool vflag,
@ -112,7 +112,7 @@ int ** dpd_charged_gpu_compute_n(const int ago, const int inum_full, const int n
host_v, dtinvsqrt, seed, timestep, boxlo, prd); host_v, dtinvsqrt, seed, timestep, boxlo, prd);
} }
void dpd_charged_gpu_compute(const int ago, const int inum_full, const int nall, void dpd_coul_slater_long_gpu_compute(const int ago, const int inum_full, const int nall,
double **host_x, int *host_type, int *ilist, int *numj, double **host_x, int *host_type, int *ilist, int *numj,
int **firstneigh, const bool eflag, const bool vflag, int **firstneigh, const bool eflag, const bool vflag,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
@ -125,17 +125,17 @@ void dpd_charged_gpu_compute(const int ago, const int inum_full, const int nall,
tag, host_v, dtinvsqrt, seed, timestep, nlocal, boxlo, prd); tag, host_v, dtinvsqrt, seed, timestep, nlocal, boxlo, prd);
} }
void dpd_charged_gpu_update_coeff(int ntypes, double **host_a0, double **host_gamma, void dpd_coul_slater_long_gpu_update_coeff(int ntypes, double **host_a0, double **host_gamma,
double **host_sigma, double **host_cut_dpd) double **host_sigma, double **host_cut_dpd)
{ {
DPDCMF.update_coeff(ntypes,host_a0,host_gamma,host_sigma, host_cut_dpd); DPDCMF.update_coeff(ntypes,host_a0,host_gamma,host_sigma, host_cut_dpd);
} }
void dpd_charged_gpu_get_extra_data(double *host_q) { void dpd_coul_slater_long_gpu_get_extra_data(double *host_q) {
DPDCMF.get_extra_data(host_q); DPDCMF.get_extra_data(host_q);
} }
double dpd_charged_gpu_bytes() { double dpd_coul_slater_long_gpu_bytes() {
return DPDCMF.host_memory_usage(); return DPDCMF.host_memory_usage();
} }

32
src/DPD-BASIC/pair_dpd_coul_slater_long.cpp
View File

@ -41,7 +41,7 @@ static constexpr double EPSILON = 1.0e-10;
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairDPDCharged::PairDPDCharged(LAMMPS *lmp) : Pair(lmp) PairDPDCoulSlaterLong::PairDPDCoulSlaterLong(LAMMPS *lmp) : Pair(lmp)
{ {
writedata = 1; writedata = 1;
ewaldflag = pppmflag = 1; ewaldflag = pppmflag = 1;
@ -51,7 +51,7 @@ PairDPDCharged::PairDPDCharged(LAMMPS *lmp) : Pair(lmp)
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairDPDCharged::~PairDPDCharged() PairDPDCoulSlaterLong::~PairDPDCoulSlaterLong()
{ {
if (copymode) return; if (copymode) return;
@ -75,7 +75,7 @@ PairDPDCharged::~PairDPDCharged()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairDPDCharged::compute(int eflag, int vflag) void PairDPDCoulSlaterLong::compute(int eflag, int vflag)
{ {
int i,j,ii,jj,inum,jnum,itype,jtype; int i,j,ii,jj,inum,jnum,itype,jtype;
double qtmp,xtmp,ytmp,ztmp,delx,dely,delz,evdwl,ecoul,fpair; double qtmp,xtmp,ytmp,ztmp,delx,dely,delz,evdwl,ecoul,fpair;
@ -215,7 +215,7 @@ void PairDPDCharged::compute(int eflag, int vflag)
allocate all arrays allocate all arrays
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::allocate() void PairDPDCoulSlaterLong::allocate()
{ {
int i,j; int i,j;
allocated = 1; allocated = 1;
@ -246,7 +246,7 @@ void PairDPDCharged::allocate()
global settings global settings
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::settings(int narg, char **arg) void PairDPDCoulSlaterLong::settings(int narg, char **arg)
{ {
// params : T cut_dpd seed lambda cut_coul // params : T cut_dpd seed lambda cut_coul
if (narg != 5) error->all(FLERR,"Illegal pair_style command"); if (narg != 5) error->all(FLERR,"Illegal pair_style command");
@ -276,7 +276,7 @@ void PairDPDCharged::settings(int narg, char **arg)
set coeffs for one or more type pairs set coeffs for one or more type pairs
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::coeff(int narg, char **arg) void PairDPDCoulSlaterLong::coeff(int narg, char **arg)
{ {
if (narg < 4 || narg > 6) if (narg < 4 || narg > 6)
error->all(FLERR,"Incorrect args for pair coefficients"); error->all(FLERR,"Incorrect args for pair coefficients");
@ -320,7 +320,7 @@ void PairDPDCharged::coeff(int narg, char **arg)
init specific to this pair style init specific to this pair style
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::init_style() void PairDPDCoulSlaterLong::init_style()
{ {
if (comm->ghost_velocity == 0) if (comm->ghost_velocity == 0)
error->all(FLERR,"Pair dpd requires ghost atoms store velocity"); error->all(FLERR,"Pair dpd requires ghost atoms store velocity");
@ -353,7 +353,7 @@ void PairDPDCharged::init_style()
return the DPD cutoff for uncharged return the DPD cutoff for uncharged
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
double PairDPDCharged::init_one(int i, int j) double PairDPDCoulSlaterLong::init_one(int i, int j)
{ {
if (setflag[i][j] == 0) error->all(FLERR,"All pair coeffs are not set"); if (setflag[i][j] == 0) error->all(FLERR,"All pair coeffs are not set");
@ -380,7 +380,7 @@ double PairDPDCharged::init_one(int i, int j)
proc 0 writes to restart file proc 0 writes to restart file
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::write_restart(FILE *fp) void PairDPDCoulSlaterLong::write_restart(FILE *fp)
{ {
write_restart_settings(fp); write_restart_settings(fp);
@ -405,7 +405,7 @@ void PairDPDCharged::write_restart(FILE *fp)
proc 0 reads from restart file, bcasts proc 0 reads from restart file, bcasts
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::read_restart(FILE *fp) void PairDPDCoulSlaterLong::read_restart(FILE *fp)
{ {
read_restart_settings(fp); read_restart_settings(fp);
@ -440,7 +440,7 @@ void PairDPDCharged::read_restart(FILE *fp)
proc 0 writes to restart file proc 0 writes to restart file
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::write_restart_settings(FILE *fp) void PairDPDCoulSlaterLong::write_restart_settings(FILE *fp)
{ {
fwrite(&temperature,sizeof(double),1,fp); fwrite(&temperature,sizeof(double),1,fp);
fwrite(&cut_global,sizeof(double),1,fp); fwrite(&cut_global,sizeof(double),1,fp);
@ -459,7 +459,7 @@ void PairDPDCharged::write_restart_settings(FILE *fp)
proc 0 reads from restart file, bcasts proc 0 reads from restart file, bcasts
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::read_restart_settings(FILE *fp) void PairDPDCoulSlaterLong::read_restart_settings(FILE *fp)
{ {
if (comm->me == 0) { if (comm->me == 0) {
utils::sfread(FLERR,&temperature,sizeof(double),1,fp,nullptr,error); utils::sfread(FLERR,&temperature,sizeof(double),1,fp,nullptr,error);
@ -493,7 +493,7 @@ void PairDPDCharged::read_restart_settings(FILE *fp)
proc 0 writes to data file proc 0 writes to data file
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::write_data(FILE *fp) void PairDPDCoulSlaterLong::write_data(FILE *fp)
{ {
for (int i = 1; i <= atom->ntypes; i++) for (int i = 1; i <= atom->ntypes; i++)
fprintf(fp,"%d %g %g\n",i,a0[i][i],gamma[i][i]); fprintf(fp,"%d %g %g\n",i,a0[i][i],gamma[i][i]);
@ -503,7 +503,7 @@ void PairDPDCharged::write_data(FILE *fp)
proc 0 writes all pairs to data file proc 0 writes all pairs to data file
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDCharged::write_data_all(FILE *fp) void PairDPDCoulSlaterLong::write_data_all(FILE *fp)
{ {
for (int i = 1; i <= atom->ntypes; i++) for (int i = 1; i <= atom->ntypes; i++)
for (int j = i; j <= atom->ntypes; j++) for (int j = i; j <= atom->ntypes; j++)
@ -512,7 +512,7 @@ void PairDPDCharged::write_data_all(FILE *fp)
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
double PairDPDCharged::single(int i, int j, int itype, int jtype, double rsq, double PairDPDCoulSlaterLong::single(int i, int j, int itype, int jtype, double rsq,
double factor_coul, double factor_dpd, double &fforce) double factor_coul, double factor_dpd, double &fforce)
{ {
double r,rinv,wd,phi; double r,rinv,wd,phi;
@ -555,7 +555,7 @@ double PairDPDCharged::single(int i, int j, int itype, int jtype, double rsq,
return energy; return energy;
} }
void *PairDPDCharged::extract(const char *str, int &dim) void *PairDPDCoulSlaterLong::extract(const char *str, int &dim)
{ {
if (strcmp(str,"cut_coul") == 0) { if (strcmp(str,"cut_coul") == 0) {
dim = 0; dim = 0;

12
src/DPD-BASIC/pair_dpd_coul_slater_long.h
View File

@ -13,21 +13,21 @@
#ifdef PAIR_CLASS #ifdef PAIR_CLASS
// clang-format off // clang-format off
PairStyle(dpd/coul/slater/long,PairDPDCharged); PairStyle(dpd/coul/slater/long,PairDPDCoulSlaterLong);
// clang-format on // clang-format on
#else #else
#ifndef LMP_PAIR_DPD_CHARGED_H #ifndef LMP_PAIR_DPD_COUL_SLATER_LONG_H
#define LMP_PAIR_DPD_CHARGED_H #define LMP_PAIR_DPD_COUL_SLATER_LONG_H
#include "pair.h" #include "pair.h"
namespace LAMMPS_NS { namespace LAMMPS_NS {
class PairDPDCharged : public Pair { class PairDPDCoulSlaterLong : public Pair {
public: public:
PairDPDCharged(class LAMMPS *); PairDPDCoulSlaterLong(class LAMMPS *);
~PairDPDCharged() override; ~PairDPDCoulSlaterLong() override;
void compute(int, int) override; void compute(int, int) override;
void settings(int, char **) override; void settings(int, char **) override;
void coeff(int, char **) override; void coeff(int, char **) override;

36
src/GPU/pair_dpd_coul_slater_long_gpu.cpp
View File

@ -39,30 +39,30 @@ using namespace EwaldConst;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int dpd_charged_gpu_init(const int ntypes, double **cutsq, double **host_a0, double **host_gamma, int dpd_coul_slater_long_gpu_init(const int ntypes, double **cutsq, double **host_a0, double **host_gamma,
double **host_sigma, double **host_cut_dpd, double **host_cut_dpdsq, double **host_cut_slatersq, double **host_sigma, double **host_cut_dpd, double **host_cut_dpdsq, double **host_cut_slatersq,
double **host_scale, double *special_lj, const int inum, double **host_scale, double *special_lj, const int inum,
const int nall, const int max_nbors, const int maxspecial, const double cell_size, const int nall, const int max_nbors, const int maxspecial, const double cell_size,
int &gpu_mode, FILE *screen, double *host_special_coul, int &gpu_mode, FILE *screen, double *host_special_coul,
const double qqrd2e, const double g_ewald, const double lamda); const double qqrd2e, const double g_ewald, const double lamda);
void dpd_charged_gpu_clear(); void dpd_coul_slater_long_gpu_clear();
int **dpd_charged_gpu_compute_n(const int ago, const int inum_full, const int nall, double **host_x, int **dpd_coul_slater_long_gpu_compute_n(const int ago, const int inum_full, const int nall, double **host_x,
int *host_type, double *sublo, double *subhi, tagint *tag, int **nspecial, int *host_type, double *sublo, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, const bool eatom, tagint **special, const bool eflag, const bool vflag, const bool eatom,
const bool vatom, int &host_start, int **ilist, int **jnum, const bool vatom, int &host_start, int **ilist, int **jnum,
const double cpu_time, bool &success, double **host_v, const double cpu_time, bool &success, double **host_v,
const double dtinvsqrt, const int seed, const int timestep, double *boxlo, const double dtinvsqrt, const int seed, const int timestep, double *boxlo,
double *prd); double *prd);
void dpd_charged_gpu_compute(const int ago, const int inum_full, const int nall, double **host_x, void dpd_coul_slater_long_gpu_compute(const int ago, const int inum_full, const int nall, double **host_x,
int *host_type, int *ilist, int *numj, int **firstneigh, const bool eflag, int *host_type, int *ilist, int *numj, int **firstneigh, const bool eflag,
const bool vflag, const bool eatom, const bool vatom, int &host_start, const bool vflag, const bool eatom, const bool vatom, int &host_start,
const double cpu_time, bool &success, tagint *tag, double **host_v, const double cpu_time, bool &success, tagint *tag, double **host_v,
const double dtinvsqrt, const int seed, const int timestep, const int nlocal, const double dtinvsqrt, const int seed, const int timestep, const int nlocal,
double *boxlo, double *prd); double *boxlo, double *prd);
void dpd_charged_gpu_get_extra_data(double *host_q); void dpd_coul_slater_long_gpu_get_extra_data(double *host_q);
double dpd_charged_gpu_bytes(); double dpd_coul_slater_long_gpu_bytes();
static constexpr double EPSILON = 1.0e-10; static constexpr double EPSILON = 1.0e-10;
@ -207,7 +207,7 @@ static constexpr double EPSILON = 1.0e-10;
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairDPDChargedGPU::PairDPDChargedGPU(LAMMPS *lmp) : PairDPDCharged(lmp), gpu_mode(GPU_FORCE) PairDPDCoulSlaterLongGPU::PairDPDCoulSlaterLongGPU(LAMMPS *lmp) : PairDPDCoulSlaterLong(lmp), gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
@ -220,14 +220,14 @@ PairDPDChargedGPU::PairDPDChargedGPU(LAMMPS *lmp) : PairDPDCharged(lmp), gpu_mod
free all arrays free all arrays
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
PairDPDChargedGPU::~PairDPDChargedGPU() PairDPDCoulSlaterLongGPU::~PairDPDCoulSlaterLongGPU()
{ {
dpd_charged_gpu_clear(); dpd_coul_slater_long_gpu_clear();
} }
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairDPDChargedGPU::compute(int eflag, int vflag) void PairDPDCoulSlaterLongGPU::compute(int eflag, int vflag)
{ {
ev_init(eflag, vflag); ev_init(eflag, vflag);
@ -240,7 +240,7 @@ void PairDPDChargedGPU::compute(int eflag, int vflag)
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
double *q = atom->q; double *q = atom->q;
dpd_charged_gpu_get_extra_data(q); dpd_coul_slater_long_gpu_get_extra_data(q);
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
double sublo[3], subhi[3]; double sublo[3], subhi[3];
@ -255,7 +255,7 @@ void PairDPDChargedGPU::compute(int eflag, int vflag)
domain->bbox(domain->sublo_lamda, domain->subhi_lamda, sublo, subhi); domain->bbox(domain->sublo_lamda, domain->subhi_lamda, sublo, subhi);
} }
inum = atom->nlocal; inum = atom->nlocal;
firstneigh = dpd_charged_gpu_compute_n( firstneigh = dpd_coul_slater_long_gpu_compute_n(
neighbor->ago, inum, nall, atom->x, atom->type, sublo, subhi, atom->tag, atom->nspecial, neighbor->ago, inum, nall, atom->x, atom->type, sublo, subhi, atom->tag, atom->nspecial,
atom->special, eflag, vflag, eflag_atom, vflag_atom, host_start, &ilist, &numneigh, atom->special, eflag, vflag, eflag_atom, vflag_atom, host_start, &ilist, &numneigh,
cpu_time, success, atom->v, dtinvsqrt, seed, update->ntimestep, domain->boxlo, domain->prd); cpu_time, success, atom->v, dtinvsqrt, seed, update->ntimestep, domain->boxlo, domain->prd);
@ -264,7 +264,7 @@ void PairDPDChargedGPU::compute(int eflag, int vflag)
ilist = list->ilist; ilist = list->ilist;
numneigh = list->numneigh; numneigh = list->numneigh;
firstneigh = list->firstneigh; firstneigh = list->firstneigh;
dpd_charged_gpu_compute(neighbor->ago, inum, nall, atom->x, atom->type, ilist, numneigh, firstneigh, dpd_coul_slater_long_gpu_compute(neighbor->ago, inum, nall, atom->x, atom->type, ilist, numneigh, firstneigh,
eflag, vflag, eflag_atom, vflag_atom, host_start, cpu_time, success, atom->tag, eflag, vflag, eflag_atom, vflag_atom, host_start, cpu_time, success, atom->tag,
atom->v, dtinvsqrt, seed, update->ntimestep, atom->nlocal, domain->boxlo, atom->v, dtinvsqrt, seed, update->ntimestep, atom->nlocal, domain->boxlo,
domain->prd); domain->prd);
@ -284,7 +284,7 @@ void PairDPDChargedGPU::compute(int eflag, int vflag)
init specific to this pair style init specific to this pair style
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairDPDChargedGPU::init_style() void PairDPDCoulSlaterLongGPU::init_style()
{ {
if (comm->ghost_velocity == 0) if (comm->ghost_velocity == 0)
@ -315,7 +315,7 @@ void PairDPDChargedGPU::init_style()
if (atom->molecular != Atom::ATOMIC) maxspecial = atom->maxspecial; if (atom->molecular != Atom::ATOMIC) maxspecial = atom->maxspecial;
int mnf = 5e-2 * neighbor->oneatom; int mnf = 5e-2 * neighbor->oneatom;
int success = int success =
dpd_charged_gpu_init(atom->ntypes + 1, cutsq, a0, gamma, sigma, dpd_coul_slater_long_gpu_init(atom->ntypes + 1, cutsq, a0, gamma, sigma,
cut_dpd, cut_dpdsq, cut_slatersq, scale, cut_dpd, cut_dpdsq, cut_slatersq, scale,
force->special_lj, atom->nlocal, force->special_lj, atom->nlocal,
atom->nlocal + atom->nghost, mnf, maxspecial, cell_size, gpu_mode, screen, atom->nlocal + atom->nghost, mnf, maxspecial, cell_size, gpu_mode, screen,
@ -328,15 +328,15 @@ void PairDPDChargedGPU::init_style()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
double PairDPDChargedGPU::memory_usage() double PairDPDCoulSlaterLongGPU::memory_usage()
{ {
double bytes = Pair::memory_usage(); double bytes = Pair::memory_usage();
return bytes + dpd_charged_gpu_bytes(); return bytes + dpd_coul_slater_long_gpu_bytes();
} }
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairDPDChargedGPU::cpu_compute(int start, int inum, int eflag, int /* vflag */, int *ilist, void PairDPDCoulSlaterLongGPU::cpu_compute(int start, int inum, int eflag, int /* vflag */, int *ilist,
int *numneigh, int **firstneigh) int *numneigh, int **firstneigh)
{ {
int i, j, ii, jj, jnum, itype, jtype; int i, j, ii, jj, jnum, itype, jtype;

12
src/GPU/pair_dpd_coul_slater_long_gpu.h
View File

@ -13,21 +13,21 @@
#ifdef PAIR_CLASS #ifdef PAIR_CLASS
// clang-format off // clang-format off
PairStyle(dpd/coul/slater/long/gpu,PairDPDChargedGPU); PairStyle(dpd/coul/slater/long/gpu,PairDPDCoulSlaterLongGPU);
// clang-format on // clang-format on
#else #else
#ifndef LMP_PAIR_DPD_CHARGED_GPU_H #ifndef LMP_PAIR_DPD_COUL_SLATER_LONG_GPU_H
#define LMP_PAIR_DPD_CHARGED_GPU_H #define LMP_PAIR_DPD_COUL_SLATER_LONG_GPU_H
#include "pair_dpd_coul_slater_long.h" #include "pair_dpd_coul_slater_long.h"
namespace LAMMPS_NS { namespace LAMMPS_NS {
class PairDPDChargedGPU : public PairDPDCharged { class PairDPDCoulSlaterLongGPU : public PairDPDCoulSlaterLong {
public: public:
PairDPDChargedGPU(LAMMPS *lmp); PairDPDCoulSlaterLongGPU(LAMMPS *lmp);
~PairDPDChargedGPU() override; ~PairDPDCoulSlaterLongGPU() override;
void cpu_compute(int, int, int, int, int *, int *, int **); void cpu_compute(int, int, int, int, int *, int *, int **);
void compute(int, int) override; void compute(int, int) override;
void init_style() override; void init_style() override;