git-svn-id: svn://svn.icms.temple.edu/lammps-ro/trunk@14220 f3b2605a-c512-4ea7-a41b-209d697bcdaa

This commit is contained in:
sjplimp
2015-10-30 20:04:06 +00:00
parent 0aa77408f8
commit 354e20f431
1219 changed files with 10111 additions and 10111 deletions

View File

@ -11,7 +11,7 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "mpi.h" #include <mpi.h>
#include "compute_erotate_asphere.h" #include "compute_erotate_asphere.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"

View File

@ -15,8 +15,8 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "mpi.h" #include <mpi.h>
#include "string.h" #include <string.h>
#include "compute_temp_asphere.h" #include "compute_temp_asphere.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"
@ -55,7 +55,7 @@ ComputeTempAsphere::ComputeTempAsphere(LAMMPS *lmp, int narg, char **arg) :
int iarg = 3; int iarg = 3;
while (iarg < narg) { while (iarg < narg) {
if (strcmp(arg[iarg],"bias") == 0) { if (strcmp(arg[iarg],"bias") == 0) {
if (iarg+2 > narg) if (iarg+2 > narg)
error->all(FLERR,"Illegal compute temp/asphere command"); error->all(FLERR,"Illegal compute temp/asphere command");
tempbias = 1; tempbias = 1;
int n = strlen(arg[iarg+1]) + 1; int n = strlen(arg[iarg+1]) + 1;
@ -63,7 +63,7 @@ ComputeTempAsphere::ComputeTempAsphere(LAMMPS *lmp, int narg, char **arg) :
strcpy(id_bias,arg[iarg+1]); strcpy(id_bias,arg[iarg+1]);
iarg += 2; iarg += 2;
} else if (strcmp(arg[iarg],"dof") == 0) { } else if (strcmp(arg[iarg],"dof") == 0) {
if (iarg+2 > narg) if (iarg+2 > narg)
error->all(FLERR,"Illegal compute temp/asphere command"); error->all(FLERR,"Illegal compute temp/asphere command");
if (strcmp(arg[iarg+1],"rotate") == 0) mode = ROTATE; if (strcmp(arg[iarg+1],"rotate") == 0) mode = ROTATE;
else if (strcmp(arg[iarg+1],"all") == 0) mode = ALL; else if (strcmp(arg[iarg+1],"all") == 0) mode = ALL;
@ -107,7 +107,7 @@ void ComputeTempAsphere::init()
if (tempbias) { if (tempbias) {
int i = modify->find_compute(id_bias); int i = modify->find_compute(id_bias);
if (i < 0) if (i < 0)
error->all(FLERR,"Could not find compute ID for temperature bias"); error->all(FLERR,"Could not find compute ID for temperature bias");
tbias = modify->compute[i]; tbias = modify->compute[i];
if (tbias->tempflag == 0) if (tbias->tempflag == 0)
@ -267,7 +267,7 @@ double ComputeTempAsphere::compute_scalar()
MPI_Allreduce(&t,&scalar,1,MPI_DOUBLE,MPI_SUM,world); MPI_Allreduce(&t,&scalar,1,MPI_DOUBLE,MPI_SUM,world);
if (dynamic || tempbias == 2) dof_compute(); if (dynamic || tempbias == 2) dof_compute();
if (dof < 0.0 && natoms_temp > 0.0) if (dof < 0.0 && natoms_temp > 0.0)
error->all(FLERR,"Temperature compute degrees of freedom < 0"); error->all(FLERR,"Temperature compute degrees of freedom < 0");
scalar *= tfactor; scalar *= tfactor;
return scalar; return scalar;

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "string.h" #include <string.h>
#include "stdlib.h" #include <stdlib.h>
#include "math.h" #include <math.h>
#include "math_extra.h" #include "math_extra.h"
#include "fix_nh_asphere.h" #include "fix_nh_asphere.h"
#include "atom.h" #include "atom.h"

View File

@ -11,7 +11,7 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "string.h" #include <string.h>
#include "fix_nph_asphere.h" #include "fix_nph_asphere.h"
#include "modify.h" #include "modify.h"
#include "error.h" #include "error.h"

View File

@ -11,7 +11,7 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "string.h" #include <string.h>
#include "fix_npt_asphere.h" #include "fix_npt_asphere.h"
#include "modify.h" #include "modify.h"
#include "error.h" #include "error.h"

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "string.h" #include <string.h>
#include "fix_nve_asphere.h" #include "fix_nve_asphere.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"

View File

@ -11,9 +11,9 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "string.h" #include <string.h>
#include "stdlib.h" #include <stdlib.h>
#include "fix_nve_asphere_noforce.h" #include "fix_nve_asphere_noforce.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"

View File

@ -11,9 +11,9 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "string.h" #include <string.h>
#include "fix_nve_line.h" #include "fix_nve_line.h"
#include "atom.h" #include "atom.h"
#include "atom_vec_line.h" #include "atom_vec_line.h"

View File

@ -11,9 +11,9 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "string.h" #include <string.h>
#include "fix_nve_tri.h" #include "fix_nve_tri.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"

View File

@ -11,7 +11,7 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "string.h" #include <string.h>
#include "fix_nvt_asphere.h" #include "fix_nvt_asphere.h"
#include "group.h" #include "group.h"
#include "modify.h" #include "modify.h"

View File

@ -15,10 +15,10 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_gayberne.h" #include "pair_gayberne.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"

View File

@ -11,10 +11,10 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_line_lj.h" #include "pair_line_lj.h"
#include "atom.h" #include "atom.h"
#include "atom_vec_line.h" #include "atom_vec_line.h"

View File

@ -15,10 +15,10 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_resquared.h" #include "pair_resquared.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"

View File

@ -11,10 +11,10 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_tri_lj.h" #include "pair_tri_lj.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"
@ -476,7 +476,7 @@ void PairTriLJ::init_style()
neighbor->request(this,instance_me); neighbor->request(this,instance_me);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
init for one type pair i,j and corresponding j,i init for one type pair i,j and corresponding j,i
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */

View File

@ -11,7 +11,7 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "stdlib.h" #include <stdlib.h>
#include "body_nparticle.h" #include "body_nparticle.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom_vec_body.h" #include "atom_vec_body.h"
@ -25,14 +25,14 @@ using namespace LAMMPS_NS;
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
BodyNparticle::BodyNparticle(LAMMPS *lmp, int narg, char **arg) : BodyNparticle::BodyNparticle(LAMMPS *lmp, int narg, char **arg) :
Body(lmp, narg, arg) Body(lmp, narg, arg)
{ {
if (narg != 3) error->all(FLERR,"Invalid body nparticle command"); if (narg != 3) error->all(FLERR,"Invalid body nparticle command");
int nmin = force->inumeric(FLERR,arg[1]); int nmin = force->inumeric(FLERR,arg[1]);
int nmax = force->inumeric(FLERR,arg[2]); int nmax = force->inumeric(FLERR,arg[2]);
if (nmin <= 0 || nmin > nmax) if (nmin <= 0 || nmin > nmax)
error->all(FLERR,"Invalid body nparticle command"); error->all(FLERR,"Invalid body nparticle command");
size_forward = 0; size_forward = 0;
@ -90,7 +90,7 @@ int BodyNparticle::unpack_border_body(AtomVecBody::Bonus *bonus, double *buf)
populate bonus data structure with data file values populate bonus data structure with data file values
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void BodyNparticle::data_body(int ibonus, int ninteger, int ndouble, void BodyNparticle::data_body(int ibonus, int ninteger, int ndouble,
char **ifile, char **dfile) char **ifile, char **dfile)
{ {
AtomVecBody::Bonus *bonus = &avec->bonus[ibonus]; AtomVecBody::Bonus *bonus = &avec->bonus[ibonus];
@ -98,22 +98,22 @@ void BodyNparticle::data_body(int ibonus, int ninteger, int ndouble,
// error in data file if any values are NULL // error in data file if any values are NULL
for (int i = 0; i < ninteger; i++) for (int i = 0; i < ninteger; i++)
if (ifile[i] == NULL) if (ifile[i] == NULL)
error->one(FLERR,"Invalid format in Bodies section of data file"); error->one(FLERR,"Invalid format in Bodies section of data file");
for (int i = 0; i < ndouble; i++) for (int i = 0; i < ndouble; i++)
if (dfile[i] == NULL) if (dfile[i] == NULL)
error->one(FLERR,"Invalid format in Bodies section of data file"); error->one(FLERR,"Invalid format in Bodies section of data file");
// set ninteger, ndouble in bonus and allocate 2 vectors of ints, doubles // set ninteger, ndouble in bonus and allocate 2 vectors of ints, doubles
if (ninteger != 1) if (ninteger != 1)
error->one(FLERR,"Incorrect # of integer values in " error->one(FLERR,"Incorrect # of integer values in "
"Bodies section of data file"); "Bodies section of data file");
int nsub = atoi(ifile[0]); int nsub = atoi(ifile[0]);
if (nsub < 1) if (nsub < 1)
error->one(FLERR,"Incorrect integer value in " error->one(FLERR,"Incorrect integer value in "
"Bodies section of data file"); "Bodies section of data file");
if (ndouble != 6 + 3*nsub) if (ndouble != 6 + 3*nsub)
error->one(FLERR,"Incorrect # of floating-point values in " error->one(FLERR,"Incorrect # of floating-point values in "
"Bodies section of data file"); "Bodies section of data file");
@ -144,7 +144,7 @@ void BodyNparticle::data_body(int ibonus, int ninteger, int ndouble,
double max; double max;
max = MAX(inertia[0],inertia[1]); max = MAX(inertia[0],inertia[1]);
max = MAX(max,inertia[2]); max = MAX(max,inertia[2]);
if (inertia[0] < EPSILON*max) inertia[0] = 0.0; if (inertia[0] < EPSILON*max) inertia[0] = 0.0;
if (inertia[1] < EPSILON*max) inertia[1] = 0.0; if (inertia[1] < EPSILON*max) inertia[1] = 0.0;
if (inertia[2] < EPSILON*max) inertia[2] = 0.0; if (inertia[2] < EPSILON*max) inertia[2] = 0.0;
@ -169,9 +169,9 @@ void BodyNparticle::data_body(int ibonus, int ninteger, int ndouble,
double cross[3]; double cross[3];
MathExtra::cross3(ex_space,ey_space,cross); MathExtra::cross3(ex_space,ey_space,cross);
if (MathExtra::dot3(cross,ez_space) < 0.0) MathExtra::negate3(ez_space); if (MathExtra::dot3(cross,ez_space) < 0.0) MathExtra::negate3(ez_space);
// create initial quaternion // create initial quaternion
MathExtra::exyz_to_q(ex_space,ey_space,ez_space,bonus->quat); MathExtra::exyz_to_q(ex_space,ey_space,ez_space,bonus->quat);
// bonus->dvalue = sub-particle displacements in body frame // bonus->dvalue = sub-particle displacements in body frame

View File

@ -11,8 +11,8 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "string.h" #include <string.h>
#include "compute_body_local.h" #include "compute_body_local.h"
#include "atom.h" #include "atom.h"
#include "atom_vec_body.h" #include "atom_vec_body.h"
@ -61,7 +61,7 @@ ComputeBodyLocal::ComputeBodyLocal(LAMMPS *lmp, int narg, char **arg) :
int indexmax = bptr->noutcol(); int indexmax = bptr->noutcol();
for (int i = 0; i < nvalues; i++) { for (int i = 0; i < nvalues; i++) {
if (which[i] == INDEX && (index[i] < 0 || index[i] >= indexmax)) if (which[i] == INDEX && (index[i] < 0 || index[i] >= indexmax))
error->all(FLERR,"Invalid index in compute body/local command"); error->all(FLERR,"Invalid index in compute body/local command");
} }
@ -92,7 +92,7 @@ void ComputeBodyLocal::init()
int nlocal = atom->nlocal; int nlocal = atom->nlocal;
for (int i = 0; i < nlocal; i++) for (int i = 0; i < nlocal; i++)
if (mask[i] & groupbit) if (mask[i] & groupbit)
if (body[i] < 0) nonbody = 1; if (body[i] < 0) nonbody = 1;
int flag; int flag;

View File

@ -11,9 +11,9 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "string.h" #include <string.h>
#include "fix_nve_body.h" #include "fix_nve_body.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"

View File

@ -11,10 +11,10 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_body.h" #include "pair_body.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"
@ -178,7 +178,7 @@ void PairBody::compute(int eflag, int vflag)
torque[i][0] += ti[0]; torque[i][0] += ti[0];
torque[i][1] += ti[1]; torque[i][1] += ti[1];
torque[i][2] += ti[2]; torque[i][2] += ti[2];
if (newton_pair || j < nlocal) { if (newton_pair || j < nlocal) {
fj[0] = -delx*fpair; fj[0] = -delx*fpair;
fj[1] = -dely*fpair; fj[1] = -dely*fpair;
@ -465,7 +465,7 @@ void PairBody::body2space(int i)
AtomVecBody::Bonus *bonus = &avec->bonus[ibonus]; AtomVecBody::Bonus *bonus = &avec->bonus[ibonus];
int nsub = bptr->nsub(bonus); int nsub = bptr->nsub(bonus);
double *coords = bptr->coords(bonus); double *coords = bptr->coords(bonus);
dnum[i] = nsub; dnum[i] = nsub;
dfirst[i] = ndiscrete; dfirst[i] = ndiscrete;

View File

@ -15,9 +15,9 @@
Contributing author: Eric Simon (Cray) Contributing author: Eric Simon (Cray)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "string.h" #include <string.h>
#include "stdlib.h" #include <stdlib.h>
#include "angle_class2.h" #include "angle_class2.h"
#include "atom.h" #include "atom.h"
#include "neighbor.h" #include "neighbor.h"

View File

@ -20,7 +20,7 @@ AngleStyle(class2,AngleClass2)
#ifndef LMP_ANGLE_CLASS2_H #ifndef LMP_ANGLE_CLASS2_H
#define LMP_ANGLE_CLASS2_H #define LMP_ANGLE_CLASS2_H
#include "stdio.h" #include <stdio.h>
#include "angle.h" #include "angle.h"
namespace LAMMPS_NS { namespace LAMMPS_NS {

View File

@ -15,8 +15,8 @@
Contributing author: Eric Simon (Cray) Contributing author: Eric Simon (Cray)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdlib.h" #include <stdlib.h>
#include "bond_class2.h" #include "bond_class2.h"
#include "atom.h" #include "atom.h"
#include "neighbor.h" #include "neighbor.h"

View File

@ -20,7 +20,7 @@ BondStyle(class2,BondClass2)
#ifndef LMP_BOND_CLASS2_H #ifndef LMP_BOND_CLASS2_H
#define LMP_BOND_CLASS2_H #define LMP_BOND_CLASS2_H
#include "stdio.h" #include <stdio.h>
#include "bond.h" #include "bond.h"
namespace LAMMPS_NS { namespace LAMMPS_NS {

View File

@ -15,9 +15,9 @@
Contributing author: Eric Simon (Cray) Contributing author: Eric Simon (Cray)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "string.h" #include <string.h>
#include "stdlib.h" #include <stdlib.h>
#include "dihedral_class2.h" #include "dihedral_class2.h"
#include "atom.h" #include "atom.h"
#include "neighbor.h" #include "neighbor.h"
@ -205,8 +205,8 @@ void DihedralClass2::compute(int eflag, int vflag)
MPI_Comm_rank(world,&me); MPI_Comm_rank(world,&me);
if (screen) { if (screen) {
char str[128]; char str[128];
sprintf(str,"Dihedral problem: %d " BIGINT_FORMAT " " sprintf(str,"Dihedral problem: %d " BIGINT_FORMAT " "
TAGINT_FORMAT " " TAGINT_FORMAT " " TAGINT_FORMAT " " TAGINT_FORMAT " "
TAGINT_FORMAT " " TAGINT_FORMAT, TAGINT_FORMAT " " TAGINT_FORMAT,
me,update->ntimestep, me,update->ntimestep,
atom->tag[i1],atom->tag[i2],atom->tag[i3],atom->tag[i4]); atom->tag[i1],atom->tag[i2],atom->tag[i3],atom->tag[i4]);

View File

@ -20,7 +20,7 @@ DihedralStyle(class2,DihedralClass2)
#ifndef LMP_DIHEDRAL_CLASS2_H #ifndef LMP_DIHEDRAL_CLASS2_H
#define LMP_DIHEDRAL_CLASS2_H #define LMP_DIHEDRAL_CLASS2_H
#include "stdio.h" #include <stdio.h>
#include "dihedral.h" #include "dihedral.h"
namespace LAMMPS_NS { namespace LAMMPS_NS {

View File

@ -15,9 +15,9 @@
Contributing author: Eric Simon (Cray) Contributing author: Eric Simon (Cray)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "string.h" #include <string.h>
#include "stdlib.h" #include <stdlib.h>
#include "improper_class2.h" #include "improper_class2.h"
#include "atom.h" #include "atom.h"
#include "neighbor.h" #include "neighbor.h"
@ -156,8 +156,8 @@ void ImproperClass2::compute(int eflag, int vflag)
MPI_Comm_rank(world,&me); MPI_Comm_rank(world,&me);
if (screen) { if (screen) {
char str[128]; char str[128];
sprintf(str,"Improper problem: %d " BIGINT_FORMAT " " sprintf(str,"Improper problem: %d " BIGINT_FORMAT " "
TAGINT_FORMAT " " TAGINT_FORMAT " " TAGINT_FORMAT " " TAGINT_FORMAT " "
TAGINT_FORMAT " " TAGINT_FORMAT, TAGINT_FORMAT " " TAGINT_FORMAT,
me,update->ntimestep, me,update->ntimestep,
atom->tag[i1],atom->tag[i2],atom->tag[i3],atom->tag[i4]); atom->tag[i1],atom->tag[i2],atom->tag[i3],atom->tag[i4]);
@ -853,5 +853,5 @@ void ImproperClass2::write_data(FILE *fp)
for (int i = 1; i <= atom->nimpropertypes; i++) for (int i = 1; i <= atom->nimpropertypes; i++)
fprintf(fp,"%d %g %g %g %g %g %g\n",i,aa_k1[i],aa_k2[i],aa_k3[i], fprintf(fp,"%d %g %g %g %g %g %g\n",i,aa_k1[i],aa_k2[i],aa_k3[i],
aa_theta0_1[i]*180.0/MY_PI,aa_theta0_2[i]*180.0/MY_PI, aa_theta0_1[i]*180.0/MY_PI,aa_theta0_2[i]*180.0/MY_PI,
aa_theta0_3[i]*180.0/MY_PI); aa_theta0_3[i]*180.0/MY_PI);
} }

View File

@ -20,7 +20,7 @@ ImproperStyle(class2,ImproperClass2)
#ifndef LMP_IMPROPER_CLASS2_H #ifndef LMP_IMPROPER_CLASS2_H
#define LMP_IMPROPER_CLASS2_H #define LMP_IMPROPER_CLASS2_H
#include "stdio.h" #include <stdio.h>
#include "improper.h" #include "improper.h"
namespace LAMMPS_NS { namespace LAMMPS_NS {

View File

@ -11,9 +11,9 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_class2.h" #include "pair_lj_class2.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"

View File

@ -11,9 +11,9 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_class2_coul_cut.h" #include "pair_lj_class2_coul_cut.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"

View File

@ -11,10 +11,10 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_lj_class2_coul_long.h" #include "pair_lj_class2_coul_long.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"
@ -136,18 +136,18 @@ void PairLJClass2CoulLong::compute(int eflag, int vflag)
prefactor = qqrd2e * qtmp*q[j]/r; prefactor = qqrd2e * qtmp*q[j]/r;
forcecoul = prefactor * (erfc + EWALD_F*grij*expm2); forcecoul = prefactor * (erfc + EWALD_F*grij*expm2);
if (factor_coul < 1.0) forcecoul -= (1.0-factor_coul)*prefactor; if (factor_coul < 1.0) forcecoul -= (1.0-factor_coul)*prefactor;
} else { } else {
union_int_float_t rsq_lookup; union_int_float_t rsq_lookup;
rsq_lookup.f = rsq; rsq_lookup.f = rsq;
itable = rsq_lookup.i & ncoulmask; itable = rsq_lookup.i & ncoulmask;
itable >>= ncoulshiftbits; itable >>= ncoulshiftbits;
fraction = (rsq_lookup.f - rtable[itable]) * drtable[itable]; fraction = (rsq_lookup.f - rtable[itable]) * drtable[itable];
table = ftable[itable] + fraction*dftable[itable]; table = ftable[itable] + fraction*dftable[itable];
forcecoul = qtmp*q[j] * table; forcecoul = qtmp*q[j] * table;
if (factor_coul < 1.0) { if (factor_coul < 1.0) {
table = ctable[itable] + fraction*dctable[itable]; table = ctable[itable] + fraction*dctable[itable];
prefactor = qtmp*q[j] * table; prefactor = qtmp*q[j] * table;
forcecoul -= (1.0-factor_coul)*prefactor; forcecoul -= (1.0-factor_coul)*prefactor;
} }
} }
} else forcecoul = 0.0; } else forcecoul = 0.0;
@ -172,11 +172,11 @@ void PairLJClass2CoulLong::compute(int eflag, int vflag)
if (eflag) { if (eflag) {
if (rsq < cut_coulsq) { if (rsq < cut_coulsq) {
if (!ncoultablebits || rsq <= tabinnersq) if (!ncoultablebits || rsq <= tabinnersq)
ecoul = prefactor*erfc; ecoul = prefactor*erfc;
else { else {
table = etable[itable] + fraction*detable[itable]; table = etable[itable] + fraction*detable[itable];
ecoul = qtmp*q[j] * table; ecoul = qtmp*q[j] * table;
} }
if (factor_coul < 1.0) ecoul -= (1.0-factor_coul)*prefactor; if (factor_coul < 1.0) ecoul -= (1.0-factor_coul)*prefactor;
} else ecoul = 0.0; } else ecoul = 0.0;
@ -251,7 +251,7 @@ void PairLJClass2CoulLong::settings(int narg, char **arg)
void PairLJClass2CoulLong::coeff(int narg, char **arg) void PairLJClass2CoulLong::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");
if (!allocated) allocate(); if (!allocated) allocate();
@ -299,7 +299,7 @@ void PairLJClass2CoulLong::init_style()
error->all(FLERR,"Pair style requires a KSpace style"); error->all(FLERR,"Pair style requires a KSpace style");
g_ewald = force->kspace->g_ewald; g_ewald = force->kspace->g_ewald;
// setup force tables // setup force tables
if (ncoultablebits) init_tables(cut_coul,NULL); if (ncoultablebits) init_tables(cut_coul,NULL);
} }

View File

@ -15,8 +15,8 @@
Contributing authors: Jeremy Lechman (SNL) Contributing authors: Jeremy Lechman (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "string.h" #include <string.h>
#include "fix_wall_colloid.h" #include "fix_wall_colloid.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"

View File

@ -15,10 +15,10 @@
Contributing author: Pieter in 't Veld (SNL) Contributing author: Pieter in 't Veld (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_colloid.h" #include "pair_colloid.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"
@ -144,7 +144,7 @@ void PairColloid::compute(int eflag, int vflag)
evdwl = 2.0/9.0*fR * evdwl = 2.0/9.0*fR *
(1.0-(K[1]*(K[1]*(K[1]/3.0+3.0*K[2])+4.2*K[4])+K[2]*K[4]) * (1.0-(K[1]*(K[1]*(K[1]/3.0+3.0*K[2])+4.2*K[4])+K[2]*K[4]) *
sigma6[itype][jtype]/K[6]) - offset[itype][jtype]; sigma6[itype][jtype]/K[6]) - offset[itype][jtype];
if (rsq <= K[1]) if (rsq <= K[1])
error->one(FLERR,"Overlapping small/large in pair colloid"); error->one(FLERR,"Overlapping small/large in pair colloid");
break; break;
@ -183,7 +183,7 @@ void PairColloid::compute(int eflag, int vflag)
if (eflag) if (eflag)
evdwl += a12[itype][jtype]/6.0 * evdwl += a12[itype][jtype]/6.0 *
(2.0*K[0]*(K[7]+K[8])-log(K[8]/K[7])) - offset[itype][jtype]; (2.0*K[0]*(K[7]+K[8])-log(K[8]/K[7])) - offset[itype][jtype];
if (r <= K[1]) if (r <= K[1])
error->one(FLERR,"Overlapping large/large in pair colloid"); error->one(FLERR,"Overlapping large/large in pair colloid");
break; break;
} }

View File

@ -15,8 +15,8 @@
Contributing authors: Randy Schunk (Sandia) Contributing authors: Randy Schunk (Sandia)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_yukawa_colloid.h" #include "pair_yukawa_colloid.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"

View File

@ -16,10 +16,10 @@
(hendrik.heenen at mytum.com) (hendrik.heenen at mytum.com)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "mpi.h" #include <mpi.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "math.h" #include <math.h>
#include "compute_temp_cs.h" #include "compute_temp_cs.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -57,12 +57,12 @@ ComputeTempCS::ComputeTempCS(LAMMPS *lmp, int narg, char **arg) :
// find and define groupbits for core and shell groups // find and define groupbits for core and shell groups
cgroup = group->find(arg[3]); cgroup = group->find(arg[3]);
if (cgroup == -1) if (cgroup == -1)
error->all(FLERR,"Cannot find specified group ID for core particles"); error->all(FLERR,"Cannot find specified group ID for core particles");
groupbit_c = group->bitmask[cgroup]; groupbit_c = group->bitmask[cgroup];
sgroup = group->find(arg[4]); sgroup = group->find(arg[4]);
if (sgroup == -1) if (sgroup == -1)
error->all(FLERR,"Cannot find specified group ID for shell particles"); error->all(FLERR,"Cannot find specified group ID for shell particles");
groupbit_s = group->bitmask[sgroup]; groupbit_s = group->bitmask[sgroup];
@ -228,9 +228,9 @@ double ComputeTempCS::compute_scalar()
vcm_pairs(); vcm_pairs();
// calculate thermal scalar in respect to atom velocities as center-of-mass // calculate thermal scalar in respect to atom velocities as center-of-mass
// velocities of its according core/shell pairs // velocities of its according core/shell pairs
double **v = atom->v; double **v = atom->v;
int *mask = atom->mask; int *mask = atom->mask;
int *type = atom->type; int *type = atom->type;
@ -256,12 +256,12 @@ double ComputeTempCS::compute_scalar()
MPI_Allreduce(&t,&scalar,1,MPI_DOUBLE,MPI_SUM,world); MPI_Allreduce(&t,&scalar,1,MPI_DOUBLE,MPI_SUM,world);
if (dynamic) dof_compute(); if (dynamic) dof_compute();
if (dof < 0.0 && natoms_temp > 0.0) if (dof < 0.0 && natoms_temp > 0.0)
error->all(FLERR,"Temperature compute degrees of freedom < 0"); error->all(FLERR,"Temperature compute degrees of freedom < 0");
scalar *= tfactor; scalar *= tfactor;
return scalar; return scalar;
} }
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void ComputeTempCS::compute_vector() void ComputeTempCS::compute_vector()
@ -313,7 +313,7 @@ void ComputeTempCS::vcm_pairs()
maxatom = atom->nmax; maxatom = atom->nmax;
memory->create(vint,maxatom,3,"temp/cs:vint"); memory->create(vint,maxatom,3,"temp/cs:vint");
} }
// vcm = COM velocity of each CS pair // vcm = COM velocity of each CS pair
// vint = internal velocity of each C/S atom, used as bias // vint = internal velocity of each C/S atom, used as bias
@ -327,7 +327,7 @@ void ComputeTempCS::vcm_pairs()
tagint partnerID; tagint partnerID;
for (i = 0; i < nlocal; i++) { for (i = 0; i < nlocal; i++) {
if ((mask[i] & groupbit) && if ((mask[i] & groupbit) &&
(mask[i] & groupbit_c || mask[i] & groupbit_s)) { (mask[i] & groupbit_c || mask[i] & groupbit_s)) {
if (rmass) massone = rmass[i]; if (rmass) massone = rmass[i];
else massone = mass[type[i]]; else massone = mass[type[i]];

View File

@ -15,10 +15,10 @@
Contributing author: Hendrik Heenen (hendrik.heenen@mytum.de) Contributing author: Hendrik Heenen (hendrik.heenen@mytum.de)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_born_coul_long_cs.h" #include "pair_born_coul_long_cs.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"
@ -118,8 +118,8 @@ void PairBornCoulLongCS::compute(int eflag, int vflag)
r = sqrt(rsq); r = sqrt(rsq);
prefactor = qqrd2e * qtmp*q[j]; prefactor = qqrd2e * qtmp*q[j];
if (factor_coul < 1.0) { if (factor_coul < 1.0) {
// When bonded parts are being calculated a minimal distance (EPS_EWALD) // When bonded parts are being calculated a minimal distance (EPS_EWALD)
// has to be added to the prefactor and erfc in order to make the // has to be added to the prefactor and erfc in order to make the
// used approximation functions valid // used approximation functions valid
grij = g_ewald * (r+EPS_EWALD); grij = g_ewald * (r+EPS_EWALD);
expm2 = exp(-grij*grij); expm2 = exp(-grij*grij);

View File

@ -15,10 +15,10 @@
Contributing author: Hendrik Heenen (hendrik.heenen@mytum.de) Contributing author: Hendrik Heenen (hendrik.heenen@mytum.de)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_buck_coul_long_cs.h" #include "pair_buck_coul_long_cs.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"
@ -118,8 +118,8 @@ void PairBuckCoulLongCS::compute(int eflag, int vflag)
r = sqrt(rsq); r = sqrt(rsq);
prefactor = qqrd2e * qtmp*q[j]; prefactor = qqrd2e * qtmp*q[j];
if (factor_coul < 1.0) { if (factor_coul < 1.0) {
// When bonded parts are being calculated a minimal distance (EPS_EWALD) // When bonded parts are being calculated a minimal distance (EPS_EWALD)
// has to be added to the prefactor and erfc in order to make the // has to be added to the prefactor and erfc in order to make the
// used approximation functions for the Ewald correction valid // used approximation functions for the Ewald correction valid
grij = g_ewald * (r+EPS_EWALD); grij = g_ewald * (r+EPS_EWALD);
expm2 = exp(-grij*grij); expm2 = exp(-grij*grij);

View File

@ -15,10 +15,10 @@
Contributing author: Hendrik Heenen (hendrik.heenen@mytum.de) Contributing author: Hendrik Heenen (hendrik.heenen@mytum.de)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_coul_long_cs.h" #include "pair_coul_long_cs.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"
@ -116,8 +116,8 @@ void PairCoulLongCS::compute(int eflag, int vflag)
r = sqrt(rsq); r = sqrt(rsq);
prefactor = qqrd2e * scale[itype][jtype] * qtmp*q[j]; prefactor = qqrd2e * scale[itype][jtype] * qtmp*q[j];
if (factor_coul < 1.0) { if (factor_coul < 1.0) {
// When bonded parts are being calculated a minimal distance (EPS_EWALD) // When bonded parts are being calculated a minimal distance (EPS_EWALD)
// has to be added to the prefactor and erfc in order to make the // has to be added to the prefactor and erfc in order to make the
// used approximation functions for the Ewald correction valid // used approximation functions for the Ewald correction valid
grij = g_ewald * (r+EPS_EWALD); grij = g_ewald * (r+EPS_EWALD);
expm2 = exp(-grij*grij); expm2 = exp(-grij*grij);

View File

@ -15,10 +15,10 @@
Contributing author: Hendrik Heenen (hendrik.heenen@mytum.de) Contributing author: Hendrik Heenen (hendrik.heenen@mytum.de)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_lj_cut_coul_long_cs.h" #include "pair_lj_cut_coul_long_cs.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"
@ -124,8 +124,8 @@ void PairLJCutCoulLongCS::compute(int eflag, int vflag)
r = sqrt(rsq); r = sqrt(rsq);
prefactor = qqrd2e * qtmp*q[j]; prefactor = qqrd2e * qtmp*q[j];
if (factor_coul < 1.0) { if (factor_coul < 1.0) {
// When bonded parts are being calculated a minimal distance (EPS_EWALD) // When bonded parts are being calculated a minimal distance (EPS_EWALD)
// has to be added to the prefactor and erfc in order to make the // has to be added to the prefactor and erfc in order to make the
// used approximation functions for the Ewald correction valid // used approximation functions for the Ewald correction valid
grij = g_ewald * (r+EPS_EWALD); grij = g_ewald * (r+EPS_EWALD);
expm2 = exp(-grij*grij); expm2 = exp(-grij*grij);

View File

@ -11,8 +11,8 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdlib.h" #include <stdlib.h>
#include "atom_vec_dipole.h" #include "atom_vec_dipole.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"

View File

@ -11,8 +11,8 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cut_dipole_cut.h" #include "pair_lj_cut_dipole_cut.h"
#include "atom.h" #include "atom.h"
#include "neighbor.h" #include "neighbor.h"
@ -22,7 +22,7 @@
#include "memory.h" #include "memory.h"
#include "error.h" #include "error.h"
#include "update.h" #include "update.h"
#include "string.h" #include <string.h>
using namespace LAMMPS_NS; using namespace LAMMPS_NS;

View File

@ -5,16 +5,16 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_lj_cut_dipole_long.h" #include "pair_lj_cut_dipole_long.h"
#include "atom.h" #include "atom.h"
#include "comm.h" #include "comm.h"
@ -26,7 +26,7 @@
#include "memory.h" #include "memory.h"
#include "error.h" #include "error.h"
#include "update.h" #include "update.h"
#include "string.h" #include <string.h>
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -50,7 +50,7 @@ PairLJCutDipoleLong::PairLJCutDipoleLong(LAMMPS *lmp) : Pair(lmp)
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
free all arrays free all arrays
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
PairLJCutDipoleLong::~PairLJCutDipoleLong() PairLJCutDipoleLong::~PairLJCutDipoleLong()
@ -153,7 +153,7 @@ void PairLJCutDipoleLong::compute(int eflag, int vflag)
pdotp = mu[i][0]*mu[j][0] + mu[i][1]*mu[j][1] + mu[i][2]*mu[j][2]; pdotp = mu[i][0]*mu[j][0] + mu[i][1]*mu[j][1] + mu[i][2]*mu[j][2];
pidotr = mu[i][0]*delx + mu[i][1]*dely + mu[i][2]*delz; pidotr = mu[i][0]*delx + mu[i][1]*dely + mu[i][2]*delz;
pjdotr = mu[j][0]*delx + mu[j][1]*dely + mu[j][2]*delz; pjdotr = mu[j][0]*delx + mu[j][1]*dely + mu[j][2]*delz;
g0 = qtmp*q[j]; g0 = qtmp*q[j];
g1 = qtmp*pjdotr - q[j]*pidotr + pdotp; g1 = qtmp*pjdotr - q[j]*pidotr + pdotp;
g2 = -pidotr*pjdotr; g2 = -pidotr*pjdotr;
@ -166,10 +166,10 @@ void PairLJCutDipoleLong::compute(int eflag, int vflag)
g0b1_g1b2_g2b3 = g0*b1 + g1*b2 + g2*b3; g0b1_g1b2_g2b3 = g0*b1 + g1*b2 + g2*b3;
fdx = delx * g0b1_g1b2_g2b3 - fdx = delx * g0b1_g1b2_g2b3 -
b1 * (qtmp*mu[j][0] - q[j]*mu[i][0]) + b1 * (qtmp*mu[j][0] - q[j]*mu[i][0]) +
b2 * (pjdotr*mu[i][0] + pidotr*mu[j][0]); b2 * (pjdotr*mu[i][0] + pidotr*mu[j][0]);
fdy = dely * g0b1_g1b2_g2b3 - fdy = dely * g0b1_g1b2_g2b3 -
b1 * (qtmp*mu[j][1] - q[j]*mu[i][1]) + b1 * (qtmp*mu[j][1] - q[j]*mu[i][1]) +
b2 * (pjdotr*mu[i][1] + pidotr*mu[j][1]); b2 * (pjdotr*mu[i][1] + pidotr*mu[j][1]);
fdz = delz * g0b1_g1b2_g2b3 - fdz = delz * g0b1_g1b2_g2b3 -
b1 * (qtmp*mu[j][2] - q[j]*mu[i][2]) + b1 * (qtmp*mu[j][2] - q[j]*mu[i][2]) +
@ -204,13 +204,13 @@ void PairLJCutDipoleLong::compute(int eflag, int vflag)
d1 = (d0 + pre1*expm2) * r2inv; d1 = (d0 + pre1*expm2) * r2inv;
d2 = (3.0*d1 + pre2*expm2) * r2inv; d2 = (3.0*d1 + pre2*expm2) * r2inv;
d3 = (5.0*d2 + pre3*expm2) * r2inv; d3 = (5.0*d2 + pre3*expm2) * r2inv;
g0d1_g1d2_g2d3 = g0*d1 + g1*d2 + g2*d3; g0d1_g1d2_g2d3 = g0*d1 + g1*d2 + g2*d3;
fax = delx * g0d1_g1d2_g2d3 - fax = delx * g0d1_g1d2_g2d3 -
d1 * (qtmp*mu[j][0] - q[j]*mu[i][0]) + d1 * (qtmp*mu[j][0] - q[j]*mu[i][0]) +
d2 * (pjdotr*mu[i][0] + pidotr*mu[j][0]); d2 * (pjdotr*mu[i][0] + pidotr*mu[j][0]);
fay = dely * g0d1_g1d2_g2d3 - fay = dely * g0d1_g1d2_g2d3 -
d1 * (qtmp*mu[j][1] - q[j]*mu[i][1]) + d1 * (qtmp*mu[j][1] - q[j]*mu[i][1]) +
d2 * (pjdotr*mu[i][1] + pidotr*mu[j][1]); d2 * (pjdotr*mu[i][1] + pidotr*mu[j][1]);
faz = delz * g0d1_g1d2_g2d3 - faz = delz * g0d1_g1d2_g2d3 -
d1 * (qtmp*mu[j][2] - q[j]*mu[i][2]) + d1 * (qtmp*mu[j][2] - q[j]*mu[i][2]) +
@ -271,7 +271,7 @@ void PairLJCutDipoleLong::compute(int eflag, int vflag)
fx = qqrd2e*forcecoulx + delx*fforce; fx = qqrd2e*forcecoulx + delx*fforce;
fy = qqrd2e*forcecouly + dely*fforce; fy = qqrd2e*forcecouly + dely*fforce;
fz = qqrd2e*forcecoulz + delz*fforce; fz = qqrd2e*forcecoulz + delz*fforce;
// force & torque accumulation // force & torque accumulation
f[i][0] += fx; f[i][0] += fx;
@ -316,7 +316,7 @@ void PairLJCutDipoleLong::compute(int eflag, int vflag)
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
allocate all arrays allocate all arrays
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairLJCutDipoleLong::allocate() void PairLJCutDipoleLong::allocate()
@ -343,12 +343,12 @@ void PairLJCutDipoleLong::allocate()
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
global settings global settings
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
void PairLJCutDipoleLong::settings(int narg, char **arg) void PairLJCutDipoleLong::settings(int narg, char **arg)
{ {
if (narg < 1 || narg > 2) if (narg < 1 || narg > 2)
error->all(FLERR,"Incorrect args in pair_style command"); error->all(FLERR,"Incorrect args in pair_style command");
cut_lj_global = force->numeric(FLERR,arg[0]); cut_lj_global = force->numeric(FLERR,arg[0]);
@ -371,7 +371,7 @@ void PairLJCutDipoleLong::settings(int narg, char **arg)
void PairLJCutDipoleLong::coeff(int narg, char **arg) void PairLJCutDipoleLong::coeff(int narg, char **arg)
{ {
if (narg < 4 || narg > 5) if (narg < 4 || narg > 5)
error->all(FLERR,"Incorrect args for pair coefficients"); error->all(FLERR,"Incorrect args for pair coefficients");
if (!allocated) allocate(); if (!allocated) allocate();
@ -419,12 +419,12 @@ double PairLJCutDipoleLong::init_one(int i, int j)
lj2[i][j] = 24.0 * epsilon[i][j] * pow(sigma[i][j],6.0); lj2[i][j] = 24.0 * epsilon[i][j] * pow(sigma[i][j],6.0);
lj3[i][j] = 4.0 * epsilon[i][j] * pow(sigma[i][j],12.0); lj3[i][j] = 4.0 * epsilon[i][j] * pow(sigma[i][j],12.0);
lj4[i][j] = 4.0 * epsilon[i][j] * pow(sigma[i][j],6.0); lj4[i][j] = 4.0 * epsilon[i][j] * pow(sigma[i][j],6.0);
if (offset_flag) { if (offset_flag) {
double ratio = sigma[i][j] / cut_lj[i][j]; double ratio = sigma[i][j] / cut_lj[i][j];
offset[i][j] = 4.0 * epsilon[i][j] * (pow(ratio,12.0) - pow(ratio,6.0)); offset[i][j] = 4.0 * epsilon[i][j] * (pow(ratio,12.0) - pow(ratio,6.0));
} else offset[i][j] = 0.0; } else offset[i][j] = 0.0;
cut_ljsq[j][i] = cut_ljsq[i][j]; cut_ljsq[j][i] = cut_ljsq[i][j];
lj1[j][i] = lj1[i][j]; lj1[j][i] = lj1[i][j];
lj2[j][i] = lj2[i][j]; lj2[j][i] = lj2[i][j];
@ -449,7 +449,7 @@ void PairLJCutDipoleLong::init_style()
// insure use of KSpace long-range solver, set g_ewald // insure use of KSpace long-range solver, set g_ewald
if (force->kspace == NULL) if (force->kspace == NULL)
error->all(FLERR,"Pair style requires a KSpace style"); error->all(FLERR,"Pair style requires a KSpace style");
g_ewald = force->kspace->g_ewald; g_ewald = force->kspace->g_ewald;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -15,10 +15,10 @@
Contributing author: Pieter J. in 't Veld and Stan Moore (Sandia) Contributing author: Pieter J. in 't Veld and Stan Moore (Sandia)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "math_const.h" #include "math_const.h"
#include "math_vector.h" #include "math_vector.h"
#include "pair_lj_long_dipole_long.h" #include "pair_lj_long_dipole_long.h"
@ -173,7 +173,7 @@ void *PairLJLongDipoleLong::extract(const char *id, int &dim)
"B", "sigma", "epsilon", "ewald_order", "ewald_cut", "ewald_mix", "B", "sigma", "epsilon", "ewald_order", "ewald_cut", "ewald_mix",
"cut_coul", "cut_vdwl", NULL}; "cut_coul", "cut_vdwl", NULL};
void *ptrs[] = { void *ptrs[] = {
lj4, sigma, epsilon, &ewald_order, &cut_coul, &mix_flag, &cut_coul, lj4, sigma, epsilon, &ewald_order, &cut_coul, &mix_flag, &cut_coul,
&cut_lj_global, NULL}; &cut_lj_global, NULL};
int i; int i;
@ -189,7 +189,7 @@ void *PairLJLongDipoleLong::extract(const char *id, int &dim)
void PairLJLongDipoleLong::coeff(int narg, char **arg) void PairLJLongDipoleLong::coeff(int narg, char **arg)
{ {
if (narg < 4 || narg > 5) if (narg < 4 || narg > 5)
error->all(FLERR,"Incorrect args for pair coefficients"); error->all(FLERR,"Incorrect args for pair coefficients");
if (!allocated) allocate(); if (!allocated) allocate();
@ -247,14 +247,14 @@ void PairLJLongDipoleLong::init_style()
// ensure use of KSpace long-range solver, set g_ewald // ensure use of KSpace long-range solver, set g_ewald
if (ewald_order&(1<<3)) { // r^-1 kspace if (ewald_order&(1<<3)) { // r^-1 kspace
if (force->kspace == NULL) if (force->kspace == NULL)
error->all(FLERR,"Pair style requires a KSpace style"); error->all(FLERR,"Pair style requires a KSpace style");
for (i=0; style3[i]&&strcmp(force->kspace_style, style3[i]); ++i); for (i=0; style3[i]&&strcmp(force->kspace_style, style3[i]); ++i);
if (!style3[i]) if (!style3[i])
error->all(FLERR,"Pair style requires use of kspace_style ewald/disp"); error->all(FLERR,"Pair style requires use of kspace_style ewald/disp");
} }
if (ewald_order&(1<<6)) { // r^-6 kspace if (ewald_order&(1<<6)) { // r^-6 kspace
if (force->kspace == NULL) if (force->kspace == NULL)
error->all(FLERR,"Pair style requires a KSpace style"); error->all(FLERR,"Pair style requires a KSpace style");
for (i=0; style6[i]&&strcmp(force->kspace_style, style6[i]); ++i); for (i=0; style6[i]&&strcmp(force->kspace_style, style6[i]); ++i);
if (!style6[i]) if (!style6[i])
@ -313,7 +313,7 @@ double PairLJLongDipoleLong::init_one(int i, int j)
//if (cut_respa && MIN(cut_lj[i][j],cut_coul) < cut_respa[3]) //if (cut_respa && MIN(cut_lj[i][j],cut_coul) < cut_respa[3])
//error->all(FLERR,"Pair cutoff < Respa interior cutoff"); //error->all(FLERR,"Pair cutoff < Respa interior cutoff");
if (offset_flag) { if (offset_flag) {
double ratio = sigma[i][j] / cut_lj[i][j]; double ratio = sigma[i][j] / cut_lj[i][j];
offset[i][j] = 4.0 * epsilon[i][j] * (pow(ratio,12.0) - pow(ratio,6.0)); offset[i][j] = 4.0 * epsilon[i][j] * (pow(ratio,12.0) - pow(ratio,6.0));
@ -423,7 +423,7 @@ void PairLJLongDipoleLong::compute(int eflag, int vflag)
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
double **x = atom->x, *x0 = x[0]; double **x = atom->x, *x0 = x[0];
double **mu = atom->mu, *mu0 = mu[0], *imu, *jmu; double **mu = atom->mu, *mu0 = mu[0], *imu, *jmu;
double **tq = atom->torque, *tq0 = tq[0], *tqi; double **tq = atom->torque, *tq0 = tq[0], *tqi;
@ -445,7 +445,7 @@ void PairLJLongDipoleLong::compute(int eflag, int vflag)
double B0, B1, B2, B3, G0, G1, G2, mudi, mudj, muij; double B0, B1, B2, B3, G0, G1, G2, mudi, mudj, muij;
vector force_d = VECTOR_NULL, ti = VECTOR_NULL, tj = VECTOR_NULL; vector force_d = VECTOR_NULL, ti = VECTOR_NULL, tj = VECTOR_NULL;
vector mui, muj, xi, d; vector mui, muj, xi, d;
double C1 = 2.0 * g_ewald / MY_PIS; double C1 = 2.0 * g_ewald / MY_PIS;
double C2 = 2.0 * g2 * C1; double C2 = 2.0 * g2 * C1;
double C3 = 2.0 * g2 * C2; double C3 = 2.0 * g2 * C2;
@ -460,14 +460,14 @@ void PairLJLongDipoleLong::compute(int eflag, int vflag)
cutsqi = cutsq[typei]; cut_ljsqi = cut_ljsq[typei]; cutsqi = cutsq[typei]; cut_ljsqi = cut_ljsq[typei];
memcpy(xi, x0+(i+(i<<1)), sizeof(vector)); memcpy(xi, x0+(i+(i<<1)), sizeof(vector));
memcpy(mui, imu = mu0+(i<<2), sizeof(vector)); memcpy(mui, imu = mu0+(i<<2), sizeof(vector));
jneighn = (jneigh = list->firstneigh[i])+list->numneigh[i]; jneighn = (jneigh = list->firstneigh[i])+list->numneigh[i];
for (; jneigh<jneighn; ++jneigh) { // loop over neighbors for (; jneigh<jneighn; ++jneigh) { // loop over neighbors
j = *jneigh; j = *jneigh;
ni = sbmask(j); // special index ni = sbmask(j); // special index
j &= NEIGHMASK; j &= NEIGHMASK;
{ register double *xj = x0+(j+(j<<1)); { register double *xj = x0+(j+(j<<1));
d[0] = xi[0] - xj[0]; // pair vector d[0] = xi[0] - xj[0]; // pair vector
d[1] = xi[1] - xj[1]; d[1] = xi[1] - xj[1];
@ -496,7 +496,7 @@ void PairLJLongDipoleLong::compute(int eflag, int vflag)
G1 = qi*mudj-qj*mudi+muij; G1 = qi*mudj-qj*mudi+muij;
G2 = -mudi*mudj; G2 = -mudi*mudj;
force_coul = G0*B1+G1*B2+G2*B3; force_coul = G0*B1+G1*B2+G2*B3;
mudi *= B2; mudj *= B2; // torque contribs mudi *= B2; mudj *= B2; // torque contribs
ti[0] = mudj*d[0]+(qj*d[0]-muj[0])*B1; ti[0] = mudj*d[0]+(qj*d[0]-muj[0])*B1;
ti[1] = mudj*d[1]+(qj*d[1]-muj[1])*B1; ti[1] = mudj*d[1]+(qj*d[1]-muj[1])*B1;
@ -552,7 +552,7 @@ void PairLJLongDipoleLong::compute(int eflag, int vflag)
register double f = special_lj[ni], t = rn*(1.0-f); register double f = special_lj[ni], t = rn*(1.0-f);
force_lj = f*(rn *= rn)*lj1i[typej]- force_lj = f*(rn *= rn)*lj1i[typej]-
g8*(((6.0*a2+6.0)*a2+3.0)*a2+1.0)*x2*rsq+t*lj2i[typej]; g8*(((6.0*a2+6.0)*a2+3.0)*a2+1.0)*x2*rsq+t*lj2i[typej];
if (eflag) evdwl = if (eflag) evdwl =
f*rn*lj3i[typej]-g6*((a2+1.0)*a2+0.5)*x2+t*lj4i[typej]; f*rn*lj3i[typej]-g6*((a2+1.0)*a2+0.5)*x2+t*lj4i[typej];
} }
} }
@ -641,8 +641,8 @@ double PairLJLongDipoleLong::single(int i, int j, int itype, int jtype,
G1 = qi*mudj-qj*mudi+muij; G1 = qi*mudj-qj*mudi+muij;
G2 = -mudi*mudj; G2 = -mudi*mudj;
force_coul = G0*B1+G1*B2+G2*B3; force_coul = G0*B1+G1*B2+G2*B3;
eng += G0*B0+G1*B1+G2*B2; eng += G0*B0+G1*B1+G2*B2;
if (factor_coul < 1.0) { // adj part, eqn 2.13 if (factor_coul < 1.0) { // adj part, eqn 2.13
force_coul -= (f = force->qqrd2e*(1.0-factor_coul)/r)*( force_coul -= (f = force->qqrd2e*(1.0-factor_coul)/r)*(
(3.0*G1+6.0*muij+15.0*G2*r2inv)*r2inv+G0); (3.0*G1+6.0*muij+15.0*G2*r2inv)*r2inv+G0);
@ -672,7 +672,7 @@ double PairLJLongDipoleLong::single(int i, int j, int itype, int jtype,
eng += factor_lj*(r6inv*(r6inv*lj3[itype][jtype]- eng += factor_lj*(r6inv*(r6inv*lj3[itype][jtype]-
lj4[itype][jtype])-offset[itype][jtype]); lj4[itype][jtype])-offset[itype][jtype]);
} }
} }
else force_lj = 0.0; else force_lj = 0.0;
fforce = (force_coul+force_lj)*r2inv; fforce = (force_coul+force_lj)*r2inv;

View File

@ -15,10 +15,10 @@
Contributing authors: Amit Kumar and Michael Bybee (UIUC) Contributing authors: Amit Kumar and Michael Bybee (UIUC)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_brownian.h" #include "pair_brownian.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -502,7 +502,7 @@ void PairBrownian::init_style()
if (strcmp(modify->fix[i]->style,"deform") == 0) if (strcmp(modify->fix[i]->style,"deform") == 0)
flagdeform = 1; flagdeform = 1;
else if (strstr(modify->fix[i]->style,"wall") != NULL) { else if (strstr(modify->fix[i]->style,"wall") != NULL) {
if (flagwall) if (flagwall)
error->all(FLERR, error->all(FLERR,
"Cannot use multiple fix wall commands with pair brownian"); "Cannot use multiple fix wall commands with pair brownian");
flagwall = 1; // Walls exist flagwall = 1; // Walls exist

View File

@ -16,10 +16,10 @@
Dave Heine (Corning), polydispersity Dave Heine (Corning), polydispersity
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_brownian_poly.h" #include "pair_brownian_poly.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -365,7 +365,7 @@ void PairBrownianPoly::init_style()
if (strcmp(modify->fix[i]->style,"deform") == 0) if (strcmp(modify->fix[i]->style,"deform") == 0)
flagdeform = 1; flagdeform = 1;
else if (strstr(modify->fix[i]->style,"wall") != NULL) { else if (strstr(modify->fix[i]->style,"wall") != NULL) {
if (flagwall) if (flagwall)
error->all(FLERR, error->all(FLERR,
"Cannot use multiple fix wall commands with pair brownian"); "Cannot use multiple fix wall commands with pair brownian");
flagwall = 1; // Walls exist flagwall = 1; // Walls exist

View File

@ -16,10 +16,10 @@
Amit Kumar and Michael Bybee (UIUC) Amit Kumar and Michael Bybee (UIUC)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_lubricate.h" #include "pair_lubricate.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -575,7 +575,7 @@ void PairLubricate::init_style()
"fix deform remap option"); "fix deform remap option");
} }
if (strstr(modify->fix[i]->style,"wall") != NULL) { if (strstr(modify->fix[i]->style,"wall") != NULL) {
if (flagwall) if (flagwall)
error->all(FLERR, error->all(FLERR,
"Cannot use multiple fix wall commands with pair lubricate"); "Cannot use multiple fix wall commands with pair lubricate");
flagwall = 1; // Walls exist flagwall = 1; // Walls exist

View File

@ -15,11 +15,11 @@
Contributing authors: Amit Kumar and Michael Bybee (UIUC) Contributing authors: Amit Kumar and Michael Bybee (UIUC)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "mpi.h" #include <mpi.h>
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_lubricateU.h" #include "pair_lubricateU.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -1800,7 +1800,7 @@ void PairLubricateU::init_style()
if (strcmp(modify->fix[i]->style,"deform") == 0) if (strcmp(modify->fix[i]->style,"deform") == 0)
flagdeform = 1; flagdeform = 1;
else if (strstr(modify->fix[i]->style,"wall") != NULL) { else if (strstr(modify->fix[i]->style,"wall") != NULL) {
if (flagwall) if (flagwall)
error->all(FLERR, error->all(FLERR,
"Cannot use multiple fix wall commands with " "Cannot use multiple fix wall commands with "
"pair lubricateU"); "pair lubricateU");

View File

@ -17,11 +17,11 @@
Dave Heine (Corning), polydispersity Dave Heine (Corning), polydispersity
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "mpi.h" #include <mpi.h>
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_lubricateU_poly.h" #include "pair_lubricateU_poly.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -1153,7 +1153,7 @@ void PairLubricateUPoly::init_style()
if (strcmp(modify->fix[i]->style,"deform") == 0) if (strcmp(modify->fix[i]->style,"deform") == 0)
flagdeform = 1; flagdeform = 1;
else if (strstr(modify->fix[i]->style,"wall") != NULL){ else if (strstr(modify->fix[i]->style,"wall") != NULL){
if (flagwall) if (flagwall)
error->all(FLERR, error->all(FLERR,
"Cannot use multiple fix wall commands with " "Cannot use multiple fix wall commands with "
"pair lubricateU"); "pair lubricateU");

View File

@ -17,10 +17,10 @@
Dave Heine (Corning), polydispersity Dave Heine (Corning), polydispersity
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_lubricate_poly.h" #include "pair_lubricate_poly.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -471,7 +471,7 @@ void PairLubricatePoly::init_style()
"fix deform remap option"); "fix deform remap option");
} }
if (strstr(modify->fix[i]->style,"wall") != NULL) { if (strstr(modify->fix[i]->style,"wall") != NULL) {
if (flagwall) if (flagwall)
error->all(FLERR, error->all(FLERR,
"Cannot use multiple fix wall commands with " "Cannot use multiple fix wall commands with "
"pair lubricate/poly"); "pair lubricate/poly");

View File

@ -11,8 +11,8 @@
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "string.h" #include <string.h>
#include "stdlib.h" #include <stdlib.h>
#include "fix_gpu.h" #include "fix_gpu.h"
#include "atom.h" #include "atom.h"
#include "force.h" #include "force.h"
@ -40,7 +40,7 @@ extern int lmp_init_device(MPI_Comm world, MPI_Comm replica,
const int first_gpu, const int last_gpu, const int first_gpu, const int last_gpu,
const int gpu_mode, const double particle_split, const int gpu_mode, const double particle_split,
const int nthreads, const int t_per_atom, const int nthreads, const int t_per_atom,
const double cell_size, char *opencl_flags, const double cell_size, char *opencl_flags,
const int block_pair); const int block_pair);
extern void lmp_clear_device(); extern void lmp_clear_device();
extern double lmp_gpu_forces(double **f, double **tor, double *eatom, extern double lmp_gpu_forces(double **f, double **tor, double *eatom,
@ -89,7 +89,7 @@ FixGPU::FixGPU(LAMMPS *lmp, int narg, char **arg) :
if (ngpu <= 0) error->all(FLERR,"Illegal package gpu command"); if (ngpu <= 0) error->all(FLERR,"Illegal package gpu command");
int first_gpu = 0; int first_gpu = 0;
int last_gpu = ngpu-1; int last_gpu = ngpu-1;
// options // options
_gpu_mode = GPU_NEIGH; _gpu_mode = GPU_NEIGH;
@ -172,7 +172,7 @@ FixGPU::FixGPU(LAMMPS *lmp, int narg, char **arg) :
if (binsize == 0.0) binsize = -1.0; if (binsize == 0.0) binsize = -1.0;
int gpu_flag = lmp_init_device(universe->uworld, world, first_gpu, last_gpu, int gpu_flag = lmp_init_device(universe->uworld, world, first_gpu, last_gpu,
_gpu_mode, _particle_split, nthreads, _gpu_mode, _particle_split, nthreads,
threads_per_atom, binsize, opencl_flags, threads_per_atom, binsize, opencl_flags,
block_pair); block_pair);
GPU_EXTRA::check_flag(gpu_flag,error,world); GPU_EXTRA::check_flag(gpu_flag,error,world);
} }
@ -200,8 +200,8 @@ int FixGPU::setmask()
void FixGPU::init() void FixGPU::init()
{ {
// GPU package cannot be used with atom_style template // GPU package cannot be used with atom_style template
if (atom->molecular == 2) if (atom->molecular == 2)
error->all(FLERR,"GPU package does not (yet) work with " error->all(FLERR,"GPU package does not (yet) work with "
"atom_style template"); "atom_style template");
@ -219,7 +219,7 @@ void FixGPU::init()
// neighbor list builds on the GPU with triclinic box is not yet supported // neighbor list builds on the GPU with triclinic box is not yet supported
if ((_gpu_mode == GPU_NEIGH || _gpu_mode == GPU_HYB_NEIGH) && if ((_gpu_mode == GPU_NEIGH || _gpu_mode == GPU_HYB_NEIGH) &&
domain->triclinic) domain->triclinic)
error->all(FLERR,"Cannot use package gpu neigh yes with triclinic box"); error->all(FLERR,"Cannot use package gpu neigh yes with triclinic box");
@ -229,7 +229,7 @@ void FixGPU::init()
error->warning(FLERR,"Using package gpu without any pair style defined"); error->warning(FLERR,"Using package gpu without any pair style defined");
// make sure fdotr virial is not accumulated multiple times // make sure fdotr virial is not accumulated multiple times
if (force->pair_match("hybrid",1) != NULL) { if (force->pair_match("hybrid",1) != NULL) {
PairHybrid *hybrid = (PairHybrid *) force->pair; PairHybrid *hybrid = (PairHybrid *) force->pair;
for (int i = 0; i < hybrid->nstyles; i++) for (int i = 0; i < hybrid->nstyles; i++)

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_beck_gpu.h" #include "pair_beck_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
#include "math_special.h" #include "math_special.h"
@ -54,7 +54,7 @@ int ** beck_gpu_compute_n(const int ago, const int inum,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, int **ilist, int **jnum,
const double cpu_time, bool &success); const double cpu_time, bool &success);
void beck_gpu_compute(const int ago, const int inum, const int nall, void beck_gpu_compute(const int ago, const int inum, 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,
@ -68,7 +68,7 @@ PairBeckGPU::PairBeckGPU(LAMMPS *lmp) : PairBeck(lmp), gpu_mode(GPU_FORCE)
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -86,10 +86,10 @@ void PairBeckGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -98,7 +98,7 @@ void PairBeckGPU::compute(int eflag, int vflag)
atom->x, atom->type, domain->sublo, atom->x, atom->type, domain->sublo,
domain->subhi, atom->tag, atom->nspecial, domain->subhi, atom->tag, atom->nspecial,
atom->special, eflag, vflag, eflag_atom, atom->special, eflag, vflag, eflag_atom,
vflag_atom, host_start, vflag_atom, host_start,
&ilist, &numneigh, cpu_time, success); &ilist, &numneigh, cpu_time, success);
} else { } else {
inum = list->inum; inum = list->inum;
@ -125,7 +125,7 @@ void PairBeckGPU::compute(int eflag, int vflag)
void PairBeckGPU::init_style() void PairBeckGPU::init_style()
{ {
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with beck/gpu pair style"); error->all(FLERR,"Cannot use newton pair with beck/gpu pair style");
// Repeat cutsq calculation because done after call to init_style // Repeat cutsq calculation because done after call to init_style
@ -171,7 +171,7 @@ double PairBeckGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairBeckGPU::cpu_compute(int start, int inum, int eflag, int vflag, void PairBeckGPU::cpu_compute(int start, int inum, int eflag, int vflag,
int *ilist, int *numneigh, int **firstneigh) { int *ilist, int *numneigh, int **firstneigh) {
int i,j,ii,jj,jnum,itype,jtype; int i,j,ii,jj,jnum,itype,jtype;
double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair; double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_born_coul_long_gpu.h" #include "pair_born_coul_long_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -33,7 +33,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "kspace.h" #include "kspace.h"
#include "gpu_extra.h" #include "gpu_extra.h"
@ -51,9 +51,9 @@ using namespace MathConst;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int borncl_gpu_init(const int ntypes, double **cutsq, double **host_rhoinv, int borncl_gpu_init(const int ntypes, double **cutsq, double **host_rhoinv,
double **host_born1, double **host_born2, double **host_born1, double **host_born2,
double **host_born3, double **host_a, double **host_born3, double **host_a,
double **host_c, double **host_d, double **host_c, double **host_d,
double **sigma, double **offset, double *special_lj, double **sigma, double **offset, double *special_lj,
const int inum, const int nall, const int max_nbors, const int inum, const int nall, const int max_nbors,
const int maxspecial, const double cell_size, const int maxspecial, const double cell_size,
@ -63,7 +63,7 @@ int borncl_gpu_init(const int ntypes, double **cutsq, double **host_rhoinv,
void borncl_gpu_clear(); void borncl_gpu_clear();
int** borncl_gpu_compute_n(const int ago, const int inum_full, const int nall, int** borncl_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,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, const double cpu_time, int **ilist, int **jnum, const double cpu_time,
@ -79,13 +79,13 @@ double borncl_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairBornCoulLongGPU::PairBornCoulLongGPU(LAMMPS *lmp) : PairBornCoulLongGPU::PairBornCoulLongGPU(LAMMPS *lmp) :
PairBornCoulLong(lmp), gpu_mode(GPU_FORCE) PairBornCoulLong(lmp), gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -103,12 +103,12 @@ void PairBornCoulLongGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
inum = atom->nlocal; inum = atom->nlocal;
firstneigh = borncl_gpu_compute_n(neighbor->ago, inum, nall, atom->x, firstneigh = borncl_gpu_compute_n(neighbor->ago, inum, nall, atom->x,
@ -147,7 +147,7 @@ void PairBornCoulLongGPU::init_style()
if (!atom->q_flag) if (!atom->q_flag)
error->all(FLERR, error->all(FLERR,
"Pair style born/coul/long/gpu requires atom attribute q"); "Pair style born/coul/long/gpu requires atom attribute q");
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR, error->all(FLERR,
"Cannot use newton pair with born/coul/long/gpu pair style"); "Cannot use newton pair with born/coul/long/gpu pair style");
@ -179,12 +179,12 @@ void PairBornCoulLongGPU::init_style()
int maxspecial=0; int maxspecial=0;
if (atom->molecular) if (atom->molecular)
maxspecial=atom->maxspecial; maxspecial=atom->maxspecial;
int success = borncl_gpu_init(atom->ntypes+1, cutsq, rhoinv, int success = borncl_gpu_init(atom->ntypes+1, cutsq, rhoinv,
born1, born2, born3, a, c, d, sigma, born1, born2, born3, a, c, d, sigma,
offset, force->special_lj, atom->nlocal, offset, force->special_lj, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
cell_size, gpu_mode, screen, cut_ljsq, cell_size, gpu_mode, screen, cut_ljsq,
cut_coulsq, force->special_coul, cut_coulsq, force->special_coul,
force->qqrd2e, g_ewald); force->qqrd2e, g_ewald);
GPU_EXTRA::check_flag(success,error,world); GPU_EXTRA::check_flag(success,error,world);
@ -284,7 +284,7 @@ void PairBornCoulLongGPU::cpu_compute(int start, int inum, int eflag,
if (factor_coul < 1.0) ecoul -= (1.0-factor_coul)*prefactor; if (factor_coul < 1.0) ecoul -= (1.0-factor_coul)*prefactor;
} else ecoul = 0.0; } else ecoul = 0.0;
if (rsq < cut_ljsq[itype][jtype]) { if (rsq < cut_ljsq[itype][jtype]) {
evdwl = a[itype][jtype]*rexp - c[itype][jtype]*r6inv evdwl = a[itype][jtype]*rexp - c[itype][jtype]*r6inv
+ d[itype][jtype]*r6inv*r2inv - offset[itype][jtype]; + d[itype][jtype]*r6inv*r2inv - offset[itype][jtype];
evdwl *= factor_lj; evdwl *= factor_lj;
} else evdwl = 0.0; } else evdwl = 0.0;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing authors: Trung Dac Nguyen (ORNL) Contributing authors: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_born_coul_wolf_gpu.h" #include "pair_born_coul_wolf_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -33,7 +33,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -42,9 +42,9 @@ using namespace MathConst;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int borncw_gpu_init(const int ntypes, double **cutsq, double **host_rhoinv, int borncw_gpu_init(const int ntypes, double **cutsq, double **host_rhoinv,
double **host_born1, double **host_born2, double **host_born1, double **host_born2,
double **host_born3, double **host_a, double **host_c, double **host_born3, double **host_a, double **host_c,
double **host_d, double **sigma, double **offset, double **host_d, double **sigma, double **offset,
double *special_lj, const int inum, double *special_lj, const int inum,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen, const double cell_size, int &gpu_mode, FILE *screen,
@ -54,7 +54,7 @@ int borncw_gpu_init(const int ntypes, double **cutsq, double **host_rhoinv,
void borncw_gpu_clear(); void borncw_gpu_clear();
int ** borncw_gpu_compute_n(const int ago, const int inum_full, const int nall, int ** borncw_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,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, const double cpu_time, int **ilist, int **jnum, const double cpu_time,
@ -70,13 +70,13 @@ double borncw_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairBornCoulWolfGPU::PairBornCoulWolfGPU(LAMMPS *lmp) : PairBornCoulWolf(lmp), PairBornCoulWolfGPU::PairBornCoulWolfGPU(LAMMPS *lmp) : PairBornCoulWolf(lmp),
gpu_mode(GPU_FORCE) gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -94,10 +94,10 @@ void PairBornCoulWolfGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -106,8 +106,8 @@ void PairBornCoulWolfGPU::compute(int eflag, int vflag)
atom->x, atom->type, domain->sublo, atom->x, atom->type, domain->sublo,
domain->subhi, atom->tag, atom->nspecial, domain->subhi, atom->tag, atom->nspecial,
atom->special, eflag, vflag, eflag_atom, atom->special, eflag, vflag, eflag_atom,
vflag_atom, host_start, vflag_atom, host_start,
&ilist, &numneigh, cpu_time, success, &ilist, &numneigh, cpu_time, success,
atom->q, domain->boxlo, domain->prd); atom->q, domain->boxlo, domain->prd);
} else { } else {
inum = list->inum; inum = list->inum;
@ -135,7 +135,7 @@ void PairBornCoulWolfGPU::compute(int eflag, int vflag)
void PairBornCoulWolfGPU::init_style() void PairBornCoulWolfGPU::init_style()
{ {
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR, error->all(FLERR,
"Cannot use newton pair with born/coul/wolf/gpu pair style"); "Cannot use newton pair with born/coul/wolf/gpu pair style");
@ -159,18 +159,18 @@ void PairBornCoulWolfGPU::init_style()
cut_coulsq = cut_coul * cut_coul; cut_coulsq = cut_coul * cut_coul;
double e_shift = erfc(alf*cut_coul)/cut_coul; double e_shift = erfc(alf*cut_coul)/cut_coul;
double f_shift = -(e_shift+ 2.0*alf/MY_PIS * exp(-alf*alf*cut_coul*cut_coul)) / double f_shift = -(e_shift+ 2.0*alf/MY_PIS * exp(-alf*alf*cut_coul*cut_coul)) /
cut_coul; cut_coul;
int maxspecial=0; int maxspecial=0;
if (atom->molecular) if (atom->molecular)
maxspecial=atom->maxspecial; maxspecial=atom->maxspecial;
int success = borncw_gpu_init(atom->ntypes+1, cutsq, rhoinv, int success = borncw_gpu_init(atom->ntypes+1, cutsq, rhoinv,
born1, born2, born3, a, c, d, sigma, offset, born1, born2, born3, a, c, d, sigma, offset,
force->special_lj, atom->nlocal, force->special_lj, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
cell_size, gpu_mode, screen, cut_ljsq, cell_size, gpu_mode, screen, cut_ljsq,
cut_coulsq, force->special_coul, force->qqrd2e, cut_coulsq, force->special_coul, force->qqrd2e,
alf, e_shift, f_shift); alf, e_shift, f_shift);
GPU_EXTRA::check_flag(success,error,world); GPU_EXTRA::check_flag(success,error,world);
@ -178,7 +178,7 @@ void PairBornCoulWolfGPU::init_style()
int irequest = neighbor->request(this,instance_me); int irequest = neighbor->request(this,instance_me);
neighbor->requests[irequest]->half = 0; neighbor->requests[irequest]->half = 0;
neighbor->requests[irequest]->full = 1; neighbor->requests[irequest]->full = 1;
} }
} }
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
@ -203,7 +203,7 @@ void PairBornCoulWolfGPU::cpu_compute(int start, int inum, int eflag, int vflag,
int *jlist; int *jlist;
evdwl = ecoul = 0.0; evdwl = ecoul = 0.0;
double **x = atom->x; double **x = atom->x;
double **f = atom->f; double **f = atom->f;
double *q = atom->q; double *q = atom->q;
@ -212,10 +212,10 @@ void PairBornCoulWolfGPU::cpu_compute(int start, int inum, int eflag, int vflag,
double *special_coul = force->special_coul; double *special_coul = force->special_coul;
double *special_lj = force->special_lj; double *special_lj = force->special_lj;
double qqrd2e = force->qqrd2e; double qqrd2e = force->qqrd2e;
double e_shift = erfc(alf*cut_coul)/cut_coul; double e_shift = erfc(alf*cut_coul)/cut_coul;
double f_shift = -(e_shift+ 2.0*alf/MY_PIS * exp(-alf*alf*cut_coul*cut_coul)) / double f_shift = -(e_shift+ 2.0*alf/MY_PIS * exp(-alf*alf*cut_coul*cut_coul)) /
cut_coul; cut_coul;
// loop over neighbors of my atoms // loop over neighbors of my atoms
@ -247,13 +247,13 @@ void PairBornCoulWolfGPU::cpu_compute(int start, int inum, int eflag, int vflag,
if (rsq < cutsq[itype][jtype]) { if (rsq < cutsq[itype][jtype]) {
r2inv = 1.0/rsq; r2inv = 1.0/rsq;
if (rsq < cut_coulsq) { if (rsq < cut_coulsq) {
r = sqrt(rsq); r = sqrt(rsq);
prefactor = qqrd2e*qtmp*q[j]/r; prefactor = qqrd2e*qtmp*q[j]/r;
erfcc = erfc(alf*r); erfcc = erfc(alf*r);
erfcd = exp(-alf*alf*r*r); erfcd = exp(-alf*alf*r*r);
v_sh = (erfcc - e_shift*r) * prefactor; v_sh = (erfcc - e_shift*r) * prefactor;
dvdrr = (erfcc/rsq + 2.0*alf/MY_PIS * erfcd/r) + f_shift; dvdrr = (erfcc/rsq + 2.0*alf/MY_PIS * erfcd/r) + f_shift;
forcecoul = dvdrr*rsq*prefactor; forcecoul = dvdrr*rsq*prefactor;
if (factor_coul < 1.0) forcecoul -= (1.0-factor_coul)*prefactor; if (factor_coul < 1.0) forcecoul -= (1.0-factor_coul)*prefactor;
@ -263,12 +263,12 @@ void PairBornCoulWolfGPU::cpu_compute(int start, int inum, int eflag, int vflag,
r6inv = r2inv*r2inv*r2inv; r6inv = r2inv*r2inv*r2inv;
r = sqrt(rsq); r = sqrt(rsq);
rexp = exp((sigma[itype][jtype]-r)*rhoinv[itype][jtype]); rexp = exp((sigma[itype][jtype]-r)*rhoinv[itype][jtype]);
forceborn = born1[itype][jtype]*r*rexp - born2[itype][jtype]*r6inv + forceborn = born1[itype][jtype]*r*rexp - born2[itype][jtype]*r6inv +
born3[itype][jtype]*r2inv*r6inv; born3[itype][jtype]*r2inv*r6inv;
} else forceborn = 0.0; } else forceborn = 0.0;
fpair = (factor_coul*forcecoul + factor_lj*forceborn) * r2inv; fpair = (factor_coul*forcecoul + factor_lj*forceborn) * r2inv;
f[i][0] += delx*fpair; f[i][0] += delx*fpair;
f[i][1] += dely*fpair; f[i][1] += dely*fpair;
f[i][2] += delz*fpair; f[i][2] += delz*fpair;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_born_gpu.h" #include "pair_born_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -40,19 +40,19 @@ using namespace LAMMPS_NS;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int born_gpu_init(const int ntypes, double **cutsq, double **host_rhoinv, int born_gpu_init(const int ntypes, double **cutsq, double **host_rhoinv,
double **host_born1, double **host_born2, double **host_born3, double **host_born1, double **host_born2, double **host_born3,
double **host_a, double **host_c, double **host_d, double **host_a, double **host_c, double **host_d,
double **host_sigma, double **offset, double *special_lj, double **host_sigma, double **offset, double *special_lj,
const int inum, const int nall, const int max_nbors, const int inum, const int nall, const int max_nbors,
const int maxspecial, const double cell_size, const int maxspecial, const double cell_size,
int &gpu_mode, FILE *screen); int &gpu_mode, FILE *screen);
void born_gpu_reinit(const int ntypes, double **host_rhoinv, void born_gpu_reinit(const int ntypes, double **host_rhoinv,
double **host_born1, double **host_born2, double **host_born3, double **host_born1, double **host_born2, double **host_born3,
double **host_a, double **host_c, double **host_d, double **host_a, double **host_c, double **host_d,
double **offset); double **offset);
void born_gpu_clear(); void born_gpu_clear();
int ** born_gpu_compute_n(const int ago, const int inum_full, int ** born_gpu_compute_n(const int ago, const int inum_full,
const int nall, double **host_x, int *host_type, const int nall, double **host_x, int *host_type,
double *sublo, double *subhi, tagint *tag, int **nspecial, double *sublo, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, tagint **special, const bool eflag, const bool vflag,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
@ -71,7 +71,7 @@ PairBornGPU::PairBornGPU(LAMMPS *lmp) : PairBorn(lmp), gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -89,10 +89,10 @@ void PairBornGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -101,7 +101,7 @@ void PairBornGPU::compute(int eflag, int vflag)
atom->x, atom->type, domain->sublo, atom->x, atom->type, domain->sublo,
domain->subhi, atom->tag, atom->nspecial, domain->subhi, atom->tag, atom->nspecial,
atom->special, eflag, vflag, eflag_atom, atom->special, eflag, vflag, eflag_atom,
vflag_atom, host_start, vflag_atom, host_start,
&ilist, &numneigh, cpu_time, success); &ilist, &numneigh, cpu_time, success);
} else { } else {
inum = list->inum; inum = list->inum;
@ -128,7 +128,7 @@ void PairBornGPU::compute(int eflag, int vflag)
void PairBornGPU::init_style() void PairBornGPU::init_style()
{ {
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with born/gpu pair style"); error->all(FLERR,"Cannot use newton pair with born/gpu pair style");
// Repeat cutsq calculation because done after call to init_style // Repeat cutsq calculation because done after call to init_style
@ -151,7 +151,7 @@ void PairBornGPU::init_style()
int maxspecial=0; int maxspecial=0;
if (atom->molecular) if (atom->molecular)
maxspecial=atom->maxspecial; maxspecial=atom->maxspecial;
int success = born_gpu_init(atom->ntypes+1, cutsq, rhoinv, int success = born_gpu_init(atom->ntypes+1, cutsq, rhoinv,
born1, born2, born3, a, c, d, sigma, born1, born2, born3, a, c, d, sigma,
offset, force->special_lj, atom->nlocal, offset, force->special_lj, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
@ -170,7 +170,7 @@ void PairBornGPU::init_style()
void PairBornGPU::reinit() void PairBornGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
born_gpu_reinit(atom->ntypes+1, rhoinv, born1, born2, born3, born_gpu_reinit(atom->ntypes+1, rhoinv, born1, born2, born3,
a, c, d, offset); a, c, d, offset);
} }
@ -185,7 +185,7 @@ double PairBornGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairBornGPU::cpu_compute(int start, int inum, int eflag, int vflag, void PairBornGPU::cpu_compute(int start, int inum, int eflag, int vflag,
int *ilist, int *numneigh, int **firstneigh) { int *ilist, int *numneigh, int **firstneigh) {
int i,j,ii,jj,jnum,itype,jtype; int i,j,ii,jj,jnum,itype,jtype;
double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair; double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair;
@ -225,7 +225,7 @@ void PairBornGPU::cpu_compute(int start, int inum, int eflag, int vflag,
r6inv = r2inv*r2inv*r2inv; r6inv = r2inv*r2inv*r2inv;
r = sqrt(rsq); r = sqrt(rsq);
rexp = exp((sigma[itype][jtype]-r)*rhoinv[itype][jtype]); rexp = exp((sigma[itype][jtype]-r)*rhoinv[itype][jtype]);
forceborn = born1[itype][jtype]*r*rexp - born2[itype][jtype]*r6inv + forceborn = born1[itype][jtype]*r*rexp - born2[itype][jtype]*r6inv +
born3[itype][jtype]*r2inv*r6inv; born3[itype][jtype]*r2inv*r6inv;
fpair = factor_lj*forceborn*r2inv; fpair = factor_lj*forceborn*r2inv;
@ -234,7 +234,7 @@ void PairBornGPU::cpu_compute(int start, int inum, int eflag, int vflag,
f[i][2] += delz*fpair; f[i][2] += delz*fpair;
if (eflag) { if (eflag) {
evdwl = a[itype][jtype]*rexp - c[itype][jtype]*r6inv + evdwl = a[itype][jtype]*rexp - c[itype][jtype]*r6inv +
d[itype][jtype]*r6inv*r2inv - offset[itype][jtype]; d[itype][jtype]*r6inv*r2inv - offset[itype][jtype];
evdwl *= factor_lj; evdwl *= factor_lj;
} }

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -15,9 +15,9 @@
Contributing authors: Trung Dac Nguyen (ORNL) Contributing authors: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_buck_coul_cut_gpu.h" #include "pair_buck_coul_cut_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;

View File

@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_buck_coul_long_gpu.h" #include "pair_buck_coul_long_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "kspace.h" #include "kspace.h"
#include "gpu_extra.h" #include "gpu_extra.h"

View File

@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_buck_gpu.h" #include "pair_buck_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -167,7 +167,7 @@ void PairBuckGPU::init_style()
void PairBuckGPU::reinit() void PairBuckGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
buck_gpu_reinit(atom->ntypes+1, cutsq, rhoinv, buck1, buck2, buck_gpu_reinit(atom->ntypes+1, cutsq, rhoinv, buck1, buck2,
a, c, offset); a, c, offset);
} }

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_colloid_gpu.h" #include "pair_colloid_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -40,22 +40,22 @@ using namespace LAMMPS_NS;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int colloid_gpu_init(const int ntypes, double **cutsq, double **host_lj1, int colloid_gpu_init(const int ntypes, double **cutsq, double **host_lj1,
double **host_lj2, double **host_lj3, double **host_lj4, double **host_lj2, double **host_lj3, double **host_lj4,
double **offset, double *special_lj, double **host_a12, double **offset, double *special_lj, double **host_a12,
double **host_a1, double **host_a2, double **host_d1, double **host_a1, double **host_a2, double **host_d1,
double **host_d2, double **host_sigma3, double **host_sigma6, double **host_d2, double **host_sigma3, double **host_sigma6,
int **host_form, const int nlocal, int **host_form, const int nlocal,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen); const double cell_size, int &gpu_mode, FILE *screen);
void colloid_gpu_clear(); void colloid_gpu_clear();
int ** colloid_gpu_compute_n(const int ago, const int inum, int ** colloid_gpu_compute_n(const int ago, const int inum,
const int nall, double **host_x, int *host_type, const int nall, double **host_x, int *host_type,
double *sublo, double *subhi, tagint *tag, int **nspecial, double *sublo, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, tagint **special, const bool eflag, const bool vflag,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, int **ilist, int **jnum,
const double cpu_time, bool &success); const double cpu_time, bool &success);
void colloid_gpu_compute(const int ago, const int inum, const int nall, void colloid_gpu_compute(const int ago, const int inum, 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,
@ -69,7 +69,7 @@ PairColloidGPU::PairColloidGPU(LAMMPS *lmp) : PairColloid(lmp), gpu_mode(GPU_FOR
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -87,10 +87,10 @@ void PairColloidGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -99,7 +99,7 @@ void PairColloidGPU::compute(int eflag, int vflag)
atom->x, atom->type, domain->sublo, atom->x, atom->type, domain->sublo,
domain->subhi, atom->tag, atom->nspecial, domain->subhi, atom->tag, atom->nspecial,
atom->special, eflag, vflag, eflag_atom, atom->special, eflag, vflag, eflag_atom,
vflag_atom, host_start, vflag_atom, host_start,
&ilist, &numneigh, cpu_time, success); &ilist, &numneigh, cpu_time, success);
} else { } else {
inum = list->inum; inum = list->inum;
@ -126,7 +126,7 @@ void PairColloidGPU::compute(int eflag, int vflag)
void PairColloidGPU::init_style() void PairColloidGPU::init_style()
{ {
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with colloid/gpu pair style"); error->all(FLERR,"Cannot use newton pair with colloid/gpu pair style");
// Repeat cutsq calculation because done after call to init_style // Repeat cutsq calculation because done after call to init_style
@ -160,7 +160,7 @@ void PairColloidGPU::init_style()
if (atom->molecular) if (atom->molecular)
maxspecial=atom->maxspecial; maxspecial=atom->maxspecial;
int success = colloid_gpu_init(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, int success = colloid_gpu_init(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4,
offset, force->special_lj, a12, a1, a2, offset, force->special_lj, a12, a1, a2,
d1, d2, sigma3, sigma6, _form, atom->nlocal, d1, d2, sigma3, sigma6, _form, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
cell_size, gpu_mode, screen); cell_size, gpu_mode, screen);
@ -184,8 +184,8 @@ double PairColloidGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairColloidGPU::cpu_compute(int start, int inum, int eflag, int vflag, void PairColloidGPU::cpu_compute(int start, int inum, int eflag, int vflag,
int *ilist, int *numneigh, int **firstneigh) int *ilist, int *numneigh, int **firstneigh)
{ {
int i,j,ii,jj,jnum,itype,jtype; int i,j,ii,jj,jnum,itype,jtype;
double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair; double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair;
@ -224,16 +224,16 @@ void PairColloidGPU::cpu_compute(int start, int inum, int eflag, int vflag,
if (rsq >= cutsq[itype][jtype]) continue; if (rsq >= cutsq[itype][jtype]) continue;
switch (form[itype][jtype]) { switch (form[itype][jtype]) {
case SMALL_SMALL: case SMALL_SMALL:
r2inv = 1.0/rsq; r2inv = 1.0/rsq;
r6inv = r2inv*r2inv*r2inv; r6inv = r2inv*r2inv*r2inv;
forcelj = r6inv * (lj1[itype][jtype]*r6inv - lj2[itype][jtype]); forcelj = r6inv * (lj1[itype][jtype]*r6inv - lj2[itype][jtype]);
fpair = factor_lj*forcelj*r2inv; fpair = factor_lj*forcelj*r2inv;
if (eflag) if (eflag)
evdwl = r6inv*(r6inv*lj3[itype][jtype]-lj4[itype][jtype]) - evdwl = r6inv*(r6inv*lj3[itype][jtype]-lj4[itype][jtype]) -
offset[itype][jtype]; offset[itype][jtype];
break; break;
case SMALL_LARGE: case SMALL_LARGE:
c2 = a2[itype][jtype]; c2 = a2[itype][jtype];
K[1] = c2*c2; K[1] = c2*c2;
@ -244,14 +244,14 @@ void PairColloidGPU::cpu_compute(int start, int inum, int eflag, int vflag,
K[3] *= K[3]*K[3]; K[3] *= K[3]*K[3];
K[6] = K[3]*K[3]; K[6] = K[3]*K[3];
fR = sigma3[itype][jtype]*a12[itype][jtype]*c2*K[1]/K[3]; fR = sigma3[itype][jtype]*a12[itype][jtype]*c2*K[1]/K[3];
fpair = 4.0/15.0*fR*factor_lj * fpair = 4.0/15.0*fR*factor_lj *
(2.0*(K[1]+K[2]) * (K[1]*(5.0*K[1]+22.0*K[2])+5.0*K[4]) * (2.0*(K[1]+K[2]) * (K[1]*(5.0*K[1]+22.0*K[2])+5.0*K[4]) *
sigma6[itype][jtype]/K[6]-5.0) / K[0]; sigma6[itype][jtype]/K[6]-5.0) / K[0];
if (eflag) if (eflag)
evdwl = 2.0/9.0*fR * evdwl = 2.0/9.0*fR *
(1.0-(K[1]*(K[1]*(K[1]/3.0+3.0*K[2])+4.2*K[4])+K[2]*K[4]) * (1.0-(K[1]*(K[1]*(K[1]/3.0+3.0*K[2])+4.2*K[4])+K[2]*K[4]) *
sigma6[itype][jtype]/K[6]) - offset[itype][jtype]; sigma6[itype][jtype]/K[6]) - offset[itype][jtype];
if (rsq <= K[1]) if (rsq <= K[1])
error->one(FLERR,"Overlapping small/large in pair colloid"); error->one(FLERR,"Overlapping small/large in pair colloid");
break; break;
@ -284,17 +284,17 @@ void PairColloidGPU::cpu_compute(int start, int inum, int eflag, int vflag,
fR = a12[itype][jtype]*sigma6[itype][jtype]/r/37800.0; fR = a12[itype][jtype]*sigma6[itype][jtype]/r/37800.0;
evdwl = fR * (h[0]-h[1]-h[2]+h[3]); evdwl = fR * (h[0]-h[1]-h[2]+h[3]);
dUR = evdwl/r + 5.0*fR*(g[0]+g[1]-g[2]-g[3]); dUR = evdwl/r + 5.0*fR*(g[0]+g[1]-g[2]-g[3]);
dUA = -a12[itype][jtype]/3.0*r*((2.0*K[0]*K[7]+1.0)*K[7] + dUA = -a12[itype][jtype]/3.0*r*((2.0*K[0]*K[7]+1.0)*K[7] +
(2.0*K[0]*K[8]-1.0)*K[8]); (2.0*K[0]*K[8]-1.0)*K[8]);
fpair = factor_lj * (dUR+dUA)/r; fpair = factor_lj * (dUR+dUA)/r;
if (eflag) if (eflag)
evdwl += a12[itype][jtype]/6.0 * evdwl += a12[itype][jtype]/6.0 *
(2.0*K[0]*(K[7]+K[8])-log(K[8]/K[7])) - offset[itype][jtype]; (2.0*K[0]*(K[7]+K[8])-log(K[8]/K[7])) - offset[itype][jtype];
if (r <= K[1]) if (r <= K[1])
error->one(FLERR,"Overlapping large/large in pair colloid"); error->one(FLERR,"Overlapping large/large in pair colloid");
break; break;
} }
if (eflag) evdwl *= factor_lj; if (eflag) evdwl *= factor_lj;
f[i][0] += delx*fpair; f[i][0] += delx*fpair;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen Contributing author: Trung Dac Nguyen
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_coul_cut_gpu.h" #include "pair_coul_cut_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -172,7 +172,7 @@ void PairCoulCutGPU::init_style()
void PairCoulCutGPU::reinit() void PairCoulCutGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
coul_gpu_reinit(atom->ntypes+1, scale); coul_gpu_reinit(atom->ntypes+1, scale);
} }

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ndtrung@umich.edu) Contributing author: Trung Dac Nguyen (ndtrung@umich.edu)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_coul_debye_gpu.h" #include "pair_coul_debye_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -46,15 +46,15 @@ int cdebye_gpu_init(const int ntypes, double **host_scale, double **cutsq,
const double qqrd2e, const double kappa); const double qqrd2e, const double kappa);
void cdebye_gpu_reinit(const int ntypes, double **host_scale); void cdebye_gpu_reinit(const int ntypes, double **host_scale);
void cdebye_gpu_clear(); void cdebye_gpu_clear();
int ** cdebye_gpu_compute_n(const int ago, const int inum, const int nall, int ** cdebye_gpu_compute_n(const int ago, const int inum, const int nall,
double **host_x, int *host_type, double **host_x, int *host_type,
double *sublo, double *subhi, tagint *tag, int **nspecial, double *sublo, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, tagint **special, const bool eflag, const bool vflag,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, const double cpu_time, int **ilist, int **jnum, const double cpu_time,
bool &success, double *host_q, double *boxlo, bool &success, double *host_q, double *boxlo,
double *prd); double *prd);
void cdebye_gpu_compute(const int ago, const int inum, const int nall, void cdebye_gpu_compute(const int ago, const int inum, const int nall,
double **host_x, int *host_type, double **host_x, int *host_type,
int *ilist, int *numj, int **firstneigh, int *ilist, int *numj, int **firstneigh,
const bool eflag, const bool vflag, const bool eatom, const bool eflag, const bool vflag, const bool eatom,
@ -65,12 +65,12 @@ double cdebye_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairCoulDebyeGPU::PairCoulDebyeGPU(LAMMPS *lmp) : PairCoulDebyeGPU::PairCoulDebyeGPU(LAMMPS *lmp) :
PairCoulDebye(lmp), gpu_mode(GPU_FORCE) PairCoulDebye(lmp), gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -88,12 +88,12 @@ void PairCoulDebyeGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
inum = atom->nlocal; inum = atom->nlocal;
firstneigh = cdebye_gpu_compute_n(neighbor->ago, inum, nall, atom->x, firstneigh = cdebye_gpu_compute_n(neighbor->ago, inum, nall, atom->x,
@ -101,7 +101,7 @@ void PairCoulDebyeGPU::compute(int eflag, int vflag)
atom->tag, atom->nspecial, atom->special, atom->tag, atom->nspecial, atom->special,
eflag, vflag, eflag_atom, vflag_atom, eflag, vflag, eflag_atom, vflag_atom,
host_start, &ilist, &numneigh, cpu_time, host_start, &ilist, &numneigh, cpu_time,
success, atom->q, domain->boxlo, success, atom->q, domain->boxlo,
domain->prd); domain->prd);
} else { } else {
inum = list->inum; inum = list->inum;
@ -132,7 +132,7 @@ void PairCoulDebyeGPU::init_style()
if (!atom->q_flag) if (!atom->q_flag)
error->all(FLERR,"Pair style coul/debye/gpu requires atom attribute q"); error->all(FLERR,"Pair style coul/debye/gpu requires atom attribute q");
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with coul/debye/gpu pair style"); error->all(FLERR,"Cannot use newton pair with coul/debye/gpu pair style");
// Repeat cutsq calculation because done after call to init_style // Repeat cutsq calculation because done after call to init_style
@ -174,7 +174,7 @@ void PairCoulDebyeGPU::init_style()
void PairCoulDebyeGPU::reinit() void PairCoulDebyeGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
cdebye_gpu_reinit(atom->ntypes+1, scale); cdebye_gpu_reinit(atom->ntypes+1, scale);
} }
@ -188,7 +188,7 @@ double PairCoulDebyeGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairCoulDebyeGPU::cpu_compute(int start, int inum, int eflag, void PairCoulDebyeGPU::cpu_compute(int start, int inum, int eflag,
int vflag, int *ilist, int vflag, int *ilist,
int *numneigh, int **firstneigh) int *numneigh, int **firstneigh)
{ {

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_coul_dsf_gpu.h" #include "pair_coul_dsf_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
#define MY_PIS 1.77245385090551602729 #define MY_PIS 1.77245385090551602729
@ -52,8 +52,8 @@ int cdsf_gpu_init(const int ntypes, const int nlocal, const int nall,
const int max_nbors, const int maxspecial, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen, const double cell_size, int &gpu_mode, FILE *screen,
const double host_cut_coulsq, const double host_cut_coulsq,
double *host_special_coul, const double qqrd2e, double *host_special_coul, const double qqrd2e,
const double e_shift, const double f_shift, const double e_shift, const double f_shift,
const double alpha); const double alpha);
void cdsf_gpu_clear(); void cdsf_gpu_clear();
int ** cdsf_gpu_compute_n(const int ago, const int inum, int ** cdsf_gpu_compute_n(const int ago, const int inum,
@ -75,7 +75,7 @@ double cdsf_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairCoulDSFGPU::PairCoulDSFGPU(LAMMPS *lmp) : PairCoulDSF(lmp), PairCoulDSFGPU::PairCoulDSFGPU(LAMMPS *lmp) : PairCoulDSF(lmp),
gpu_mode(GPU_FORCE) gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
@ -164,10 +164,10 @@ void PairCoulDSFGPU::init_style()
double cell_size = sqrt(maxcut) + neighbor->skin; double cell_size = sqrt(maxcut) + neighbor->skin;
cut_coulsq = cut_coul * cut_coul; cut_coulsq = cut_coul * cut_coul;
double erfcc = erfc(alpha*cut_coul); double erfcc = erfc(alpha*cut_coul);
double erfcd = exp(-alpha*alpha*cut_coul*cut_coul); double erfcd = exp(-alpha*alpha*cut_coul*cut_coul);
f_shift = -(erfcc/cut_coulsq + 2.0/MY_PIS*alpha*erfcd/cut_coul); f_shift = -(erfcc/cut_coulsq + 2.0/MY_PIS*alpha*erfcd/cut_coul);
e_shift = erfcc/cut_coul - f_shift*cut_coul; e_shift = erfcc/cut_coul - f_shift*cut_coul;
int maxspecial=0; int maxspecial=0;
if (atom->molecular) if (atom->molecular)
@ -248,10 +248,10 @@ void PairCoulDSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
erfcd = exp(-alpha*alpha*r*r); erfcd = exp(-alpha*alpha*r*r);
t = 1.0 / (1.0 + EWALD_P*alpha*r); t = 1.0 / (1.0 + EWALD_P*alpha*r);
erfcc = t * (A1+t*(A2+t*(A3+t*(A4+t*A5)))) * erfcd; erfcc = t * (A1+t*(A2+t*(A3+t*(A4+t*A5)))) * erfcd;
forcecoul = prefactor * (erfcc/r + 2.0*alpha/MY_PIS * erfcd + forcecoul = prefactor * (erfcc/r + 2.0*alpha/MY_PIS * erfcd +
r*f_shift) * r; r*f_shift) * r;
if (factor_coul < 1.0) forcecoul -= (1.0-factor_coul)*prefactor; if (factor_coul < 1.0) forcecoul -= (1.0-factor_coul)*prefactor;
fpair = forcecoul * r2inv; fpair = forcecoul * r2inv;
f[i][0] += delx*fpair; f[i][0] += delx*fpair;
f[i][1] += dely*fpair; f[i][1] += dely*fpair;

View File

@ -15,9 +15,9 @@
Contributing author: Axel Kohlmeyer (Temple) Contributing author: Axel Kohlmeyer (Temple)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_coul_long_gpu.h" #include "pair_coul_long_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "kspace.h" #include "kspace.h"
#include "gpu_extra.h" #include "gpu_extra.h"
@ -181,7 +181,7 @@ void PairCoulLongGPU::init_style()
void PairCoulLongGPU::reinit() void PairCoulLongGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
cl_gpu_reinit(atom->ntypes+1, scale); cl_gpu_reinit(atom->ntypes+1, scale);
} }

View File

@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_dpd_gpu.h" #include "pair_dpd_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -33,7 +33,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -49,10 +49,10 @@ void dpd_gpu_clear();
int ** dpd_gpu_compute_n(const int ago, const int inum_full, const int nall, int ** dpd_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,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, const double cpu_time, bool &success, int **ilist, int **jnum, const double cpu_time, bool &success,
double **host_v, const double dtinvsqrt, double **host_v, const double dtinvsqrt,
const int seed, const int timestep, const int seed, const int timestep,
double *boxlo, double *prd); double *boxlo, double *prd);
void dpd_gpu_compute(const int ago, const int inum_full, const int nall, void dpd_gpu_compute(const int ago, const int inum_full, const int nall,
@ -60,7 +60,7 @@ void dpd_gpu_compute(const int ago, const int inum_full, const int nall,
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,
const double cpu_time, bool &success, tagint *tag, const double cpu_time, bool &success, tagint *tag,
double **host_v, const double dtinvsqrt, double **host_v, const double dtinvsqrt,
const int seed, const int timestep, const int seed, const int timestep,
const int nlocal, double *boxlo, double *prd); const int nlocal, double *boxlo, double *prd);
double dpd_gpu_bytes(); double dpd_gpu_bytes();
@ -75,7 +75,7 @@ double dpd_gpu_bytes();
#define _USE_UNIFORM_SARU_LCG #define _USE_UNIFORM_SARU_LCG
#endif #endif
// References: // References:
// 1. Y. Afshar, F. Schmid, A. Pishevar, S. Worley, Comput. Phys. Comm. 184 (2013), 11191128. // 1. Y. Afshar, F. Schmid, A. Pishevar, S. Worley, Comput. Phys. Comm. 184 (2013), 11191128.
// 2. C. L. Phillips, J. A. Anderson, S. C. Glotzer, Comput. Phys. Comm. 230 (2011), 7191-7201. // 2. C. L. Phillips, J. A. Anderson, S. C. Glotzer, Comput. Phys. Comm. 230 (2011), 7191-7201.
// PRNG period = 3666320093*2^32 ~ 2^64 ~ 10^19 // PRNG period = 3666320093*2^32 ~ 2^64 ~ 10^19
@ -87,9 +87,9 @@ double dpd_gpu_bytes();
#define TWO_N32 0.232830643653869628906250e-9f /* 2^-32 */ #define TWO_N32 0.232830643653869628906250e-9f /* 2^-32 */
// specifically implemented for steps = 1; high = 1.0; low = -1.0 // specifically implemented for steps = 1; high = 1.0; low = -1.0
// returns uniformly distributed random numbers u in [-1.0;1.0] // returns uniformly distributed random numbers u in [-1.0;1.0]
// using the inherent LCG, then multiply u with sqrt(3) to "match" // using the inherent LCG, then multiply u with sqrt(3) to "match"
// with a normal random distribution. // with a normal random distribution.
// Afshar et al. mutlplies u in [-0.5;0.5] with sqrt(12) // Afshar et al. mutlplies u in [-0.5;0.5] with sqrt(12)
// Curly brackets to make variables local to the scope. // Curly brackets to make variables local to the scope.
#ifdef _USE_UNIFORM_SARU_LCG #ifdef _USE_UNIFORM_SARU_LCG
@ -119,8 +119,8 @@ double dpd_gpu_bytes();
#endif #endif
// specifically implemented for steps = 1; high = 1.0; low = -1.0 // specifically implemented for steps = 1; high = 1.0; low = -1.0
// returns uniformly distributed random numbers u in [-1.0;1.0] using TEA8 // returns uniformly distributed random numbers u in [-1.0;1.0] using TEA8
// then multiply u with sqrt(3) to "match" with a normal random distribution // then multiply u with sqrt(3) to "match" with a normal random distribution
// Afshar et al. mutlplies u in [-0.5;0.5] with sqrt(12) // Afshar et al. mutlplies u in [-0.5;0.5] with sqrt(12)
#ifdef _USE_UNIFORM_SARU_TEA8 #ifdef _USE_UNIFORM_SARU_TEA8
#define numtyp double #define numtyp double
@ -159,7 +159,7 @@ double dpd_gpu_bytes();
#endif #endif
// specifically implemented for steps = 1; high = 1.0; low = -1.0 // specifically implemented for steps = 1; high = 1.0; low = -1.0
// returns two uniformly distributed random numbers r1 and r2 in [-1.0;1.0], // returns two uniformly distributed random numbers r1 and r2 in [-1.0;1.0],
// and uses the polar method (Marsaglia's) to transform to a normal random value // and uses the polar method (Marsaglia's) to transform to a normal random value
// This is used to compared with CPU DPD using RandMars::gaussian() // This is used to compared with CPU DPD using RandMars::gaussian()
#ifdef _USE_GAUSSIAN_SARU_LCG #ifdef _USE_GAUSSIAN_SARU_LCG
@ -232,7 +232,7 @@ void PairDPDGPU::compute(int eflag, int vflag)
int inum, host_start; int inum, host_start;
double dtinvsqrt = 1.0/sqrt(update->dt); double dtinvsqrt = 1.0/sqrt(update->dt);
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -242,7 +242,7 @@ void PairDPDGPU::compute(int eflag, int vflag)
atom->tag, atom->nspecial, atom->special, atom->tag, atom->nspecial, atom->special,
eflag, vflag, eflag_atom, vflag_atom, eflag, vflag, eflag_atom, vflag_atom,
host_start, &ilist, &numneigh, cpu_time, host_start, &ilist, &numneigh, cpu_time,
success, atom->v, dtinvsqrt, seed, success, atom->v, dtinvsqrt, seed,
update->ntimestep, update->ntimestep,
domain->boxlo, domain->prd); domain->boxlo, domain->prd);
} else { } else {
@ -252,8 +252,8 @@ void PairDPDGPU::compute(int eflag, int vflag)
firstneigh = list->firstneigh; firstneigh = list->firstneigh;
dpd_gpu_compute(neighbor->ago, inum, nall, atom->x, atom->type, dpd_gpu_compute(neighbor->ago, inum, nall, atom->x, atom->type,
ilist, numneigh, firstneigh, eflag, vflag, eflag_atom, ilist, numneigh, firstneigh, eflag, vflag, eflag_atom,
vflag_atom, host_start, cpu_time, success, vflag_atom, host_start, cpu_time, success,
atom->tag, atom->v, dtinvsqrt, seed, atom->tag, atom->v, dtinvsqrt, seed,
update->ntimestep, update->ntimestep,
atom->nlocal, domain->boxlo, domain->prd); atom->nlocal, domain->boxlo, domain->prd);
} }
@ -296,7 +296,7 @@ void PairDPDGPU::init_style()
int maxspecial=0; int maxspecial=0;
if (atom->molecular) if (atom->molecular)
maxspecial=atom->maxspecial; maxspecial=atom->maxspecial;
int success = dpd_gpu_init(atom->ntypes+1, cutsq, a0, gamma, sigma, int success = dpd_gpu_init(atom->ntypes+1, cutsq, a0, gamma, sigma,
cut, force->special_lj, false, atom->nlocal, cut, force->special_lj, false, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
cell_size, gpu_mode, screen); cell_size, gpu_mode, screen);

View File

@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_dpd_tstat_gpu.h" #include "pair_dpd_tstat_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -33,7 +33,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -49,8 +49,8 @@ void dpd_gpu_clear();
int ** dpd_gpu_compute_n(const int ago, const int inum_full, const int nall, int ** dpd_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,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, const double cpu_time, bool &success, int **ilist, int **jnum, const double cpu_time, bool &success,
double **host_v, const double dtinvsqrt, double **host_v, const double dtinvsqrt,
const int seed, const int timestep, const int seed, const int timestep,
@ -60,7 +60,7 @@ void dpd_gpu_compute(const int ago, const int inum_full, const int nall,
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,
const double cpu_time, bool &success, tagint *tag, const double cpu_time, bool &success, tagint *tag,
double **host_v, const double dtinvsqrt, double **host_v, const double dtinvsqrt,
const int seed, const int timestep, const int seed, const int timestep,
const int nlocal, double *boxlo, double *prd); const int nlocal, double *boxlo, double *prd);
void dpd_gpu_update_coeff(int ntypes, double **host_a0, double **host_gamma, void dpd_gpu_update_coeff(int ntypes, double **host_a0, double **host_gamma,
@ -77,7 +77,7 @@ double dpd_gpu_bytes();
#define _USE_UNIFORM_SARU_LCG #define _USE_UNIFORM_SARU_LCG
#endif #endif
// References: // References:
// 1. Y. Afshar, F. Schmid, A. Pishevar, S. Worley, Comput. Phys. Comm. 184 (2013), 11191128. // 1. Y. Afshar, F. Schmid, A. Pishevar, S. Worley, Comput. Phys. Comm. 184 (2013), 11191128.
// 2. C. L. Phillips, J. A. Anderson, S. C. Glotzer, Comput. Phys. Comm. 230 (2011), 7191-7201. // 2. C. L. Phillips, J. A. Anderson, S. C. Glotzer, Comput. Phys. Comm. 230 (2011), 7191-7201.
// PRNG period = 3666320093*2^32 ~ 2^64 ~ 10^19 // PRNG period = 3666320093*2^32 ~ 2^64 ~ 10^19
@ -89,9 +89,9 @@ double dpd_gpu_bytes();
#define TWO_N32 0.232830643653869628906250e-9f /* 2^-32 */ #define TWO_N32 0.232830643653869628906250e-9f /* 2^-32 */
// specifically implemented for steps = 1; high = 1.0; low = -1.0 // specifically implemented for steps = 1; high = 1.0; low = -1.0
// returns uniformly distributed random numbers u in [-1.0;1.0] // returns uniformly distributed random numbers u in [-1.0;1.0]
// using the inherent LCG, then multiply u with sqrt(3) to "match" // using the inherent LCG, then multiply u with sqrt(3) to "match"
// with a normal random distribution. // with a normal random distribution.
// Afshar et al. mutlplies u in [-0.5;0.5] with sqrt(12) // Afshar et al. mutlplies u in [-0.5;0.5] with sqrt(12)
// Curly brackets to make variables local to the scope. // Curly brackets to make variables local to the scope.
#ifdef _USE_UNIFORM_SARU_LCG #ifdef _USE_UNIFORM_SARU_LCG
@ -121,8 +121,8 @@ double dpd_gpu_bytes();
#endif #endif
// specifically implemented for steps = 1; high = 1.0; low = -1.0 // specifically implemented for steps = 1; high = 1.0; low = -1.0
// returns uniformly distributed random numbers u in [-1.0;1.0] using TEA8 // returns uniformly distributed random numbers u in [-1.0;1.0] using TEA8
// then multiply u with sqrt(3) to "match" with a normal random distribution // then multiply u with sqrt(3) to "match" with a normal random distribution
// Afshar et al. mutlplies u in [-0.5;0.5] with sqrt(12) // Afshar et al. mutlplies u in [-0.5;0.5] with sqrt(12)
#ifdef _USE_UNIFORM_SARU_TEA8 #ifdef _USE_UNIFORM_SARU_TEA8
#define numtyp double #define numtyp double
@ -161,7 +161,7 @@ double dpd_gpu_bytes();
#endif #endif
// specifically implemented for steps = 1; high = 1.0; low = -1.0 // specifically implemented for steps = 1; high = 1.0; low = -1.0
// returns two uniformly distributed random numbers r1 and r2 in [-1.0;1.0], // returns two uniformly distributed random numbers r1 and r2 in [-1.0;1.0],
// and uses the polar method (Marsaglia's) to transform to a normal random value // and uses the polar method (Marsaglia's) to transform to a normal random value
// This is used to compared with CPU DPD using RandMars::gaussian() // This is used to compared with CPU DPD using RandMars::gaussian()
#ifdef _USE_GAUSSIAN_SARU_LCG #ifdef _USE_GAUSSIAN_SARU_LCG
@ -241,7 +241,7 @@ void PairDPDTstatGPU::compute(int eflag, int vflag)
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++)
sigma[i][j] = sigma[j][i] = sqrt(2.0*boltz*temperature*gamma[i][j]); sigma[i][j] = sigma[j][i] = sqrt(2.0*boltz*temperature*gamma[i][j]);
dpd_gpu_update_coeff(atom->ntypes+1, a0, gamma, sigma, cut); dpd_gpu_update_coeff(atom->ntypes+1, a0, gamma, sigma, cut);
} }
@ -249,7 +249,7 @@ void PairDPDTstatGPU::compute(int eflag, int vflag)
int inum, host_start; int inum, host_start;
double dtinvsqrt = 1.0/sqrt(update->dt); double dtinvsqrt = 1.0/sqrt(update->dt);
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -259,7 +259,7 @@ void PairDPDTstatGPU::compute(int eflag, int vflag)
atom->tag, atom->nspecial, atom->special, atom->tag, atom->nspecial, atom->special,
eflag, vflag, eflag_atom, vflag_atom, eflag, vflag, eflag_atom, vflag_atom,
host_start, &ilist, &numneigh, cpu_time, host_start, &ilist, &numneigh, cpu_time,
success, atom->v, dtinvsqrt, seed, success, atom->v, dtinvsqrt, seed,
update->ntimestep, update->ntimestep,
domain->boxlo, domain->prd); domain->boxlo, domain->prd);
} else { } else {
@ -269,8 +269,8 @@ void PairDPDTstatGPU::compute(int eflag, int vflag)
firstneigh = list->firstneigh; firstneigh = list->firstneigh;
dpd_gpu_compute(neighbor->ago, inum, nall, atom->x, atom->type, dpd_gpu_compute(neighbor->ago, inum, nall, atom->x, atom->type,
ilist, numneigh, firstneigh, eflag, vflag, eflag_atom, ilist, numneigh, firstneigh, eflag, vflag, eflag_atom,
vflag_atom, host_start, cpu_time, success, vflag_atom, host_start, cpu_time, success,
atom->tag, atom->v, dtinvsqrt, seed, atom->tag, atom->v, dtinvsqrt, seed,
update->ntimestep, update->ntimestep,
atom->nlocal, domain->boxlo, domain->prd); atom->nlocal, domain->boxlo, domain->prd);
} }
@ -313,7 +313,7 @@ void PairDPDTstatGPU::init_style()
int maxspecial=0; int maxspecial=0;
if (atom->molecular) if (atom->molecular)
maxspecial=atom->maxspecial; maxspecial=atom->maxspecial;
int success = dpd_gpu_init(atom->ntypes+1, cutsq, a0, gamma, sigma, int success = dpd_gpu_init(atom->ntypes+1, cutsq, a0, gamma, sigma,
cut, force->special_lj, true, atom->nlocal, cut, force->special_lj, true, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
cell_size, gpu_mode, screen); cell_size, gpu_mode, screen);
@ -410,7 +410,7 @@ void PairDPDTstatGPU::cpu_compute(int start, int inum, int eflag, int vflag,
f[i][0] += delx*fpair; f[i][0] += delx*fpair;
f[i][1] += dely*fpair; f[i][1] += dely*fpair;
f[i][2] += delz*fpair; f[i][2] += delz*fpair;
if (evflag) ev_tally_full(i,0.0,0.0,fpair,delx,dely,delz); if (evflag) ev_tally_full(i,0.0,0.0,fpair,delx,dely,delz);
} }
} }

View File

@ -15,9 +15,9 @@
Contributing authors: Trung Dac Nguyen (ORNL), W. Michael Brown (ORNL) Contributing authors: Trung Dac Nguyen (ORNL), W. Michael Brown (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_eam_alloy_gpu.h" #include "pair_eam_alloy_gpu.h"
#include "atom.h" #include "atom.h"
#include "force.h" #include "force.h"
@ -39,10 +39,10 @@ int eam_alloy_gpu_init(const int ntypes, double host_cutforcesq,
int **host_type2rhor, int **host_type2z2r, int **host_type2rhor, int **host_type2z2r,
int *host_type2frho, double ***host_rhor_spline, int *host_type2frho, double ***host_rhor_spline,
double ***host_z2r_spline, double ***host_frho_spline, double ***host_z2r_spline, double ***host_frho_spline,
double rdr, double rdrho, double rhomax, double rdr, double rdrho, double rhomax,
int nrhor, int nrho, int nz2r, int nfrho, int nr, int nrhor, int nrho, int nz2r, int nfrho, int nr,
const int nlocal, const int nall, const int max_nbors, const int nlocal, const int nall, const int max_nbors,
const int maxspecial, const double cell_size, int &gpu_mode, const int maxspecial, const double cell_size, int &gpu_mode,
FILE *screen, int &fp_size); FILE *screen, int &fp_size);
void eam_alloy_gpu_clear(); void eam_alloy_gpu_clear();
int** eam_alloy_gpu_compute_n(const int ago, const int inum_full, const int nall, int** eam_alloy_gpu_compute_n(const int ago, const int inum_full, const int nall,
@ -233,7 +233,7 @@ double PairEAMAlloyGPU::single(int i, int j, int itype, int jtype,
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
int PairEAMAlloyGPU::pack_forward_comm(int n, int *list, double *buf, int PairEAMAlloyGPU::pack_forward_comm(int n, int *list, double *buf,
int pbc_flag,int *pbc) int pbc_flag,int *pbc)
{ {
int i,j,m; int i,j,m;

View File

@ -15,9 +15,9 @@
Contributing authors: Trung Dac Nguyen (ORNL), W. Michael Brown (ORNL) Contributing authors: Trung Dac Nguyen (ORNL), W. Michael Brown (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_eam_fs_gpu.h" #include "pair_eam_fs_gpu.h"
#include "atom.h" #include "atom.h"
#include "force.h" #include "force.h"
@ -39,10 +39,10 @@ int eam_fs_gpu_init(const int ntypes, double host_cutforcesq,
int **host_type2rhor, int **host_type2z2r, int **host_type2rhor, int **host_type2z2r,
int *host_type2frho, double ***host_rhor_spline, int *host_type2frho, double ***host_rhor_spline,
double ***host_z2r_spline, double ***host_frho_spline, double ***host_z2r_spline, double ***host_frho_spline,
double rdr, double rdrho, double rhomax, double rdr, double rdrho, double rhomax,
int nrhor, int nrho, int nz2r, int nfrho, int nr, int nrhor, int nrho, int nz2r, int nfrho, int nr,
const int nlocal, const int nall, const int max_nbors, const int nlocal, const int nall, const int max_nbors,
const int maxspecial, const double cell_size, int &gpu_mode, const int maxspecial, const double cell_size, int &gpu_mode,
FILE *screen, int &fp_size); FILE *screen, int &fp_size);
void eam_fs_gpu_clear(); void eam_fs_gpu_clear();
int** eam_fs_gpu_compute_n(const int ago, const int inum_full, const int nall, int** eam_fs_gpu_compute_n(const int ago, const int inum_full, const int nall,
@ -233,7 +233,7 @@ double PairEAMFSGPU::single(int i, int j, int itype, int jtype,
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
int PairEAMFSGPU::pack_forward_comm(int n, int *list, double *buf, int PairEAMFSGPU::pack_forward_comm(int n, int *list, double *buf,
int pbc_flag,int *pbc) int pbc_flag,int *pbc)
{ {
int i,j,m; int i,j,m;

View File

@ -15,10 +15,10 @@
Contributing authors: Trung Dac Nguyen (ORNL), W. Michael Brown (ORNL) Contributing authors: Trung Dac Nguyen (ORNL), W. Michael Brown (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "string.h" #include <string.h>
#include "pair_eam_gpu.h" #include "pair_eam_gpu.h"
#include "atom.h" #include "atom.h"
#include "force.h" #include "force.h"
@ -41,10 +41,10 @@ int eam_gpu_init(const int ntypes, double host_cutforcesq,
int **host_type2rhor, int **host_type2z2r, int **host_type2rhor, int **host_type2z2r,
int *host_type2frho, double ***host_rhor_spline, int *host_type2frho, double ***host_rhor_spline,
double ***host_z2r_spline, double ***host_frho_spline, double ***host_z2r_spline, double ***host_frho_spline,
double rdr, double rdrho, double rhomax, double rdr, double rdrho, double rhomax,
int nrhor, int nrho, int nz2r, int nfrho, int nr, int nrhor, int nrho, int nz2r, int nfrho, int nr,
const int nlocal, const int nall, const int max_nbors, const int nlocal, const int nall, const int max_nbors,
const int maxspecial, const double cell_size, int &gpu_mode, const int maxspecial, const double cell_size, int &gpu_mode,
FILE *screen, int &fp_size); FILE *screen, int &fp_size);
void eam_gpu_clear(); void eam_gpu_clear();
int** eam_gpu_compute_n(const int ago, const int inum_full, const int nall, int** eam_gpu_compute_n(const int ago, const int inum_full, const int nall,
@ -237,7 +237,7 @@ double PairEAMGPU::single(int i, int j, int itype, int jtype,
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
int PairEAMGPU::pack_forward_comm(int n, int *list, double *buf, int PairEAMGPU::pack_forward_comm(int n, int *list, double *buf,
int pbc_flag,int *pbc) int pbc_flag,int *pbc)
{ {
int i,j,m; int i,j,m;

View File

@ -20,7 +20,7 @@ PairStyle(eam/gpu,PairEAMGPU)
#ifndef LMP_PAIR_EAM_GPU_H #ifndef LMP_PAIR_EAM_GPU_H
#define LMP_PAIR_EAM_GPU_H #define LMP_PAIR_EAM_GPU_H
#include "stdio.h" #include <stdio.h>
#include "pair_eam.h" #include "pair_eam.h"
namespace LAMMPS_NS { namespace LAMMPS_NS {

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_gauss_gpu.h" #include "pair_gauss_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -40,20 +40,20 @@ using namespace LAMMPS_NS;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int gauss_gpu_init(const int ntypes, double **cutsq, double **host_a, int gauss_gpu_init(const int ntypes, double **cutsq, double **host_a,
double **b, double **offset, double *special_lj, const int nlocal, double **b, double **offset, double *special_lj, const int nlocal,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen); const double cell_size, int &gpu_mode, FILE *screen);
int gauss_gpu_reinit(const int ntypes, double **cutsq, double **host_a, int gauss_gpu_reinit(const int ntypes, double **cutsq, double **host_a,
double **b, double **offset); double **b, double **offset);
void gauss_gpu_clear(); void gauss_gpu_clear();
int ** gauss_gpu_compute_n(const int ago, const int inum, int ** gauss_gpu_compute_n(const int ago, const int inum,
const int nall, double **host_x, int *host_type, const int nall, double **host_x, int *host_type,
double *sublo, double *subhi, tagint *tag, int **nspecial, double *sublo, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, tagint **special, const bool eflag, const bool vflag,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, int **ilist, int **jnum,
const double cpu_time, bool &success); const double cpu_time, bool &success);
void gauss_gpu_compute(const int ago, const int inum, const int nall, void gauss_gpu_compute(const int ago, const int inum, 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,
@ -66,7 +66,7 @@ PairGaussGPU::PairGaussGPU(LAMMPS *lmp) : PairGauss(lmp), gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -84,10 +84,10 @@ void PairGaussGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -96,7 +96,7 @@ void PairGaussGPU::compute(int eflag, int vflag)
atom->x, atom->type, domain->sublo, atom->x, atom->type, domain->sublo,
domain->subhi, atom->tag, atom->nspecial, domain->subhi, atom->tag, atom->nspecial,
atom->special, eflag, vflag, eflag_atom, atom->special, eflag, vflag, eflag_atom,
vflag_atom, host_start, vflag_atom, host_start,
&ilist, &numneigh, cpu_time, success); &ilist, &numneigh, cpu_time, success);
} else { } else {
inum = list->inum; inum = list->inum;
@ -123,7 +123,7 @@ void PairGaussGPU::compute(int eflag, int vflag)
void PairGaussGPU::init_style() void PairGaussGPU::init_style()
{ {
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with gauss/gpu pair style"); error->all(FLERR,"Cannot use newton pair with gauss/gpu pair style");
// Repeat cutsq calculation because done after call to init_style // Repeat cutsq calculation because done after call to init_style
@ -164,7 +164,7 @@ void PairGaussGPU::init_style()
void PairGaussGPU::reinit() void PairGaussGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
gauss_gpu_reinit(atom->ntypes+1, cutsq, a, b, offset); gauss_gpu_reinit(atom->ntypes+1, cutsq, a, b, offset);
} }
@ -178,7 +178,7 @@ double PairGaussGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairGaussGPU::cpu_compute(int start, int inum, int eflag, int vflag, void PairGaussGPU::cpu_compute(int start, int inum, int eflag, int vflag,
int *ilist, int *numneigh, int **firstneigh) { int *ilist, int *numneigh, int **firstneigh) {
int i,j,ii,jj,jnum,itype,jtype; int i,j,ii,jj,jnum,itype,jtype;
double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair; double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair;
@ -214,8 +214,8 @@ void PairGaussGPU::cpu_compute(int start, int inum, int eflag, int vflag,
if (rsq < cutsq[itype][jtype]) { if (rsq < cutsq[itype][jtype]) {
r2inv = 1.0/rsq; r2inv = 1.0/rsq;
forcelj = - 2.0*a[itype][jtype]*b[itype][jtype] * rsq * forcelj = - 2.0*a[itype][jtype]*b[itype][jtype] * rsq *
exp(-b[itype][jtype]*rsq); exp(-b[itype][jtype]*rsq);
fpair = factor_lj*forcelj*r2inv; fpair = factor_lj*forcelj*r2inv;
f[i][0] += delx*fpair; f[i][0] += delx*fpair;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_gayberne_gpu.h" #include "pair_gayberne_gpu.h"
#include "math_extra.h" #include "math_extra.h"
#include "atom.h" #include "atom.h"
@ -34,7 +34,7 @@
#include "universe.h" #include "universe.h"
#include "domain.h" #include "domain.h"
#include "update.h" #include "update.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj96_cut_gpu.h" #include "pair_lj96_cut_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_charmm_coul_long_gpu.h" #include "pair_lj_charmm_coul_long_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "kspace.h" #include "kspace.h"
#include "gpu_extra.h" #include "gpu_extra.h"

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_class2_coul_long_gpu.h" #include "pair_lj_class2_coul_long_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "kspace.h" #include "kspace.h"
#include "gpu_extra.h" #include "gpu_extra.h"

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_class2_gpu.h" #include "pair_lj_class2_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;

View File

@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ndactrung@gmail.com) Contributing author: Trung Dac Nguyen (ndactrung@gmail.com)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cubic_gpu.h" #include "pair_lj_cubic_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -40,9 +40,9 @@ using namespace PairLJCubicConstants;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int ljcb_gpu_init(const int ntypes, double **cutsq, double **cut_inner_sq, int ljcb_gpu_init(const int ntypes, double **cutsq, double **cut_inner_sq,
double **cut_inner, double **sigma, double **epsilon, double **cut_inner, double **sigma, double **epsilon,
double **host_lj1, double **host_lj2, double **host_lj3, double **host_lj1, double **host_lj2, double **host_lj3,
double **host_lj4, double *special_lj, const int nlocal, double **host_lj4, double *special_lj, const int nlocal,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen); const double cell_size, int &gpu_mode, FILE *screen);
@ -64,7 +64,7 @@ double ljcb_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairLJCubicGPU::PairLJCubicGPU(LAMMPS *lmp) : PairLJCubic(lmp), PairLJCubicGPU::PairLJCubicGPU(LAMMPS *lmp) : PairLJCubic(lmp),
gpu_mode(GPU_FORCE) gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
@ -91,7 +91,7 @@ void PairLJCubicGPU::compute(int eflag, int vflag)
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -151,7 +151,7 @@ void PairLJCubicGPU::init_style()
if (atom->molecular) if (atom->molecular)
maxspecial=atom->maxspecial; maxspecial=atom->maxspecial;
int success = ljcb_gpu_init(atom->ntypes+1, cutsq, cut_inner_sq, int success = ljcb_gpu_init(atom->ntypes+1, cutsq, cut_inner_sq,
cut_inner, sigma, epsilon, lj1, lj2, cut_inner, sigma, epsilon, lj1, lj2,
lj3, lj4, force->special_lj, atom->nlocal, lj3, lj4, force->special_lj, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
cell_size, gpu_mode, screen); cell_size, gpu_mode, screen);

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cut_coul_cut_gpu.h" #include "pair_lj_cut_coul_cut_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cut_coul_debye_gpu.h" #include "pair_lj_cut_coul_debye_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -40,23 +40,23 @@ using namespace LAMMPS_NS;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int ljcd_gpu_init(const int ntypes, double **cutsq, double **host_lj1, int ljcd_gpu_init(const int ntypes, double **cutsq, double **host_lj1,
double **host_lj2, double **host_lj3, double **host_lj4, double **host_lj2, double **host_lj3, double **host_lj4,
double **offset, double *special_lj, const int nlocal, double **offset, double *special_lj, const int nlocal,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen, const double cell_size, int &gpu_mode, FILE *screen,
double **host_cut_ljsq, double **host_cut_coulsq, double **host_cut_ljsq, double **host_cut_coulsq,
double *host_special_coul, const double qqrd2e, double *host_special_coul, const double qqrd2e,
const double kappa); const double kappa);
void ljcd_gpu_clear(); void ljcd_gpu_clear();
int ** ljcd_gpu_compute_n(const int ago, const int inum, const int nall, int ** ljcd_gpu_compute_n(const int ago, const int inum, const int nall,
double **host_x, int *host_type, double **host_x, int *host_type,
double *sublo, double *subhi, tagint *tag, int **nspecial, double *sublo, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, tagint **special, const bool eflag, const bool vflag,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, const double cpu_time, int **ilist, int **jnum, const double cpu_time,
bool &success, double *host_q, double *boxlo, bool &success, double *host_q, double *boxlo,
double *prd); double *prd);
void ljcd_gpu_compute(const int ago, const int inum, const int nall, void ljcd_gpu_compute(const int ago, const int inum, const int nall,
double **host_x, int *host_type, double **host_x, int *host_type,
int *ilist, int *numj, int **firstneigh, int *ilist, int *numj, int **firstneigh,
const bool eflag, const bool vflag, const bool eatom, const bool eflag, const bool vflag, const bool eatom,
@ -67,13 +67,13 @@ double ljcd_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairLJCutCoulDebyeGPU::PairLJCutCoulDebyeGPU(LAMMPS *lmp) : PairLJCutCoulDebyeGPU::PairLJCutCoulDebyeGPU(LAMMPS *lmp) :
PairLJCutCoulDebye(lmp), gpu_mode(GPU_FORCE) PairLJCutCoulDebye(lmp), gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -91,12 +91,12 @@ void PairLJCutCoulDebyeGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
inum = atom->nlocal; inum = atom->nlocal;
firstneigh = ljcd_gpu_compute_n(neighbor->ago, inum, nall, atom->x, firstneigh = ljcd_gpu_compute_n(neighbor->ago, inum, nall, atom->x,
@ -104,7 +104,7 @@ void PairLJCutCoulDebyeGPU::compute(int eflag, int vflag)
atom->tag, atom->nspecial, atom->special, atom->tag, atom->nspecial, atom->special,
eflag, vflag, eflag_atom, vflag_atom, eflag, vflag, eflag_atom, vflag_atom,
host_start, &ilist, &numneigh, cpu_time, host_start, &ilist, &numneigh, cpu_time,
success, atom->q, domain->boxlo, success, atom->q, domain->boxlo,
domain->prd); domain->prd);
} else { } else {
inum = list->inum; inum = list->inum;
@ -135,7 +135,7 @@ void PairLJCutCoulDebyeGPU::init_style()
if (!atom->q_flag) if (!atom->q_flag)
error->all(FLERR,"Pair style lj/cut/coul/debye/gpu requires atom attribute q"); error->all(FLERR,"Pair style lj/cut/coul/debye/gpu requires atom attribute q");
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with lj/cut/coul/debye/gpu pair style"); error->all(FLERR,"Cannot use newton pair with lj/cut/coul/debye/gpu pair style");
// Repeat cutsq calculation because done after call to init_style // Repeat cutsq calculation because done after call to init_style
@ -161,8 +161,8 @@ void PairLJCutCoulDebyeGPU::init_style()
int success = ljcd_gpu_init(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, int success = ljcd_gpu_init(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4,
offset, force->special_lj, atom->nlocal, offset, force->special_lj, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
cell_size, gpu_mode, screen, cut_ljsq, cell_size, gpu_mode, screen, cut_ljsq,
cut_coulsq, force->special_coul, cut_coulsq, force->special_coul,
force->qqrd2e, kappa); force->qqrd2e, kappa);
GPU_EXTRA::check_flag(success,error,world); GPU_EXTRA::check_flag(success,error,world);
@ -183,7 +183,7 @@ double PairLJCutCoulDebyeGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairLJCutCoulDebyeGPU::cpu_compute(int start, int inum, int eflag, void PairLJCutCoulDebyeGPU::cpu_compute(int start, int inum, int eflag,
int vflag, int *ilist, int vflag, int *ilist,
int *numneigh, int **firstneigh) int *numneigh, int **firstneigh)
{ {

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cut_coul_dsf_gpu.h" #include "pair_lj_cut_coul_dsf_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
#define MY_PIS 1.77245385090551602729 #define MY_PIS 1.77245385090551602729
@ -54,8 +54,8 @@ int ljd_gpu_init(const int ntypes, double **cutsq, double **host_lj1,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen, const double cell_size, int &gpu_mode, FILE *screen,
double **host_cut_ljsq, const double host_cut_coulsq, double **host_cut_ljsq, const double host_cut_coulsq,
double *host_special_coul, const double qqrd2e, double *host_special_coul, const double qqrd2e,
const double e_shift, const double f_shift, const double e_shift, const double f_shift,
const double alpha); const double alpha);
void ljd_gpu_clear(); void ljd_gpu_clear();
int ** ljd_gpu_compute_n(const int ago, const int inum, int ** ljd_gpu_compute_n(const int ago, const int inum,
@ -165,10 +165,10 @@ void PairLJCutCoulDSFGPU::init_style()
double cell_size = sqrt(maxcut) + neighbor->skin; double cell_size = sqrt(maxcut) + neighbor->skin;
cut_coulsq = cut_coul * cut_coul; cut_coulsq = cut_coul * cut_coul;
double erfcc = erfc(alpha*cut_coul); double erfcc = erfc(alpha*cut_coul);
double erfcd = exp(-alpha*alpha*cut_coul*cut_coul); double erfcd = exp(-alpha*alpha*cut_coul*cut_coul);
f_shift = -(erfcc/cut_coulsq + 2.0/MY_PIS*alpha*erfcd/cut_coul); f_shift = -(erfcc/cut_coulsq + 2.0/MY_PIS*alpha*erfcd/cut_coul);
e_shift = erfcc/cut_coul - f_shift*cut_coul; e_shift = erfcc/cut_coul - f_shift*cut_coul;
int maxspecial=0; int maxspecial=0;
if (atom->molecular) if (atom->molecular)
@ -262,7 +262,7 @@ void PairLJCutCoulDSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
erfcd = exp(-alpha*alpha*r*r); erfcd = exp(-alpha*alpha*r*r);
t = 1.0 / (1.0 + EWALD_P*alpha*r); t = 1.0 / (1.0 + EWALD_P*alpha*r);
erfcc = t * (A1+t*(A2+t*(A3+t*(A4+t*A5)))) * erfcd; erfcc = t * (A1+t*(A2+t*(A3+t*(A4+t*A5)))) * erfcd;
forcecoul = prefactor * (erfcc/r + 2.0*alpha/MY_PIS * erfcd + forcecoul = prefactor * (erfcc/r + 2.0*alpha/MY_PIS * erfcd +
r*f_shift) * r; r*f_shift) * r;
if (factor_coul < 1.0) forcecoul -= (1.0-factor_coul)*prefactor; if (factor_coul < 1.0) forcecoul -= (1.0-factor_coul)*prefactor;
} }
@ -278,7 +278,7 @@ void PairLJCutCoulDSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
offset[itype][jtype]; offset[itype][jtype];
evdwl *= factor_lj; evdwl *= factor_lj;
} else evdwl = 0.0; } else evdwl = 0.0;
if (rsq < cut_coulsq) { if (rsq < cut_coulsq) {
ecoul = prefactor * (erfcc - r*e_shift - rsq*f_shift); ecoul = prefactor * (erfcc - r*e_shift - rsq*f_shift);
if (factor_coul < 1.0) ecoul -= (1.0-factor_coul)*prefactor; if (factor_coul < 1.0) ecoul -= (1.0-factor_coul)*prefactor;

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cut_coul_long_gpu.h" #include "pair_lj_cut_coul_long_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "kspace.h" #include "kspace.h"
#include "gpu_extra.h" #include "gpu_extra.h"
@ -200,7 +200,7 @@ void PairLJCutCoulLongGPU::init_style()
void PairLJCutCoulLongGPU::reinit() void PairLJCutCoulLongGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
ljcl_gpu_reinit(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, offset, cut_ljsq); ljcl_gpu_reinit(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, offset, cut_ljsq);
} }

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cut_coul_msm_gpu.h" #include "pair_lj_cut_coul_msm_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -33,7 +33,7 @@
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "kspace.h" #include "kspace.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -56,7 +56,7 @@ int ** ljcm_gpu_compute_n(const int ago, const int inum,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, const double cpu_time, int **ilist, int **jnum, const double cpu_time,
bool &success, double *host_q, double *boxlo, double *prd); bool &success, double *host_q, double *boxlo, double *prd);
void ljcm_gpu_compute(const int ago, const int inum, const int nall, void ljcm_gpu_compute(const int ago, const int inum, 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,
@ -72,7 +72,7 @@ PairLJCutCoulMSMGPU::PairLJCutCoulMSMGPU(LAMMPS *lmp) :
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -90,10 +90,10 @@ void PairLJCutCoulMSMGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -132,8 +132,8 @@ void PairLJCutCoulMSMGPU::compute(int eflag, int vflag)
void PairLJCutCoulMSMGPU::init_style() void PairLJCutCoulMSMGPU::init_style()
{ {
cut_respa = NULL; cut_respa = NULL;
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with lj/cut/coul/msm/gpu pair style"); error->all(FLERR,"Cannot use newton pair with lj/cut/coul/msm/gpu pair style");
if (force->kspace->scalar_pressure_flag) if (force->kspace->scalar_pressure_flag)
@ -161,7 +161,7 @@ void PairLJCutCoulMSMGPU::init_style()
// setup force tables // setup force tables
if (ncoultablebits) init_tables(cut_coul,cut_respa); if (ncoultablebits) init_tables(cut_coul,cut_respa);
int maxspecial=0; int maxspecial=0;
if (atom->molecular) if (atom->molecular)
maxspecial=atom->maxspecial; maxspecial=atom->maxspecial;
@ -192,7 +192,7 @@ double PairLJCutCoulMSMGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairLJCutCoulMSMGPU::cpu_compute(int start, int inum, int eflag, int vflag, void PairLJCutCoulMSMGPU::cpu_compute(int start, int inum, int eflag, int vflag,
int *ilist, int *numneigh, int **firstneigh) { int *ilist, int *numneigh, int **firstneigh) {
int i,j,ii,jj,jnum,itype,jtype,itable; int i,j,ii,jj,jnum,itype,jtype,itable;
double qtmp,xtmp,ytmp,ztmp,delx,dely,delz,evdwl,ecoul,fpair; double qtmp,xtmp,ytmp,ztmp,delx,dely,delz,evdwl,ecoul,fpair;
@ -209,7 +209,7 @@ void PairLJCutCoulMSMGPU::cpu_compute(int start, int inum, int eflag, int vflag,
double *special_coul = force->special_coul; double *special_coul = force->special_coul;
double *special_lj = force->special_lj; double *special_lj = force->special_lj;
double qqrd2e = force->qqrd2e; double qqrd2e = force->qqrd2e;
// loop over neighbors of my atoms // loop over neighbors of my atoms
for (ii = start; ii < inum; ii++) { for (ii = start; ii < inum; ii++) {
@ -289,7 +289,7 @@ void PairLJCutCoulMSMGPU::cpu_compute(int start, int inum, int eflag, int vflag,
evdwl *= factor_lj; evdwl *= factor_lj;
} else evdwl = 0.0; } else evdwl = 0.0;
} }
if (evflag) ev_tally_full(i,evdwl,ecoul,fpair,delx,dely,delz); if (evflag) ev_tally_full(i,evdwl,ecoul,fpair,delx,dely,delz);
} }
} }

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cut_dipole_cut_gpu.h" #include "pair_lj_cut_dipole_cut_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -40,40 +40,40 @@ using namespace LAMMPS_NS;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int dpl_gpu_init(const int ntypes, double **cutsq, double **host_lj1, int dpl_gpu_init(const int ntypes, double **cutsq, double **host_lj1,
double **host_lj2, double **host_lj3, double **host_lj4, double **host_lj2, double **host_lj3, double **host_lj4,
double **offset, double *special_lj, const int nlocal, double **offset, double *special_lj, const int nlocal,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen, const double cell_size, int &gpu_mode, FILE *screen,
double **host_cut_ljsq, double **host_cut_coulsq, double **host_cut_ljsq, double **host_cut_coulsq,
double *host_special_coul, const double qqrd2e); double *host_special_coul, const double qqrd2e);
void dpl_gpu_clear(); void dpl_gpu_clear();
int ** dpl_gpu_compute_n(const int ago, const int inum, int ** dpl_gpu_compute_n(const int ago, const int inum,
const int nall, double **host_x, int *host_type, const int nall, double **host_x, int *host_type,
double *sublo, double *subhi, tagint *tag, double *sublo, double *subhi, tagint *tag,
int **nspecial, tagint **special, const bool eflag, int **nspecial, tagint **special, const bool eflag,
const bool vflag, const bool eatom, const bool vatom, const bool vflag, const bool eatom, const bool vatom,
int &host_start, int **ilist, int **jnum, int &host_start, int **ilist, int **jnum,
const double cpu_time, bool &success, const double cpu_time, bool &success,
double *host_q, double **host_mu, double *host_q, double **host_mu,
double *boxlo, double *prd); double *boxlo, double *prd);
void dpl_gpu_compute(const int ago, const int inum, void dpl_gpu_compute(const int ago, const int inum,
const int nall, double **host_x, int *host_type, const int nall, double **host_x, int *host_type,
int *ilist, int *numj, int **firstneigh, int *ilist, int *numj, int **firstneigh,
const bool eflag, const bool vflag, const bool eatom, const bool eflag, const bool vflag, const bool eatom,
const bool vatom, int &host_start, const double cpu_time, const bool vatom, int &host_start, const double cpu_time,
bool &success, double *host_q, double **host_mu, bool &success, double *host_q, double **host_mu,
const int nlocal, double *boxlo, double *prd); const int nlocal, double *boxlo, double *prd);
double dpl_gpu_bytes(); double dpl_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairLJCutDipoleCutGPU::PairLJCutDipoleCutGPU(LAMMPS *lmp) : PairLJCutDipoleCut(lmp), PairLJCutDipoleCutGPU::PairLJCutDipoleCutGPU(LAMMPS *lmp) : PairLJCutDipoleCut(lmp),
gpu_mode(GPU_FORCE) gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -91,12 +91,12 @@ void PairLJCutDipoleCutGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
inum = atom->nlocal; inum = atom->nlocal;
firstneigh = dpl_gpu_compute_n(neighbor->ago, inum, nall, atom->x, firstneigh = dpl_gpu_compute_n(neighbor->ago, inum, nall, atom->x,
@ -104,7 +104,7 @@ void PairLJCutDipoleCutGPU::compute(int eflag, int vflag)
atom->tag, atom->nspecial, atom->special, atom->tag, atom->nspecial, atom->special,
eflag, vflag, eflag_atom, vflag_atom, eflag, vflag, eflag_atom, vflag_atom,
host_start, &ilist, &numneigh, cpu_time, host_start, &ilist, &numneigh, cpu_time,
success, atom->q, atom->mu, domain->boxlo, success, atom->q, atom->mu, domain->boxlo,
domain->prd); domain->prd);
} else { } else {
inum = list->inum; inum = list->inum;
@ -134,8 +134,8 @@ void PairLJCutDipoleCutGPU::init_style()
{ {
if (!atom->q_flag || !atom->mu_flag || !atom->torque_flag) if (!atom->q_flag || !atom->mu_flag || !atom->torque_flag)
error->all(FLERR,"Pair dipole/cut/gpu requires atom attributes q, mu, torque"); error->all(FLERR,"Pair dipole/cut/gpu requires atom attributes q, mu, torque");
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with dipole/cut/gpu pair style"); error->all(FLERR,"Cannot use newton pair with dipole/cut/gpu pair style");
if (strcmp(update->unit_style,"electron") == 0) if (strcmp(update->unit_style,"electron") == 0)
@ -186,7 +186,7 @@ double PairLJCutDipoleCutGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vflag, void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vflag,
int *ilist, int *numneigh, int *ilist, int *numneigh,
int **firstneigh) int **firstneigh)
{ {
int i,j,ii,jj,jnum,itype,jtype; int i,j,ii,jj,jnum,itype,jtype;
@ -247,7 +247,7 @@ void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vfla
forcecoulx = forcecouly = forcecoulz = 0.0; forcecoulx = forcecouly = forcecoulz = 0.0;
tixcoul = tiycoul = tizcoul = 0.0; tixcoul = tiycoul = tizcoul = 0.0;
tjxcoul = tjycoul = tjzcoul = 0.0; tjxcoul = tjycoul = tjzcoul = 0.0;
if (rsq < cut_coulsq[itype][jtype]) { if (rsq < cut_coulsq[itype][jtype]) {
if (qtmp != 0.0 && q[j] != 0.0) { if (qtmp != 0.0 && q[j] != 0.0) {
@ -259,7 +259,7 @@ void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vfla
forcecoulz += pre1*delz; forcecoulz += pre1*delz;
} }
if (mu[i][3] > 0.0 && mu[j][3] > 0.0) { if (mu[i][3] > 0.0 && mu[j][3] > 0.0) {
r3inv = r2inv*rinv; r3inv = r2inv*rinv;
r5inv = r3inv*r2inv; r5inv = r3inv*r2inv;
r7inv = r5inv*r2inv; r7inv = r5inv*r2inv;
@ -276,7 +276,7 @@ void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vfla
forcecoulx += pre1*delx + pre2*mu[i][0] + pre3*mu[j][0]; forcecoulx += pre1*delx + pre2*mu[i][0] + pre3*mu[j][0];
forcecouly += pre1*dely + pre2*mu[i][1] + pre3*mu[j][1]; forcecouly += pre1*dely + pre2*mu[i][1] + pre3*mu[j][1];
forcecoulz += pre1*delz + pre2*mu[i][2] + pre3*mu[j][2]; forcecoulz += pre1*delz + pre2*mu[i][2] + pre3*mu[j][2];
crossx = pre4 * (mu[i][1]*mu[j][2] - mu[i][2]*mu[j][1]); crossx = pre4 * (mu[i][1]*mu[j][2] - mu[i][2]*mu[j][1]);
crossy = pre4 * (mu[i][2]*mu[j][0] - mu[i][0]*mu[j][2]); crossy = pre4 * (mu[i][2]*mu[j][0] - mu[i][0]*mu[j][2]);
crossz = pre4 * (mu[i][0]*mu[j][1] - mu[i][1]*mu[j][0]); crossz = pre4 * (mu[i][0]*mu[j][1] - mu[i][1]*mu[j][0]);
@ -289,7 +289,7 @@ void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vfla
tjzcoul += -crossz + pre3 * (mu[j][0]*dely - mu[j][1]*delx); tjzcoul += -crossz + pre3 * (mu[j][0]*dely - mu[j][1]*delx);
} }
if (mu[i][3] > 0.0 && q[j] != 0.0) { if (mu[i][3] > 0.0 && q[j] != 0.0) {
r3inv = r2inv*rinv; r3inv = r2inv*rinv;
r5inv = r3inv*r2inv; r5inv = r3inv*r2inv;
pidotr = mu[i][0]*delx + mu[i][1]*dely + mu[i][2]*delz; pidotr = mu[i][0]*delx + mu[i][1]*dely + mu[i][2]*delz;
@ -304,7 +304,7 @@ void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vfla
tizcoul += pre2 * (mu[i][0]*dely - mu[i][1]*delx); tizcoul += pre2 * (mu[i][0]*dely - mu[i][1]*delx);
} }
if (mu[j][3] > 0.0 && qtmp != 0.0) { if (mu[j][3] > 0.0 && qtmp != 0.0) {
r3inv = r2inv*rinv; r3inv = r2inv*rinv;
r5inv = r3inv*r2inv; r5inv = r3inv*r2inv;
pjdotr = mu[j][0]*delx + mu[j][1]*dely + mu[j][2]*delz; pjdotr = mu[j][0]*delx + mu[j][1]*dely + mu[j][2]*delz;
@ -327,7 +327,7 @@ void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vfla
forcelj = r6inv * (lj1[itype][jtype]*r6inv - lj2[itype][jtype]); forcelj = r6inv * (lj1[itype][jtype]*r6inv - lj2[itype][jtype]);
forcelj *= factor_lj * r2inv; forcelj *= factor_lj * r2inv;
} else forcelj = 0.0; } else forcelj = 0.0;
// total force // total force
fq = factor_coul*qqrd2e; fq = factor_coul*qqrd2e;
@ -349,7 +349,7 @@ void PairLJCutDipoleCutGPU::cpu_compute(int start, int inum, int eflag, int vfla
ecoul = qtmp*q[j]*rinv; ecoul = qtmp*q[j]*rinv;
if (mu[i][3] > 0.0 && mu[j][3] > 0.0) if (mu[i][3] > 0.0 && mu[j][3] > 0.0)
ecoul += r3inv*pdotp - 3.0*r5inv*pidotr*pjdotr; ecoul += r3inv*pdotp - 3.0*r5inv*pidotr*pjdotr;
if (mu[i][3] > 0.0 && q[j] != 0.0) if (mu[i][3] > 0.0 && q[j] != 0.0)
ecoul += -q[j]*r3inv*pidotr; ecoul += -q[j]*r3inv*pidotr;
if (mu[j][3] > 0.0 && qtmp != 0.0) if (mu[j][3] > 0.0 && qtmp != 0.0)
ecoul += qtmp*r3inv*pjdotr; ecoul += qtmp*r3inv*pjdotr;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_cut_gpu.h" #include "pair_lj_cut_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -91,7 +91,7 @@ void PairLJCutGPU::compute(int eflag, int vflag)
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -170,7 +170,7 @@ void PairLJCutGPU::init_style()
void PairLJCutGPU::reinit() void PairLJCutGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
ljl_gpu_reinit(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, offset); ljl_gpu_reinit(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, offset);
} }

View File

@ -15,9 +15,9 @@
Contributing author: Inderaj Bains (NVIDIA), ibains@nvidia.com Contributing author: Inderaj Bains (NVIDIA), ibains@nvidia.com
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_expand_gpu.h" #include "pair_lj_expand_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -167,7 +167,7 @@ void PairLJExpandGPU::init_style()
void PairLJExpandGPU::reinit() void PairLJExpandGPU::reinit()
{ {
Pair::reinit(); Pair::reinit();
lje_gpu_reinit(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, offset, shift); lje_gpu_reinit(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, offset, shift);
} }

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_gromacs_gpu.h" #include "pair_lj_gromacs_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "kspace.h" #include "kspace.h"
#include "gpu_extra.h" #include "gpu_extra.h"
@ -46,7 +46,7 @@ int ljgrm_gpu_init(const int ntypes, double **cutsq, double **host_lj1,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen, const double cell_size, int &gpu_mode, FILE *screen,
double **host_ljsw1, double **host_ljsw2, double **host_ljsw3, double **host_ljsw1, double **host_ljsw2, double **host_ljsw3,
double **host_ljsw4, double **host_ljsw5, double **host_ljsw4, double **host_ljsw5,
double **cut_inner, double **cut_innersq); double **cut_inner, double **cut_innersq);
void ljgrm_gpu_clear(); void ljgrm_gpu_clear();
int ** ljgrm_gpu_compute_n(const int ago, const int inum_full, int ** ljgrm_gpu_compute_n(const int ago, const int inum_full,
@ -65,13 +65,13 @@ double ljgrm_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairLJGromacsGPU::PairLJGromacsGPU(LAMMPS *lmp) : PairLJGromacsGPU::PairLJGromacsGPU(LAMMPS *lmp) :
PairLJGromacs(lmp), gpu_mode(GPU_FORCE) PairLJGromacs(lmp), gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -89,12 +89,12 @@ void PairLJGromacsGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
inum = atom->nlocal; inum = atom->nlocal;
firstneigh = ljgrm_gpu_compute_n(neighbor->ago, inum, nall, firstneigh = ljgrm_gpu_compute_n(neighbor->ago, inum, nall,
@ -128,7 +128,7 @@ void PairLJGromacsGPU::compute(int eflag, int vflag)
void PairLJGromacsGPU::init_style() void PairLJGromacsGPU::init_style()
{ {
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with lj/gromacs/gpu pair style"); error->all(FLERR,"Cannot use newton pair with lj/gromacs/gpu pair style");
// Repeat cutsq calculation because done after call to init_style // Repeat cutsq calculation because done after call to init_style
@ -155,7 +155,7 @@ void PairLJGromacsGPU::init_style()
int success = ljgrm_gpu_init(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4, int success = ljgrm_gpu_init(atom->ntypes+1, cutsq, lj1, lj2, lj3, lj4,
force->special_lj, atom->nlocal, force->special_lj, atom->nlocal,
atom->nlocal+atom->nghost, 300, maxspecial, atom->nlocal+atom->nghost, 300, maxspecial,
cell_size, gpu_mode, screen, ljsw1, ljsw2, cell_size, gpu_mode, screen, ljsw1, ljsw2,
ljsw3, ljsw4, ljsw5, cut_inner, cut_inner_sq); ljsw3, ljsw4, ljsw5, cut_inner, cut_inner_sq);
GPU_EXTRA::check_flag(success,error,world); GPU_EXTRA::check_flag(success,error,world);
@ -218,7 +218,7 @@ void PairLJGromacsGPU::cpu_compute(int start, int inum, int eflag,
r6inv = r2inv*r2inv*r2inv; r6inv = r2inv*r2inv*r2inv;
forcelj = r6inv * (lj1[itype][jtype]*r6inv - lj2[itype][jtype]); forcelj = r6inv * (lj1[itype][jtype]*r6inv - lj2[itype][jtype]);
if (rsq > cut_inner_sq[itype][jtype]) { if (rsq > cut_inner_sq[itype][jtype]) {
r = sqrt(rsq); r = sqrt(rsq);
t = r - cut_inner[itype][jtype]; t = r - cut_inner[itype][jtype];
fswitch = r*t*t*(ljsw1[itype][jtype] + ljsw2[itype][jtype]*t); fswitch = r*t*t*(ljsw1[itype][jtype] + ljsw2[itype][jtype]*t);
forcelj += fswitch; forcelj += fswitch;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_sdk_coul_long_gpu.h" #include "pair_lj_sdk_coul_long_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "kspace.h" #include "kspace.h"
#include "gpu_extra.h" #include "gpu_extra.h"

View File

@ -15,9 +15,9 @@
Contributing author: Mike Brown (SNL) Contributing author: Mike Brown (SNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_sdk_gpu.h" #include "pair_lj_sdk_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_lj_sf_dipole_sf_gpu.h" #include "pair_lj_sf_dipole_sf_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -40,20 +40,20 @@ using namespace LAMMPS_NS;
// External functions from cuda library for atom decomposition // External functions from cuda library for atom decomposition
int dplsf_gpu_init(const int ntypes, double **cutsq, double **host_lj1, int dplsf_gpu_init(const int ntypes, double **cutsq, double **host_lj1,
double **host_lj2, double **host_lj3, double **host_lj4, double **host_lj2, double **host_lj3, double **host_lj4,
double *special_lj, const int nlocal, double *special_lj, const int nlocal,
const int nall, const int max_nbors, const int maxspecial, const int nall, const int max_nbors, const int maxspecial,
const double cell_size, int &gpu_mode, FILE *screen, const double cell_size, int &gpu_mode, FILE *screen,
double **host_cut_ljsq, double **host_cut_coulsq, double **host_cut_ljsq, double **host_cut_coulsq,
double *host_special_coul, const double qqrd2e); double *host_special_coul, const double qqrd2e);
void dplsf_gpu_clear(); void dplsf_gpu_clear();
int ** dplsf_gpu_compute_n(const int ago, const int inum, int ** dplsf_gpu_compute_n(const int ago, const int inum,
const int nall, double **host_x, int *host_type, const int nall, double **host_x, int *host_type,
double *sublo, double *subhi, tagint *tag, int **nspecial, double *sublo, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, tagint **special, const bool eflag, const bool vflag,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, const double cpu_time, int **ilist, int **jnum, const double cpu_time,
bool &success, double *host_q, double **host_mu, bool &success, double *host_q, double **host_mu,
double *boxlo, double *prd); double *boxlo, double *prd);
void dplsf_gpu_compute(const int ago, const int inum, void dplsf_gpu_compute(const int ago, const int inum,
const int nall, double **host_x, int *host_type, const int nall, double **host_x, int *host_type,
@ -66,13 +66,13 @@ double dplsf_gpu_bytes();
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
PairLJSFDipoleSFGPU::PairLJSFDipoleSFGPU(LAMMPS *lmp) : PairLJSFDipoleSF(lmp), PairLJSFDipoleSFGPU::PairLJSFDipoleSFGPU(LAMMPS *lmp) : PairLJSFDipoleSF(lmp),
gpu_mode(GPU_FORCE) gpu_mode(GPU_FORCE)
{ {
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -90,12 +90,12 @@ void PairLJSFDipoleSFGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
inum = atom->nlocal; inum = atom->nlocal;
firstneigh = dplsf_gpu_compute_n(neighbor->ago, inum, nall, atom->x, firstneigh = dplsf_gpu_compute_n(neighbor->ago, inum, nall, atom->x,
@ -103,7 +103,7 @@ void PairLJSFDipoleSFGPU::compute(int eflag, int vflag)
atom->tag, atom->nspecial, atom->special, atom->tag, atom->nspecial, atom->special,
eflag, vflag, eflag_atom, vflag_atom, eflag, vflag, eflag_atom, vflag_atom,
host_start, &ilist, &numneigh, cpu_time, host_start, &ilist, &numneigh, cpu_time,
success, atom->q, atom->mu, domain->boxlo, success, atom->q, atom->mu, domain->boxlo,
domain->prd); domain->prd);
} else { } else {
inum = list->inum; inum = list->inum;
@ -133,8 +133,8 @@ void PairLJSFDipoleSFGPU::init_style()
{ {
if (!atom->q_flag || !atom->mu_flag || !atom->torque_flag) if (!atom->q_flag || !atom->mu_flag || !atom->torque_flag)
error->all(FLERR,"Pair dipole/sf/gpu requires atom attributes q, mu, torque"); error->all(FLERR,"Pair dipole/sf/gpu requires atom attributes q, mu, torque");
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with dipole/sf/gpu pair style"); error->all(FLERR,"Cannot use newton pair with dipole/sf/gpu pair style");
if (strcmp(update->unit_style,"electron") == 0) if (strcmp(update->unit_style,"electron") == 0)
@ -261,7 +261,7 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
forcecoulz += pre1*delz; forcecoulz += pre1*delz;
} }
if (mu[i][3] > 0.0 && mu[j][3] > 0.0) { if (mu[i][3] > 0.0 && mu[j][3] > 0.0) {
r3inv = r2inv*rinv; r3inv = r2inv*rinv;
r5inv = r3inv*r2inv; r5inv = r3inv*r2inv;
rcutcoul2inv=1.0/cut_coulsq[itype][jtype]; rcutcoul2inv=1.0/cut_coulsq[itype][jtype];
@ -269,7 +269,7 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
pdotp = mu[i][0]*mu[j][0] + mu[i][1]*mu[j][1] + mu[i][2]*mu[j][2]; pdotp = mu[i][0]*mu[j][0] + mu[i][1]*mu[j][1] + mu[i][2]*mu[j][2];
pidotr = mu[i][0]*delx + mu[i][1]*dely + mu[i][2]*delz; pidotr = mu[i][0]*delx + mu[i][1]*dely + mu[i][2]*delz;
pjdotr = mu[j][0]*delx + mu[j][1]*dely + mu[j][2]*delz; pjdotr = mu[j][0]*delx + mu[j][1]*dely + mu[j][2]*delz;
afac = 1.0 - rsq*rsq * rcutcoul2inv*rcutcoul2inv; afac = 1.0 - rsq*rsq * rcutcoul2inv*rcutcoul2inv;
pre1 = afac * ( pdotp - 3.0 * r2inv * pidotr * pjdotr ); pre1 = afac * ( pdotp - 3.0 * r2inv * pidotr * pjdotr );
aforcecoulx = pre1*delx; aforcecoulx = pre1*delx;
@ -282,15 +282,15 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
bforcecoulx = bfac * (pjdotr*mu[i][0]+pidotr*mu[j][0]-presf*delx); bforcecoulx = bfac * (pjdotr*mu[i][0]+pidotr*mu[j][0]-presf*delx);
bforcecouly = bfac * (pjdotr*mu[i][1]+pidotr*mu[j][1]-presf*dely); bforcecouly = bfac * (pjdotr*mu[i][1]+pidotr*mu[j][1]-presf*dely);
bforcecoulz = bfac * (pjdotr*mu[i][2]+pidotr*mu[j][2]-presf*delz); bforcecoulz = bfac * (pjdotr*mu[i][2]+pidotr*mu[j][2]-presf*delz);
forcecoulx += 3.0 * r5inv * ( aforcecoulx + bforcecoulx ); forcecoulx += 3.0 * r5inv * ( aforcecoulx + bforcecoulx );
forcecouly += 3.0 * r5inv * ( aforcecouly + bforcecouly ); forcecouly += 3.0 * r5inv * ( aforcecouly + bforcecouly );
forcecoulz += 3.0 * r5inv * ( aforcecoulz + bforcecoulz ); forcecoulz += 3.0 * r5inv * ( aforcecoulz + bforcecoulz );
pre2 = 3.0 * bfac * r5inv * pjdotr; pre2 = 3.0 * bfac * r5inv * pjdotr;
pre3 = 3.0 * bfac * r5inv * pidotr; pre3 = 3.0 * bfac * r5inv * pidotr;
pre4 = -bfac * r3inv; pre4 = -bfac * r3inv;
crossx = pre4 * (mu[i][1]*mu[j][2] - mu[i][2]*mu[j][1]); crossx = pre4 * (mu[i][1]*mu[j][2] - mu[i][2]*mu[j][1]);
crossy = pre4 * (mu[i][2]*mu[j][0] - mu[i][0]*mu[j][2]); crossy = pre4 * (mu[i][2]*mu[j][0] - mu[i][0]*mu[j][2]);
crossz = pre4 * (mu[i][0]*mu[j][1] - mu[i][1]*mu[j][0]); crossz = pre4 * (mu[i][0]*mu[j][1] - mu[i][1]*mu[j][0]);
@ -303,13 +303,13 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
tjzcoul += -crossz + pre3 * (mu[j][0]*dely - mu[j][1]*delx); tjzcoul += -crossz + pre3 * (mu[j][0]*dely - mu[j][1]*delx);
} }
if (mu[i][3] > 0.0 && q[j] != 0.0) { if (mu[i][3] > 0.0 && q[j] != 0.0) {
r3inv = r2inv*rinv; r3inv = r2inv*rinv;
r5inv = r3inv*r2inv; r5inv = r3inv*r2inv;
pidotr = mu[i][0]*delx + mu[i][1]*dely + mu[i][2]*delz; pidotr = mu[i][0]*delx + mu[i][1]*dely + mu[i][2]*delz;
rcutcoul2inv=1.0/cut_coulsq[itype][jtype]; rcutcoul2inv=1.0/cut_coulsq[itype][jtype];
pre1 = 3.0 * q[j] * r5inv * pidotr * (1-rsq*rcutcoul2inv); pre1 = 3.0 * q[j] * r5inv * pidotr * (1-rsq*rcutcoul2inv);
pqfac = 1.0 - 3.0*rsq*rcutcoul2inv + pqfac = 1.0 - 3.0*rsq*rcutcoul2inv +
2.0*rsq*sqrt(rsq)*rcutcoul2inv*sqrt(rcutcoul2inv); 2.0*rsq*sqrt(rsq)*rcutcoul2inv*sqrt(rcutcoul2inv);
pre2 = q[j] * r3inv * pqfac; pre2 = q[j] * r3inv * pqfac;
@ -321,7 +321,7 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
tizcoul += pre2 * (mu[i][0]*dely - mu[i][1]*delx); tizcoul += pre2 * (mu[i][0]*dely - mu[i][1]*delx);
} }
if (mu[j][3] > 0.0 && qtmp != 0.0) { if (mu[j][3] > 0.0 && qtmp != 0.0) {
r3inv = r2inv*rinv; r3inv = r2inv*rinv;
r5inv = r3inv*r2inv; r5inv = r3inv*r2inv;
pjdotr = mu[j][0]*delx + mu[j][1]*dely + mu[j][2]*delz; pjdotr = mu[j][0]*delx + mu[j][1]*dely + mu[j][2]*delz;
@ -337,7 +337,7 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
tjxcoul += -pre2 * (mu[j][1]*delz - mu[j][2]*dely); tjxcoul += -pre2 * (mu[j][1]*delz - mu[j][2]*dely);
tjycoul += -pre2 * (mu[j][2]*delx - mu[j][0]*delz); tjycoul += -pre2 * (mu[j][2]*delx - mu[j][0]*delz);
tjzcoul += -pre2 * (mu[j][0]*dely - mu[j][1]*delx); tjzcoul += -pre2 * (mu[j][0]*dely - mu[j][1]*delx);
} }
} }
// LJ interaction // LJ interaction
@ -345,22 +345,22 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
if (rsq < cut_ljsq[itype][jtype]) { if (rsq < cut_ljsq[itype][jtype]) {
r6inv = r2inv*r2inv*r2inv; r6inv = r2inv*r2inv*r2inv;
forceljcut = r6inv*(lj1[itype][jtype]*r6inv-lj2[itype][jtype])*r2inv; forceljcut = r6inv*(lj1[itype][jtype]*r6inv-lj2[itype][jtype])*r2inv;
rcutlj2inv = 1.0 / cut_ljsq[itype][jtype]; rcutlj2inv = 1.0 / cut_ljsq[itype][jtype];
rcutlj6inv = rcutlj2inv * rcutlj2inv * rcutlj2inv; rcutlj6inv = rcutlj2inv * rcutlj2inv * rcutlj2inv;
forceljsf = (lj1[itype][jtype]*rcutlj6inv - lj2[itype][jtype]) * forceljsf = (lj1[itype][jtype]*rcutlj6inv - lj2[itype][jtype]) *
rcutlj6inv * rcutlj2inv; rcutlj6inv * rcutlj2inv;
forcelj = factor_lj * (forceljcut - forceljsf); forcelj = factor_lj * (forceljcut - forceljsf);
} else forcelj = 0.0; } else forcelj = 0.0;
// total force // total force
fq = factor_coul*qqrd2e; fq = factor_coul*qqrd2e;
fx = fq*forcecoulx + delx*forcelj; fx = fq*forcecoulx + delx*forcelj;
fy = fq*forcecouly + dely*forcelj; fy = fq*forcecouly + dely*forcelj;
fz = fq*forcecoulz + delz*forcelj; fz = fq*forcecoulz + delz*forcelj;
// force & torque accumulation // force & torque accumulation
f[i][0] += fx; f[i][0] += fx;
@ -376,7 +376,7 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
pow((1.0-sqrt(rsq)/sqrt(cut_coulsq[itype][jtype])),2); pow((1.0-sqrt(rsq)/sqrt(cut_coulsq[itype][jtype])),2);
if (mu[i][3] > 0.0 && mu[j][3] > 0.0) if (mu[i][3] > 0.0 && mu[j][3] > 0.0)
ecoul += bfac * (r3inv*pdotp - 3.0*r5inv*pidotr*pjdotr); ecoul += bfac * (r3inv*pdotp - 3.0*r5inv*pidotr*pjdotr);
if (mu[i][3] > 0.0 && q[j] != 0.0) if (mu[i][3] > 0.0 && q[j] != 0.0)
ecoul += -q[j]*r3inv * pqfac * pidotr; ecoul += -q[j]*r3inv * pqfac * pidotr;
if (mu[j][3] > 0.0 && qtmp != 0.0) if (mu[j][3] > 0.0 && qtmp != 0.0)
ecoul += qtmp*r3inv * qpfac * pjdotr; ecoul += qtmp*r3inv * qpfac * pjdotr;
@ -389,9 +389,9 @@ void PairLJSFDipoleSFGPU::cpu_compute(int start, int inum, int eflag, int vflag,
rsq*rcutlj2inv + rsq*rcutlj2inv +
rcutlj6inv*(-7*lj3[itype][jtype]*rcutlj6inv+4*lj4[itype][jtype]); rcutlj6inv*(-7*lj3[itype][jtype]*rcutlj6inv+4*lj4[itype][jtype]);
evdwl *= factor_lj; evdwl *= factor_lj;
} else evdwl = 0.0; } else evdwl = 0.0;
} }
if (evflag) ev_tally_xyz_full(i,evdwl,ecoul, if (evflag) ev_tally_xyz_full(i,evdwl,ecoul,
fx,fy,fz,delx,dely,delz); fx,fy,fz,delx,dely,delz);
} }

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

View File

@ -2,12 +2,12 @@
LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator LAMMPS - Large-scale Atomic/Molecular Massively Parallel Simulator
http://lammps.sandia.gov, Sandia National Laboratories http://lammps.sandia.gov, Sandia National Laboratories
Steve Plimpton, sjplimp@sandia.gov Steve Plimpton, sjplimp@sandia.gov
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
@ -15,9 +15,9 @@
Contributing author: Trung Dac Nguyen (ORNL) Contributing author: Trung Dac Nguyen (ORNL)
------------------------------------------------------------------------- */ ------------------------------------------------------------------------- */
#include "math.h" #include <math.h>
#include "stdio.h" #include <stdio.h>
#include "stdlib.h" #include <stdlib.h>
#include "pair_mie_cut_gpu.h" #include "pair_mie_cut_gpu.h"
#include "atom.h" #include "atom.h"
#include "atom_vec.h" #include "atom_vec.h"
@ -32,7 +32,7 @@
#include "universe.h" #include "universe.h"
#include "update.h" #include "update.h"
#include "domain.h" #include "domain.h"
#include "string.h" #include <string.h>
#include "gpu_extra.h" #include "gpu_extra.h"
using namespace LAMMPS_NS; using namespace LAMMPS_NS;
@ -47,13 +47,13 @@ int mie_gpu_init(const int ntypes, double **cutsq, double **host_mie1,
const double cell_size, int &gpu_mode, FILE *screen); const double cell_size, int &gpu_mode, FILE *screen);
void mie_gpu_clear(); void mie_gpu_clear();
int ** mie_gpu_compute_n(const int ago, const int inum, int ** mie_gpu_compute_n(const int ago, const int inum,
const int nall, double **host_x, int *host_type, const int nall, double **host_x, int *host_type,
double *sublo, double *subhi, tagint *tag, int **nspecial, double *sublo, double *subhi, tagint *tag, int **nspecial,
tagint **special, const bool eflag, const bool vflag, tagint **special, const bool eflag, const bool vflag,
const bool eatom, const bool vatom, int &host_start, const bool eatom, const bool vatom, int &host_start,
int **ilist, int **jnum, int **ilist, int **jnum,
const double cpu_time, bool &success); const double cpu_time, bool &success);
void mie_gpu_compute(const int ago, const int inum, const int nall, void mie_gpu_compute(const int ago, const int inum, 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,
@ -67,7 +67,7 @@ PairMIECutGPU::PairMIECutGPU(LAMMPS *lmp) : PairMIECut(lmp), gpu_mode(GPU_FORCE)
respa_enable = 0; respa_enable = 0;
reinitflag = 0; reinitflag = 0;
cpu_time = 0.0; cpu_time = 0.0;
GPU_EXTRA::gpu_ready(lmp->modify, lmp->error); GPU_EXTRA::gpu_ready(lmp->modify, lmp->error);
} }
/* ---------------------------------------------------------------------- /* ----------------------------------------------------------------------
@ -85,10 +85,10 @@ void PairMIECutGPU::compute(int eflag, int vflag)
{ {
if (eflag || vflag) ev_setup(eflag,vflag); if (eflag || vflag) ev_setup(eflag,vflag);
else evflag = vflag_fdotr = 0; else evflag = vflag_fdotr = 0;
int nall = atom->nlocal + atom->nghost; int nall = atom->nlocal + atom->nghost;
int inum, host_start; int inum, host_start;
bool success = true; bool success = true;
int *ilist, *numneigh, **firstneigh; int *ilist, *numneigh, **firstneigh;
if (gpu_mode != GPU_FORCE) { if (gpu_mode != GPU_FORCE) {
@ -125,8 +125,8 @@ void PairMIECutGPU::compute(int eflag, int vflag)
void PairMIECutGPU::init_style() void PairMIECutGPU::init_style()
{ {
cut_respa = NULL; cut_respa = NULL;
if (force->newton_pair) if (force->newton_pair)
error->all(FLERR,"Cannot use newton pair with mie/cut/gpu pair style"); error->all(FLERR,"Cannot use newton pair with mie/cut/gpu pair style");
// Repeat cutsq calculation because done after call to init_style // Repeat cutsq calculation because done after call to init_style
@ -172,7 +172,7 @@ double PairMIECutGPU::memory_usage()
/* ---------------------------------------------------------------------- */ /* ---------------------------------------------------------------------- */
void PairMIECutGPU::cpu_compute(int start, int inum, int eflag, int vflag, void PairMIECutGPU::cpu_compute(int start, int inum, int eflag, int vflag,
int *ilist, int *numneigh, int **firstneigh) { int *ilist, int *numneigh, int **firstneigh) {
int i,j,ii,jj,jnum,itype,jtype; int i,j,ii,jj,jnum,itype,jtype;
double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair; double xtmp,ytmp,ztmp,delx,dely,delz,evdwl,fpair;

View File

@ -5,7 +5,7 @@
Copyright (2003) Sandia Corporation. Under the terms of Contract Copyright (2003) Sandia Corporation. Under the terms of Contract
DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains DE-AC04-94AL85000 with Sandia Corporation, the U.S. Government retains
certain rights in this software. This software is distributed under certain rights in this software. This software is distributed under
the GNU General Public License. the GNU General Public License.
See the README file in the top-level LAMMPS directory. See the README file in the top-level LAMMPS directory.

Some files were not shown because too many files have changed in this diff Show More