From 15a3364c2c68e62bfc724e135f73c09540d620fa Mon Sep 17 00:00:00 2001
From: Stan Moore <stamoor@sandia.gov>
Date: Fri, 1 Dec 2017 09:13:31 -0700
Subject: [PATCH] Make memory_kokkos its own class
---
src/KOKKOS/Install.sh | 1 -
src/KOKKOS/angle_charmm_kokkos.cpp | 14 +-
src/KOKKOS/angle_class2_kokkos.cpp | 14 +-
src/KOKKOS/angle_harmonic_kokkos.cpp | 14 +-
src/KOKKOS/atom_kokkos.cpp | 150 +++++++++---------
src/KOKKOS/atom_vec_angle_kokkos.cpp | 38 ++---
src/KOKKOS/atom_vec_atomic_kokkos.cpp | 16 +-
src/KOKKOS/atom_vec_bond_kokkos.cpp | 32 ++--
src/KOKKOS/atom_vec_charge_kokkos.cpp | 18 +--
src/KOKKOS/atom_vec_dpd_kokkos.cpp | 32 ++--
src/KOKKOS/atom_vec_full_kokkos.cpp | 64 ++++----
src/KOKKOS/atom_vec_hybrid_kokkos.cpp | 2 +-
src/KOKKOS/atom_vec_molecular_kokkos.cpp | 62 ++++----
src/KOKKOS/bond_class2_kokkos.cpp | 14 +-
src/KOKKOS/bond_fene_kokkos.cpp | 14 +-
src/KOKKOS/bond_harmonic_kokkos.cpp | 14 +-
src/KOKKOS/comm_kokkos.cpp | 14 +-
src/KOKKOS/comm_tiled_kokkos.cpp | 2 +-
src/KOKKOS/dihedral_charmm_kokkos.cpp | 14 +-
src/KOKKOS/dihedral_class2_kokkos.cpp | 14 +-
src/KOKKOS/dihedral_opls_kokkos.cpp | 14 +-
src/KOKKOS/fix_eos_table_rx_kokkos.cpp | 14 +-
src/KOKKOS/fix_langevin_kokkos.cpp | 30 ++--
src/KOKKOS/fix_nh_kokkos.cpp | 2 +-
src/KOKKOS/fix_property_atom_kokkos.cpp | 4 +-
src/KOKKOS/fix_qeq_reax_kokkos.cpp | 2 +-
src/KOKKOS/fix_reaxc_bonds_kokkos.cpp | 6 +-
src/KOKKOS/fix_reaxc_species_kokkos.cpp | 2 +-
src/KOKKOS/fix_rx_kokkos.cpp | 60 +++----
src/KOKKOS/fix_setforce_kokkos.cpp | 10 +-
src/KOKKOS/fix_shardlow_kokkos.cpp | 2 +-
src/KOKKOS/gridcomm_kokkos.cpp | 6 +-
src/KOKKOS/improper_class2_kokkos.cpp | 14 +-
src/KOKKOS/improper_harmonic_kokkos.cpp | 14 +-
src/KOKKOS/kokkos.cpp | 3 +
src/KOKKOS/memory_kokkos.h | 19 +++
src/KOKKOS/nbin_ssa_kokkos.cpp | 2 +-
src/KOKKOS/neigh_bond_kokkos.cpp | 26 +--
src/KOKKOS/neigh_list_kokkos.cpp | 2 +-
src/KOKKOS/neighbor_kokkos.cpp | 58 +++----
src/KOKKOS/pair_buck_coul_cut_kokkos.cpp | 20 +--
src/KOKKOS/pair_buck_coul_long_kokkos.cpp | 20 +--
src/KOKKOS/pair_buck_kokkos.cpp | 16 +-
src/KOKKOS/pair_coul_cut_kokkos.cpp | 14 +-
src/KOKKOS/pair_coul_debye_kokkos.cpp | 14 +-
src/KOKKOS/pair_coul_dsf_kokkos.cpp | 14 +-
src/KOKKOS/pair_coul_long_kokkos.cpp | 18 +--
src/KOKKOS/pair_coul_wolf_kokkos.cpp | 14 +-
src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp | 34 ++--
src/KOKKOS/pair_eam_alloy_kokkos.cpp | 14 +-
src/KOKKOS/pair_eam_fs_kokkos.cpp | 14 +-
src/KOKKOS/pair_eam_kokkos.cpp | 14 +-
src/KOKKOS/pair_exp6_rx_kokkos.cpp | 30 ++--
src/KOKKOS/pair_hybrid_kokkos.cpp | 2 +-
..._lj_charmm_coul_charmm_implicit_kokkos.cpp | 20 +--
.../pair_lj_charmm_coul_charmm_kokkos.cpp | 20 +--
.../pair_lj_charmm_coul_long_kokkos.cpp | 20 +--
src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp | 22 +--
.../pair_lj_class2_coul_long_kokkos.cpp | 20 +--
src/KOKKOS/pair_lj_class2_kokkos.cpp | 12 +-
src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp | 22 +--
src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp | 22 +--
src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp | 22 +--
src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp | 24 +--
src/KOKKOS/pair_lj_cut_kokkos.cpp | 16 +-
src/KOKKOS/pair_lj_expand_kokkos.cpp | 12 +-
.../pair_lj_gromacs_coul_gromacs_kokkos.cpp | 20 +--
src/KOKKOS/pair_lj_gromacs_kokkos.cpp | 20 +--
src/KOKKOS/pair_lj_sdk_kokkos.cpp | 12 +-
src/KOKKOS/pair_morse_kokkos.cpp | 16 +-
src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp | 38 ++---
src/KOKKOS/pair_reaxc_kokkos.cpp | 26 +--
src/KOKKOS/pair_sw_kokkos.cpp | 16 +-
src/KOKKOS/pair_table_kokkos.cpp | 60 +++----
src/KOKKOS/pair_table_rx_kokkos.cpp | 68 ++++----
src/KOKKOS/pair_tersoff_kokkos.cpp | 14 +-
src/KOKKOS/pair_tersoff_mod_kokkos.cpp | 14 +-
src/KOKKOS/pair_tersoff_zbl_kokkos.cpp | 14 +-
src/KOKKOS/pair_vashishta_kokkos.cpp | 16 +-
src/KOKKOS/pppm_kokkos.cpp | 28 ++--
src/KOKKOS/verlet_kokkos.cpp | 2 +-
src/accelerator_kokkos.h | 9 ++
src/lammps.cpp | 3 +
src/lammps.h | 1 +
src/memory.h | 13 --
src/pointers.h | 2 +
src/special.cpp | 3 +-
87 files changed, 861 insertions(+), 837 deletions(-)
diff --git a/src/KOKKOS/Install.sh b/src/KOKKOS/Install.sh
index df5fc3e5f1..f86ef990d5 100644
--- a/src/KOKKOS/Install.sh
+++ b/src/KOKKOS/Install.sh
@@ -29,7 +29,6 @@ action () {
# force rebuild of files with LMP_KOKKOS switch
touch ../accelerator_kokkos.h
-touch ../memory.h
# list of files with optional dependcies
diff --git a/src/KOKKOS/angle_charmm_kokkos.cpp b/src/KOKKOS/angle_charmm_kokkos.cpp
index 401a00c856..59a20c25df 100644
--- a/src/KOKKOS/angle_charmm_kokkos.cpp
+++ b/src/KOKKOS/angle_charmm_kokkos.cpp
@@ -24,7 +24,7 @@
#include "comm.h"
#include "force.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -51,8 +51,8 @@ template<class DeviceType>
AngleCharmmKokkos<DeviceType>::~AngleCharmmKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -71,15 +71,15 @@ void AngleCharmmKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
if (eflag_atom) {
//if(k_eatom.dimension_0()<maxeatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
d_eatom = k_eatom.template view<DeviceType>();
//}
}
if (vflag_atom) {
//if(k_vatom.dimension_0()<maxvatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
d_vatom = k_vatom.template view<DeviceType>();
//}
}
diff --git a/src/KOKKOS/angle_class2_kokkos.cpp b/src/KOKKOS/angle_class2_kokkos.cpp
index e851e9d500..108b4f48f2 100644
--- a/src/KOKKOS/angle_class2_kokkos.cpp
+++ b/src/KOKKOS/angle_class2_kokkos.cpp
@@ -24,7 +24,7 @@
#include "comm.h"
#include "force.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -51,8 +51,8 @@ template<class DeviceType>
AngleClass2Kokkos<DeviceType>::~AngleClass2Kokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -70,13 +70,13 @@ void AngleClass2Kokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"angle:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"angle:eatom");
d_eatom = k_eatom.template view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"angle:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"angle:vatom");
d_vatom = k_vatom.template view<DeviceType>();
}
diff --git a/src/KOKKOS/angle_harmonic_kokkos.cpp b/src/KOKKOS/angle_harmonic_kokkos.cpp
index 9fd237ddb3..dd5a1e26c7 100644
--- a/src/KOKKOS/angle_harmonic_kokkos.cpp
+++ b/src/KOKKOS/angle_harmonic_kokkos.cpp
@@ -24,7 +24,7 @@
#include "comm.h"
#include "force.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -51,8 +51,8 @@ template<class DeviceType>
AngleHarmonicKokkos<DeviceType>::~AngleHarmonicKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -70,13 +70,13 @@ void AngleHarmonicKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"angle:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"angle:eatom");
d_eatom = k_eatom.template view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"angle:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"angle:vatom");
d_vatom = k_vatom.template view<DeviceType>();
}
diff --git a/src/KOKKOS/atom_kokkos.cpp b/src/KOKKOS/atom_kokkos.cpp
index 31b33dbdc9..4ecead5b1d 100644
--- a/src/KOKKOS/atom_kokkos.cpp
+++ b/src/KOKKOS/atom_kokkos.cpp
@@ -19,7 +19,7 @@
#include "update.h"
#include "domain.h"
#include "atom_masks.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "kokkos.h"
@@ -33,59 +33,59 @@ AtomKokkos::AtomKokkos(LAMMPS *lmp) : Atom(lmp) {}
AtomKokkos::~AtomKokkos()
{
- memory->destroy_kokkos(k_tag, tag);
- memory->destroy_kokkos(k_mask, mask);
- memory->destroy_kokkos(k_type, type);
- memory->destroy_kokkos(k_image, image);
- memory->destroy_kokkos(k_molecule, molecule);
-
- memory->destroy_kokkos(k_x, x);
- memory->destroy_kokkos(k_v, v);
- memory->destroy_kokkos(k_f, f);
-
- memory->destroy_kokkos(k_mass, mass);
- memory->destroy_kokkos(k_q, q);
-
- memory->destroy_kokkos(k_radius, radius);
- memory->destroy_kokkos(k_rmass, rmass);
- memory->destroy_kokkos(k_omega, omega);
- memory->destroy_kokkos(k_angmom, angmom);
- memory->destroy_kokkos(k_torque, torque);
-
- memory->destroy_kokkos(k_nspecial, nspecial);
- memory->destroy_kokkos(k_special, special);
- memory->destroy_kokkos(k_num_bond, num_bond);
- memory->destroy_kokkos(k_bond_type, bond_type);
- memory->destroy_kokkos(k_bond_atom, bond_atom);
- memory->destroy_kokkos(k_num_angle, num_angle);
- memory->destroy_kokkos(k_angle_type, angle_type);
- memory->destroy_kokkos(k_angle_atom1, angle_atom1);
- memory->destroy_kokkos(k_angle_atom2, angle_atom2);
- memory->destroy_kokkos(k_angle_atom3, angle_atom3);
- memory->destroy_kokkos(k_num_dihedral, num_dihedral);
- memory->destroy_kokkos(k_dihedral_type, dihedral_type);
- memory->destroy_kokkos(k_dihedral_atom1, dihedral_atom1);
- memory->destroy_kokkos(k_dihedral_atom2, dihedral_atom2);
- memory->destroy_kokkos(k_dihedral_atom3, dihedral_atom3);
- memory->destroy_kokkos(k_dihedral_atom4, dihedral_atom4);
- memory->destroy_kokkos(k_num_improper, num_improper);
- memory->destroy_kokkos(k_improper_type, improper_type);
- memory->destroy_kokkos(k_improper_atom1, improper_atom1);
- memory->destroy_kokkos(k_improper_atom2, improper_atom2);
- memory->destroy_kokkos(k_improper_atom3, improper_atom3);
- memory->destroy_kokkos(k_improper_atom4, improper_atom4);
+ memoryKK->destroy_kokkos(k_tag, tag);
+ memoryKK->destroy_kokkos(k_mask, mask);
+ memoryKK->destroy_kokkos(k_type, type);
+ memoryKK->destroy_kokkos(k_image, image);
+ memoryKK->destroy_kokkos(k_molecule, molecule);
+
+ memoryKK->destroy_kokkos(k_x, x);
+ memoryKK->destroy_kokkos(k_v, v);
+ memoryKK->destroy_kokkos(k_f, f);
+
+ memoryKK->destroy_kokkos(k_mass, mass);
+ memoryKK->destroy_kokkos(k_q, q);
+
+ memoryKK->destroy_kokkos(k_radius, radius);
+ memoryKK->destroy_kokkos(k_rmass, rmass);
+ memoryKK->destroy_kokkos(k_omega, omega);
+ memoryKK->destroy_kokkos(k_angmom, angmom);
+ memoryKK->destroy_kokkos(k_torque, torque);
+
+ memoryKK->destroy_kokkos(k_nspecial, nspecial);
+ memoryKK->destroy_kokkos(k_special, special);
+ memoryKK->destroy_kokkos(k_num_bond, num_bond);
+ memoryKK->destroy_kokkos(k_bond_type, bond_type);
+ memoryKK->destroy_kokkos(k_bond_atom, bond_atom);
+ memoryKK->destroy_kokkos(k_num_angle, num_angle);
+ memoryKK->destroy_kokkos(k_angle_type, angle_type);
+ memoryKK->destroy_kokkos(k_angle_atom1, angle_atom1);
+ memoryKK->destroy_kokkos(k_angle_atom2, angle_atom2);
+ memoryKK->destroy_kokkos(k_angle_atom3, angle_atom3);
+ memoryKK->destroy_kokkos(k_num_dihedral, num_dihedral);
+ memoryKK->destroy_kokkos(k_dihedral_type, dihedral_type);
+ memoryKK->destroy_kokkos(k_dihedral_atom1, dihedral_atom1);
+ memoryKK->destroy_kokkos(k_dihedral_atom2, dihedral_atom2);
+ memoryKK->destroy_kokkos(k_dihedral_atom3, dihedral_atom3);
+ memoryKK->destroy_kokkos(k_dihedral_atom4, dihedral_atom4);
+ memoryKK->destroy_kokkos(k_num_improper, num_improper);
+ memoryKK->destroy_kokkos(k_improper_type, improper_type);
+ memoryKK->destroy_kokkos(k_improper_atom1, improper_atom1);
+ memoryKK->destroy_kokkos(k_improper_atom2, improper_atom2);
+ memoryKK->destroy_kokkos(k_improper_atom3, improper_atom3);
+ memoryKK->destroy_kokkos(k_improper_atom4, improper_atom4);
// USER-DPD package
- memory->destroy_kokkos(k_uCond,uCond);
- memory->destroy_kokkos(k_uMech,uMech);
- memory->destroy_kokkos(k_uChem,uChem);
- memory->destroy_kokkos(k_uCG,uCG);
- memory->destroy_kokkos(k_uCGnew,uCGnew);
- memory->destroy_kokkos(k_rho,rho);
- memory->destroy_kokkos(k_dpdTheta,dpdTheta);
- memory->destroy_kokkos(k_duChem,duChem);
-
- memory->destroy_kokkos(k_dvector,dvector);
+ memoryKK->destroy_kokkos(k_uCond,uCond);
+ memoryKK->destroy_kokkos(k_uMech,uMech);
+ memoryKK->destroy_kokkos(k_uChem,uChem);
+ memoryKK->destroy_kokkos(k_uCG,uCG);
+ memoryKK->destroy_kokkos(k_uCGnew,uCGnew);
+ memoryKK->destroy_kokkos(k_rho,rho);
+ memoryKK->destroy_kokkos(k_dpdTheta,dpdTheta);
+ memoryKK->destroy_kokkos(k_duChem,duChem);
+
+ memoryKK->destroy_kokkos(k_dvector,dvector);
dvector = NULL;
}
@@ -232,10 +232,10 @@ void AtomKokkos::sort()
void AtomKokkos::grow(unsigned int mask){
if (mask & SPECIAL_MASK){
- memory->destroy_kokkos(k_special, special);
+ memoryKK->destroy_kokkos(k_special, special);
sync(Device, mask);
modified(Device, mask);
- memory->grow_kokkos(k_special,special,nmax,maxspecial,"atom:special");
+ memoryKK->grow_kokkos(k_special,special,nmax,maxspecial,"atom:special");
avec->grow_reset();
sync(Host, mask);
}
@@ -270,7 +270,7 @@ int AtomKokkos::add_custom(const char *name, int flag)
int n = strlen(name) + 1;
dname[index] = new char[n];
strcpy(dname[index],name);
- memory->grow_kokkos(k_dvector,dvector,ndvector,nmax,
+ memoryKK->grow_kokkos(k_dvector,dvector,ndvector,nmax,
"atom:dvector");
}
@@ -291,7 +291,7 @@ void AtomKokkos::remove_custom(int flag, int index)
delete [] iname[index];
iname[index] = NULL;
} else {
- //memory->destroy_kokkos(dvector);
+ //memoryKK->destroy_kokkos(dvector);
dvector[index] = NULL;
delete [] dname[index];
dname[index] = NULL;
@@ -302,25 +302,25 @@ void AtomKokkos::remove_custom(int flag, int index)
void AtomKokkos::deallocate_topology()
{
- memory->destroy_kokkos(k_bond_type, bond_type);
- memory->destroy_kokkos(k_bond_atom, bond_atom);
-
- memory->destroy_kokkos(k_angle_type, angle_type);
- memory->destroy_kokkos(k_angle_atom1, angle_atom1);
- memory->destroy_kokkos(k_angle_atom2, angle_atom2);
- memory->destroy_kokkos(k_angle_atom3, angle_atom3);
-
- memory->destroy_kokkos(k_dihedral_type, dihedral_type);
- memory->destroy_kokkos(k_dihedral_atom1, dihedral_atom1);
- memory->destroy_kokkos(k_dihedral_atom2, dihedral_atom2);
- memory->destroy_kokkos(k_dihedral_atom3, dihedral_atom3);
- memory->destroy_kokkos(k_dihedral_atom4, dihedral_atom4);
-
- memory->destroy_kokkos(k_improper_type, improper_type);
- memory->destroy_kokkos(k_improper_atom1, improper_atom1);
- memory->destroy_kokkos(k_improper_atom2, improper_atom2);
- memory->destroy_kokkos(k_improper_atom3, improper_atom3);
- memory->destroy_kokkos(k_improper_atom4, improper_atom4);
+ memoryKK->destroy_kokkos(k_bond_type, bond_type);
+ memoryKK->destroy_kokkos(k_bond_atom, bond_atom);
+
+ memoryKK->destroy_kokkos(k_angle_type, angle_type);
+ memoryKK->destroy_kokkos(k_angle_atom1, angle_atom1);
+ memoryKK->destroy_kokkos(k_angle_atom2, angle_atom2);
+ memoryKK->destroy_kokkos(k_angle_atom3, angle_atom3);
+
+ memoryKK->destroy_kokkos(k_dihedral_type, dihedral_type);
+ memoryKK->destroy_kokkos(k_dihedral_atom1, dihedral_atom1);
+ memoryKK->destroy_kokkos(k_dihedral_atom2, dihedral_atom2);
+ memoryKK->destroy_kokkos(k_dihedral_atom3, dihedral_atom3);
+ memoryKK->destroy_kokkos(k_dihedral_atom4, dihedral_atom4);
+
+ memoryKK->destroy_kokkos(k_improper_type, improper_type);
+ memoryKK->destroy_kokkos(k_improper_atom1, improper_atom1);
+ memoryKK->destroy_kokkos(k_improper_atom2, improper_atom2);
+ memoryKK->destroy_kokkos(k_improper_atom3, improper_atom3);
+ memoryKK->destroy_kokkos(k_improper_atom4, improper_atom4);
}
/* ----------------------------------------------------------------------
diff --git a/src/KOKKOS/atom_vec_angle_kokkos.cpp b/src/KOKKOS/atom_vec_angle_kokkos.cpp
index 05414cf2e4..a9e55f530a 100644
--- a/src/KOKKOS/atom_vec_angle_kokkos.cpp
+++ b/src/KOKKOS/atom_vec_angle_kokkos.cpp
@@ -19,7 +19,7 @@
#include "modify.h"
#include "fix.h"
#include "atom_masks.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
using namespace LAMMPS_NS;
@@ -68,33 +68,33 @@ void AtomVecAngleKokkos::grow(int n)
sync(Device,ALL_MASK);
modified(Device,ALL_MASK);
- memory->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
- memory->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
- memory->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
- memory->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
+ memoryKK->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
+ memoryKK->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
+ memoryKK->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
+ memoryKK->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
- memory->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
- memory->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
- memory->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
+ memoryKK->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
+ memoryKK->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
+ memoryKK->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
- memory->grow_kokkos(atomKK->k_molecule,atomKK->molecule,nmax,"atom:molecule");
- memory->grow_kokkos(atomKK->k_nspecial,atomKK->nspecial,nmax,3,"atom:nspecial");
- memory->grow_kokkos(atomKK->k_special,atomKK->special,nmax,atomKK->maxspecial,
+ memoryKK->grow_kokkos(atomKK->k_molecule,atomKK->molecule,nmax,"atom:molecule");
+ memoryKK->grow_kokkos(atomKK->k_nspecial,atomKK->nspecial,nmax,3,"atom:nspecial");
+ memoryKK->grow_kokkos(atomKK->k_special,atomKK->special,nmax,atomKK->maxspecial,
"atom:special");
- memory->grow_kokkos(atomKK->k_num_bond,atomKK->num_bond,nmax,"atom:num_bond");
- memory->grow_kokkos(atomKK->k_bond_type,atomKK->bond_type,nmax,atomKK->bond_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_num_bond,atomKK->num_bond,nmax,"atom:num_bond");
+ memoryKK->grow_kokkos(atomKK->k_bond_type,atomKK->bond_type,nmax,atomKK->bond_per_atom,
"atom:bond_type");
- memory->grow_kokkos(atomKK->k_bond_atom,atomKK->bond_atom,nmax,atomKK->bond_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_bond_atom,atomKK->bond_atom,nmax,atomKK->bond_per_atom,
"atom:bond_atom");
- memory->grow_kokkos(atomKK->k_num_angle,atomKK->num_angle,nmax,"atom:num_angle");
- memory->grow_kokkos(atomKK->k_angle_type,atomKK->angle_type,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_num_angle,atomKK->num_angle,nmax,"atom:num_angle");
+ memoryKK->grow_kokkos(atomKK->k_angle_type,atomKK->angle_type,nmax,atomKK->angle_per_atom,
"atom:angle_type");
- memory->grow_kokkos(atomKK->k_angle_atom1,atomKK->angle_atom1,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom1,atomKK->angle_atom1,nmax,atomKK->angle_per_atom,
"atom:angle_atom1");
- memory->grow_kokkos(atomKK->k_angle_atom2,atomKK->angle_atom2,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom2,atomKK->angle_atom2,nmax,atomKK->angle_per_atom,
"atom:angle_atom2");
- memory->grow_kokkos(atomKK->k_angle_atom3,atomKK->angle_atom3,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom3,atomKK->angle_atom3,nmax,atomKK->angle_per_atom,
"atom:angle_atom3");
grow_reset();
diff --git a/src/KOKKOS/atom_vec_atomic_kokkos.cpp b/src/KOKKOS/atom_vec_atomic_kokkos.cpp
index 6c610c8c11..f021c45db6 100644
--- a/src/KOKKOS/atom_vec_atomic_kokkos.cpp
+++ b/src/KOKKOS/atom_vec_atomic_kokkos.cpp
@@ -19,7 +19,7 @@
#include "modify.h"
#include "fix.h"
#include "atom_masks.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
using namespace LAMMPS_NS;
@@ -64,14 +64,14 @@ void AtomVecAtomicKokkos::grow(int n)
sync(Device,ALL_MASK);
modified(Device,ALL_MASK);
- memory->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
- memory->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
- memory->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
- memory->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
+ memoryKK->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
+ memoryKK->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
+ memoryKK->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
+ memoryKK->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
- memory->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
- memory->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
- memory->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
+ memoryKK->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
+ memoryKK->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
+ memoryKK->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
grow_reset();
sync(Host,ALL_MASK);
diff --git a/src/KOKKOS/atom_vec_bond_kokkos.cpp b/src/KOKKOS/atom_vec_bond_kokkos.cpp
index 076144420c..bf682c507f 100644
--- a/src/KOKKOS/atom_vec_bond_kokkos.cpp
+++ b/src/KOKKOS/atom_vec_bond_kokkos.cpp
@@ -19,7 +19,7 @@
#include "modify.h"
#include "fix.h"
#include "atom_masks.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
using namespace LAMMPS_NS;
@@ -65,21 +65,21 @@ void AtomVecBondKokkos::grow(int n)
sync(Device,ALL_MASK);
modified(Device,ALL_MASK);
- memory->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
- memory->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
- memory->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
- memory->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
-
- memory->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
- memory->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
- memory->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
-
- memory->grow_kokkos(atomKK->k_molecule,atomKK->molecule,nmax,"atom:molecule");
- memory->grow_kokkos(atomKK->k_nspecial,atomKK->nspecial,nmax,3,"atom:nspecial");
- memory->grow_kokkos(atomKK->k_special,atomKK->special,nmax,atomKK->maxspecial,"atom:special");
- memory->grow_kokkos(atomKK->k_num_bond,atomKK->num_bond,nmax,"atom:num_bond");
- memory->grow_kokkos(atomKK->k_bond_type,atomKK->bond_type,nmax,atomKK->bond_per_atom,"atom:bond_type");
- memory->grow_kokkos(atomKK->k_bond_atom,atomKK->bond_atom,nmax,atomKK->bond_per_atom,"atom:bond_atom");
+ memoryKK->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
+ memoryKK->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
+ memoryKK->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
+ memoryKK->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
+
+ memoryKK->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
+ memoryKK->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
+ memoryKK->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
+
+ memoryKK->grow_kokkos(atomKK->k_molecule,atomKK->molecule,nmax,"atom:molecule");
+ memoryKK->grow_kokkos(atomKK->k_nspecial,atomKK->nspecial,nmax,3,"atom:nspecial");
+ memoryKK->grow_kokkos(atomKK->k_special,atomKK->special,nmax,atomKK->maxspecial,"atom:special");
+ memoryKK->grow_kokkos(atomKK->k_num_bond,atomKK->num_bond,nmax,"atom:num_bond");
+ memoryKK->grow_kokkos(atomKK->k_bond_type,atomKK->bond_type,nmax,atomKK->bond_per_atom,"atom:bond_type");
+ memoryKK->grow_kokkos(atomKK->k_bond_atom,atomKK->bond_atom,nmax,atomKK->bond_per_atom,"atom:bond_atom");
grow_reset();
sync(Host,ALL_MASK);
diff --git a/src/KOKKOS/atom_vec_charge_kokkos.cpp b/src/KOKKOS/atom_vec_charge_kokkos.cpp
index 7b8b74b405..a9ae5cc2d1 100644
--- a/src/KOKKOS/atom_vec_charge_kokkos.cpp
+++ b/src/KOKKOS/atom_vec_charge_kokkos.cpp
@@ -19,7 +19,7 @@
#include "modify.h"
#include "fix.h"
#include "atom_masks.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
using namespace LAMMPS_NS;
@@ -67,16 +67,16 @@ void AtomVecChargeKokkos::grow(int n)
sync(Device,ALL_MASK);
modified(Device,ALL_MASK);
- memory->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
- memory->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
- memory->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
- memory->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
+ memoryKK->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
+ memoryKK->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
+ memoryKK->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
+ memoryKK->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
- memory->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
- memory->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
- memory->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
+ memoryKK->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
+ memoryKK->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
+ memoryKK->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
- memory->grow_kokkos(atomKK->k_q,atomKK->q,nmax,"atom:q");
+ memoryKK->grow_kokkos(atomKK->k_q,atomKK->q,nmax,"atom:q");
grow_reset();
sync(Host,ALL_MASK);
diff --git a/src/KOKKOS/atom_vec_dpd_kokkos.cpp b/src/KOKKOS/atom_vec_dpd_kokkos.cpp
index c4e493bd85..9c54ffccc5 100644
--- a/src/KOKKOS/atom_vec_dpd_kokkos.cpp
+++ b/src/KOKKOS/atom_vec_dpd_kokkos.cpp
@@ -19,7 +19,7 @@
#include "modify.h"
#include "fix.h"
#include "atom_masks.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
using namespace LAMMPS_NS;
@@ -67,24 +67,24 @@ void AtomVecDPDKokkos::grow(int n)
sync(Device,ALL_MASK);
modified(Device,ALL_MASK);
- memory->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
- memory->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
- memory->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
- memory->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
+ memoryKK->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
+ memoryKK->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
+ memoryKK->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
+ memoryKK->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
- memory->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
- memory->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
- memory->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
+ memoryKK->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
+ memoryKK->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
+ memoryKK->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
- memory->grow_kokkos(atomKK->k_rho,atomKK->rho,nmax,"atom:rho");
- memory->grow_kokkos(atomKK->k_dpdTheta,atomKK->dpdTheta,nmax,"atom:dpdTheta");
- memory->grow_kokkos(atomKK->k_uCond,atomKK->uCond,nmax,"atom:uCond");
- memory->grow_kokkos(atomKK->k_uMech,atomKK->uMech,nmax,"atom:uMech");
- memory->grow_kokkos(atomKK->k_uChem,atomKK->uChem,nmax,"atom:uChem");
- memory->grow_kokkos(atomKK->k_uCG,atomKK->uCG,nmax,"atom:uCG");
- memory->grow_kokkos(atomKK->k_uCGnew,atomKK->uCGnew,nmax,"atom:uCGnew");
- memory->grow_kokkos(atomKK->k_duChem,atomKK->duChem,nmax,"atom:duChem");
+ memoryKK->grow_kokkos(atomKK->k_rho,atomKK->rho,nmax,"atom:rho");
+ memoryKK->grow_kokkos(atomKK->k_dpdTheta,atomKK->dpdTheta,nmax,"atom:dpdTheta");
+ memoryKK->grow_kokkos(atomKK->k_uCond,atomKK->uCond,nmax,"atom:uCond");
+ memoryKK->grow_kokkos(atomKK->k_uMech,atomKK->uMech,nmax,"atom:uMech");
+ memoryKK->grow_kokkos(atomKK->k_uChem,atomKK->uChem,nmax,"atom:uChem");
+ memoryKK->grow_kokkos(atomKK->k_uCG,atomKK->uCG,nmax,"atom:uCG");
+ memoryKK->grow_kokkos(atomKK->k_uCGnew,atomKK->uCGnew,nmax,"atom:uCGnew");
+ memoryKK->grow_kokkos(atomKK->k_duChem,atomKK->duChem,nmax,"atom:duChem");
if (atom->nextra_grow)
for (int iextra = 0; iextra < atom->nextra_grow; iextra++)
diff --git a/src/KOKKOS/atom_vec_full_kokkos.cpp b/src/KOKKOS/atom_vec_full_kokkos.cpp
index 8e9abe4067..9369d7e844 100644
--- a/src/KOKKOS/atom_vec_full_kokkos.cpp
+++ b/src/KOKKOS/atom_vec_full_kokkos.cpp
@@ -19,7 +19,7 @@
#include "modify.h"
#include "fix.h"
#include "atom_masks.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
using namespace LAMMPS_NS;
@@ -67,59 +67,59 @@ void AtomVecFullKokkos::grow(int n)
sync(Device,ALL_MASK);
modified(Device,ALL_MASK);
- memory->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
- memory->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
- memory->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
- memory->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
+ memoryKK->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
+ memoryKK->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
+ memoryKK->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
+ memoryKK->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
- memory->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
- memory->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
- memory->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
+ memoryKK->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
+ memoryKK->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
+ memoryKK->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
- memory->grow_kokkos(atomKK->k_q,atomKK->q,nmax,"atom:q");
- memory->grow_kokkos(atomKK->k_molecule,atomKK->molecule,nmax,"atom:molecule");
+ memoryKK->grow_kokkos(atomKK->k_q,atomKK->q,nmax,"atom:q");
+ memoryKK->grow_kokkos(atomKK->k_molecule,atomKK->molecule,nmax,"atom:molecule");
- memory->grow_kokkos(atomKK->k_nspecial,atomKK->nspecial,nmax,3,"atom:nspecial");
- memory->grow_kokkos(atomKK->k_special,atomKK->special,nmax,atomKK->maxspecial,
+ memoryKK->grow_kokkos(atomKK->k_nspecial,atomKK->nspecial,nmax,3,"atom:nspecial");
+ memoryKK->grow_kokkos(atomKK->k_special,atomKK->special,nmax,atomKK->maxspecial,
"atom:special");
- memory->grow_kokkos(atomKK->k_num_bond,atomKK->num_bond,nmax,"atom:num_bond");
- memory->grow_kokkos(atomKK->k_bond_type,atomKK->bond_type,nmax,atomKK->bond_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_num_bond,atomKK->num_bond,nmax,"atom:num_bond");
+ memoryKK->grow_kokkos(atomKK->k_bond_type,atomKK->bond_type,nmax,atomKK->bond_per_atom,
"atom:bond_type");
- memory->grow_kokkos(atomKK->k_bond_atom,atomKK->bond_atom,nmax,atomKK->bond_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_bond_atom,atomKK->bond_atom,nmax,atomKK->bond_per_atom,
"atom:bond_atom");
- memory->grow_kokkos(atomKK->k_num_angle,atomKK->num_angle,nmax,"atom:num_angle");
- memory->grow_kokkos(atomKK->k_angle_type,atomKK->angle_type,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_num_angle,atomKK->num_angle,nmax,"atom:num_angle");
+ memoryKK->grow_kokkos(atomKK->k_angle_type,atomKK->angle_type,nmax,atomKK->angle_per_atom,
"atom:angle_type");
- memory->grow_kokkos(atomKK->k_angle_atom1,atomKK->angle_atom1,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom1,atomKK->angle_atom1,nmax,atomKK->angle_per_atom,
"atom:angle_atom1");
- memory->grow_kokkos(atomKK->k_angle_atom2,atomKK->angle_atom2,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom2,atomKK->angle_atom2,nmax,atomKK->angle_per_atom,
"atom:angle_atom2");
- memory->grow_kokkos(atomKK->k_angle_atom3,atomKK->angle_atom3,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom3,atomKK->angle_atom3,nmax,atomKK->angle_per_atom,
"atom:angle_atom3");
- memory->grow_kokkos(atomKK->k_num_dihedral,atomKK->num_dihedral,nmax,"atom:num_dihedral");
- memory->grow_kokkos(atomKK->k_dihedral_type,atomKK->dihedral_type,nmax,
+ memoryKK->grow_kokkos(atomKK->k_num_dihedral,atomKK->num_dihedral,nmax,"atom:num_dihedral");
+ memoryKK->grow_kokkos(atomKK->k_dihedral_type,atomKK->dihedral_type,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_type");
- memory->grow_kokkos(atomKK->k_dihedral_atom1,atomKK->dihedral_atom1,nmax,
+ memoryKK->grow_kokkos(atomKK->k_dihedral_atom1,atomKK->dihedral_atom1,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_atom1");
- memory->grow_kokkos(atomKK->k_dihedral_atom2,atomKK->dihedral_atom2,nmax,
+ memoryKK->grow_kokkos(atomKK->k_dihedral_atom2,atomKK->dihedral_atom2,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_atom2");
- memory->grow_kokkos(atomKK->k_dihedral_atom3,atomKK->dihedral_atom3,nmax,
+ memoryKK->grow_kokkos(atomKK->k_dihedral_atom3,atomKK->dihedral_atom3,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_atom3");
- memory->grow_kokkos(atomKK->k_dihedral_atom4,atomKK->dihedral_atom4,nmax,
+ memoryKK->grow_kokkos(atomKK->k_dihedral_atom4,atomKK->dihedral_atom4,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_atom4");
- memory->grow_kokkos(atomKK->k_num_improper,atomKK->num_improper,nmax,"atom:num_improper");
- memory->grow_kokkos(atomKK->k_improper_type,atomKK->improper_type,nmax,
+ memoryKK->grow_kokkos(atomKK->k_num_improper,atomKK->num_improper,nmax,"atom:num_improper");
+ memoryKK->grow_kokkos(atomKK->k_improper_type,atomKK->improper_type,nmax,
atomKK->improper_per_atom,"atom:improper_type");
- memory->grow_kokkos(atomKK->k_improper_atom1,atomKK->improper_atom1,nmax,
+ memoryKK->grow_kokkos(atomKK->k_improper_atom1,atomKK->improper_atom1,nmax,
atomKK->improper_per_atom,"atom:improper_atom1");
- memory->grow_kokkos(atomKK->k_improper_atom2,atomKK->improper_atom2,nmax,
+ memoryKK->grow_kokkos(atomKK->k_improper_atom2,atomKK->improper_atom2,nmax,
atomKK->improper_per_atom,"atom:improper_atom2");
- memory->grow_kokkos(atomKK->k_improper_atom3,atomKK->improper_atom3,nmax,
+ memoryKK->grow_kokkos(atomKK->k_improper_atom3,atomKK->improper_atom3,nmax,
atomKK->improper_per_atom,"atom:improper_atom3");
- memory->grow_kokkos(atomKK->k_improper_atom4,atomKK->improper_atom4,nmax,
+ memoryKK->grow_kokkos(atomKK->k_improper_atom4,atomKK->improper_atom4,nmax,
atomKK->improper_per_atom,"atom:improper_atom4");
grow_reset();
diff --git a/src/KOKKOS/atom_vec_hybrid_kokkos.cpp b/src/KOKKOS/atom_vec_hybrid_kokkos.cpp
index e5e361e70a..b5aadb18d6 100644
--- a/src/KOKKOS/atom_vec_hybrid_kokkos.cpp
+++ b/src/KOKKOS/atom_vec_hybrid_kokkos.cpp
@@ -18,7 +18,7 @@
#include "domain.h"
#include "modify.h"
#include "fix.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
diff --git a/src/KOKKOS/atom_vec_molecular_kokkos.cpp b/src/KOKKOS/atom_vec_molecular_kokkos.cpp
index dbf6a857b2..6f232a319b 100644
--- a/src/KOKKOS/atom_vec_molecular_kokkos.cpp
+++ b/src/KOKKOS/atom_vec_molecular_kokkos.cpp
@@ -19,7 +19,7 @@
#include "modify.h"
#include "fix.h"
#include "atom_masks.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
using namespace LAMMPS_NS;
@@ -67,57 +67,57 @@ void AtomVecMolecularKokkos::grow(int n)
sync(Device,ALL_MASK);
modified(Device,ALL_MASK);
- memory->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
- memory->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
- memory->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
- memory->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
+ memoryKK->grow_kokkos(atomKK->k_tag,atomKK->tag,nmax,"atom:tag");
+ memoryKK->grow_kokkos(atomKK->k_type,atomKK->type,nmax,"atom:type");
+ memoryKK->grow_kokkos(atomKK->k_mask,atomKK->mask,nmax,"atom:mask");
+ memoryKK->grow_kokkos(atomKK->k_image,atomKK->image,nmax,"atom:image");
- memory->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
- memory->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
- memory->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
+ memoryKK->grow_kokkos(atomKK->k_x,atomKK->x,nmax,3,"atom:x");
+ memoryKK->grow_kokkos(atomKK->k_v,atomKK->v,nmax,3,"atom:v");
+ memoryKK->grow_kokkos(atomKK->k_f,atomKK->f,nmax,3,"atom:f");
- memory->grow_kokkos(atomKK->k_molecule,atomKK->molecule,nmax,"atom:molecule");
- memory->grow_kokkos(atomKK->k_nspecial,atomKK->nspecial,nmax,3,"atom:nspecial");
- memory->grow_kokkos(atomKK->k_special,atomKK->special,nmax,atomKK->maxspecial,
+ memoryKK->grow_kokkos(atomKK->k_molecule,atomKK->molecule,nmax,"atom:molecule");
+ memoryKK->grow_kokkos(atomKK->k_nspecial,atomKK->nspecial,nmax,3,"atom:nspecial");
+ memoryKK->grow_kokkos(atomKK->k_special,atomKK->special,nmax,atomKK->maxspecial,
"atom:special");
- memory->grow_kokkos(atomKK->k_num_bond,atomKK->num_bond,nmax,"atom:num_bond");
- memory->grow_kokkos(atomKK->k_bond_type,atomKK->bond_type,nmax,atomKK->bond_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_num_bond,atomKK->num_bond,nmax,"atom:num_bond");
+ memoryKK->grow_kokkos(atomKK->k_bond_type,atomKK->bond_type,nmax,atomKK->bond_per_atom,
"atom:bond_type");
- memory->grow_kokkos(atomKK->k_bond_atom,atomKK->bond_atom,nmax,atomKK->bond_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_bond_atom,atomKK->bond_atom,nmax,atomKK->bond_per_atom,
"atom:bond_atom");
- memory->grow_kokkos(atomKK->k_num_angle,atomKK->num_angle,nmax,"atom:num_angle");
- memory->grow_kokkos(atomKK->k_angle_type,atomKK->angle_type,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_num_angle,atomKK->num_angle,nmax,"atom:num_angle");
+ memoryKK->grow_kokkos(atomKK->k_angle_type,atomKK->angle_type,nmax,atomKK->angle_per_atom,
"atom:angle_type");
- memory->grow_kokkos(atomKK->k_angle_atom1,atomKK->angle_atom1,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom1,atomKK->angle_atom1,nmax,atomKK->angle_per_atom,
"atom:angle_atom1");
- memory->grow_kokkos(atomKK->k_angle_atom2,atomKK->angle_atom2,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom2,atomKK->angle_atom2,nmax,atomKK->angle_per_atom,
"atom:angle_atom2");
- memory->grow_kokkos(atomKK->k_angle_atom3,atomKK->angle_atom3,nmax,atomKK->angle_per_atom,
+ memoryKK->grow_kokkos(atomKK->k_angle_atom3,atomKK->angle_atom3,nmax,atomKK->angle_per_atom,
"atom:angle_atom3");
- memory->grow_kokkos(atomKK->k_num_dihedral,atomKK->num_dihedral,nmax,"atom:num_dihedral");
- memory->grow_kokkos(atomKK->k_dihedral_type,atomKK->dihedral_type,nmax,
+ memoryKK->grow_kokkos(atomKK->k_num_dihedral,atomKK->num_dihedral,nmax,"atom:num_dihedral");
+ memoryKK->grow_kokkos(atomKK->k_dihedral_type,atomKK->dihedral_type,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_type");
- memory->grow_kokkos(atomKK->k_dihedral_atom1,atomKK->dihedral_atom1,nmax,
+ memoryKK->grow_kokkos(atomKK->k_dihedral_atom1,atomKK->dihedral_atom1,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_atom1");
- memory->grow_kokkos(atomKK->k_dihedral_atom2,atomKK->dihedral_atom2,nmax,
+ memoryKK->grow_kokkos(atomKK->k_dihedral_atom2,atomKK->dihedral_atom2,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_atom2");
- memory->grow_kokkos(atomKK->k_dihedral_atom3,atomKK->dihedral_atom3,nmax,
+ memoryKK->grow_kokkos(atomKK->k_dihedral_atom3,atomKK->dihedral_atom3,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_atom3");
- memory->grow_kokkos(atomKK->k_dihedral_atom4,atomKK->dihedral_atom4,nmax,
+ memoryKK->grow_kokkos(atomKK->k_dihedral_atom4,atomKK->dihedral_atom4,nmax,
atomKK->dihedral_per_atom,"atom:dihedral_atom4");
- memory->grow_kokkos(atomKK->k_num_improper,atomKK->num_improper,nmax,"atom:num_improper");
- memory->grow_kokkos(atomKK->k_improper_type,atomKK->improper_type,nmax,
+ memoryKK->grow_kokkos(atomKK->k_num_improper,atomKK->num_improper,nmax,"atom:num_improper");
+ memoryKK->grow_kokkos(atomKK->k_improper_type,atomKK->improper_type,nmax,
atomKK->improper_per_atom,"atom:improper_type");
- memory->grow_kokkos(atomKK->k_improper_atom1,atomKK->improper_atom1,nmax,
+ memoryKK->grow_kokkos(atomKK->k_improper_atom1,atomKK->improper_atom1,nmax,
atomKK->improper_per_atom,"atom:improper_atom1");
- memory->grow_kokkos(atomKK->k_improper_atom2,atomKK->improper_atom2,nmax,
+ memoryKK->grow_kokkos(atomKK->k_improper_atom2,atomKK->improper_atom2,nmax,
atomKK->improper_per_atom,"atom:improper_atom2");
- memory->grow_kokkos(atomKK->k_improper_atom3,atomKK->improper_atom3,nmax,
+ memoryKK->grow_kokkos(atomKK->k_improper_atom3,atomKK->improper_atom3,nmax,
atomKK->improper_per_atom,"atom:improper_atom3");
- memory->grow_kokkos(atomKK->k_improper_atom4,atomKK->improper_atom4,nmax,
+ memoryKK->grow_kokkos(atomKK->k_improper_atom4,atomKK->improper_atom4,nmax,
atomKK->improper_per_atom,"atom:improper_atom4");
grow_reset();
diff --git a/src/KOKKOS/bond_class2_kokkos.cpp b/src/KOKKOS/bond_class2_kokkos.cpp
index b3c11c9a06..df2f2c1e9b 100644
--- a/src/KOKKOS/bond_class2_kokkos.cpp
+++ b/src/KOKKOS/bond_class2_kokkos.cpp
@@ -23,7 +23,7 @@
#include "domain.h"
#include "comm.h"
#include "force.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -47,8 +47,8 @@ template<class DeviceType>
BondClass2Kokkos<DeviceType>::~BondClass2Kokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -67,15 +67,15 @@ void BondClass2Kokkos<DeviceType>::compute(int eflag_in, int vflag_in)
if (eflag_atom) {
//if(k_eatom.dimension_0()<maxeatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
d_eatom = k_eatom.template view<DeviceType>();
//}
}
if (vflag_atom) {
//if(k_vatom.dimension_0()<maxvatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
d_vatom = k_vatom.template view<DeviceType>();
//}
}
diff --git a/src/KOKKOS/bond_fene_kokkos.cpp b/src/KOKKOS/bond_fene_kokkos.cpp
index 8a716a98ef..20c2054208 100644
--- a/src/KOKKOS/bond_fene_kokkos.cpp
+++ b/src/KOKKOS/bond_fene_kokkos.cpp
@@ -23,7 +23,7 @@
#include "domain.h"
#include "comm.h"
#include "force.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -56,8 +56,8 @@ template<class DeviceType>
BondFENEKokkos<DeviceType>::~BondFENEKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -75,13 +75,13 @@ void BondFENEKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"bond:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"bond:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"bond:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"bond:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/bond_harmonic_kokkos.cpp b/src/KOKKOS/bond_harmonic_kokkos.cpp
index da45c70d6c..c4e0c3a817 100644
--- a/src/KOKKOS/bond_harmonic_kokkos.cpp
+++ b/src/KOKKOS/bond_harmonic_kokkos.cpp
@@ -23,7 +23,7 @@
#include "domain.h"
#include "comm.h"
#include "force.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -48,8 +48,8 @@ template<class DeviceType>
BondHarmonicKokkos<DeviceType>::~BondHarmonicKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -68,15 +68,15 @@ void BondHarmonicKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
if (eflag_atom) {
//if(k_eatom.dimension_0()<maxeatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
d_eatom = k_eatom.template view<DeviceType>();
//}
}
if (vflag_atom) {
//if(k_vatom.dimension_0()<maxvatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
d_vatom = k_vatom.template view<DeviceType>();
//}
}
diff --git a/src/KOKKOS/comm_kokkos.cpp b/src/KOKKOS/comm_kokkos.cpp
index 5534341342..7bb416df93 100644
--- a/src/KOKKOS/comm_kokkos.cpp
+++ b/src/KOKKOS/comm_kokkos.cpp
@@ -20,7 +20,7 @@
#include "domain.h"
#include "atom_masks.h"
#include "error.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "force.h"
#include "pair.h"
#include "fix.h"
@@ -71,7 +71,7 @@ CommKokkos::CommKokkos(LAMMPS *lmp) : CommBrick(lmp)
for (int i = 0; i < maxswap; i++) {
maxsendlist[i] = BUFMIN;
}
- memory->create_kokkos(k_sendlist,sendlist,maxswap,BUFMIN,"comm:sendlist");
+ memoryKK->create_kokkos(k_sendlist,sendlist,maxswap,BUFMIN,"comm:sendlist");
max_buf_pair = 0;
k_buf_send_pair = DAT::tdual_xfloat_1d("comm:k_buf_send_pair",1);
@@ -82,11 +82,11 @@ CommKokkos::CommKokkos(LAMMPS *lmp) : CommBrick(lmp)
CommKokkos::~CommKokkos()
{
- memory->destroy_kokkos(k_sendlist,sendlist);
+ memoryKK->destroy_kokkos(k_sendlist,sendlist);
sendlist = NULL;
- memory->destroy_kokkos(k_buf_send,buf_send);
+ memoryKK->destroy_kokkos(k_buf_send,buf_send);
buf_send = NULL;
- memory->destroy_kokkos(k_buf_recv,buf_recv);
+ memoryKK->destroy_kokkos(k_buf_recv,buf_recv);
buf_recv = NULL;
}
@@ -1067,7 +1067,7 @@ void CommKokkos::grow_list(int iswap, int n)
k_sendlist.modify<LMPHostType>();
}
- memory->grow_kokkos(k_sendlist,sendlist,maxswap,size,"comm:sendlist");
+ memoryKK->grow_kokkos(k_sendlist,sendlist,maxswap,size,"comm:sendlist");
for(int i=0;i<maxswap;i++) {
maxsendlist[i]=size; sendlist[i]=&k_sendlist.view<LMPHostType>()(i,0);
@@ -1095,7 +1095,7 @@ void CommKokkos::grow_swap(int n)
k_sendlist.modify<LMPHostType>();
}
- memory->grow_kokkos(k_sendlist,sendlist,maxswap,size,"comm:sendlist");
+ memoryKK->grow_kokkos(k_sendlist,sendlist,maxswap,size,"comm:sendlist");
memory->grow(maxsendlist,n,"comm:maxsendlist");
for (int i=0;i<maxswap;i++) maxsendlist[i]=size;
diff --git a/src/KOKKOS/comm_tiled_kokkos.cpp b/src/KOKKOS/comm_tiled_kokkos.cpp
index adcc634aa1..33cd8eaa6e 100644
--- a/src/KOKKOS/comm_tiled_kokkos.cpp
+++ b/src/KOKKOS/comm_tiled_kokkos.cpp
@@ -25,7 +25,7 @@
#include "compute.h"
#include "output.h"
#include "dump.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
diff --git a/src/KOKKOS/dihedral_charmm_kokkos.cpp b/src/KOKKOS/dihedral_charmm_kokkos.cpp
index 7f2117c97f..71635ec76c 100644
--- a/src/KOKKOS/dihedral_charmm_kokkos.cpp
+++ b/src/KOKKOS/dihedral_charmm_kokkos.cpp
@@ -25,7 +25,7 @@
#include "force.h"
#include "pair.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -56,8 +56,8 @@ template<class DeviceType>
DihedralCharmmKokkos<DeviceType>::~DihedralCharmmKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -81,8 +81,8 @@ void DihedralCharmmKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
if (eflag_atom) {
//if(k_eatom.dimension_0()<maxeatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"dihedral:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"dihedral:eatom");
d_eatom = k_eatom.template view<DeviceType>();
k_eatom_pair = Kokkos::DualView<E_FLOAT*,Kokkos::LayoutRight,DeviceType>("dihedral:eatom_pair",maxeatom);
d_eatom_pair = k_eatom.template view<DeviceType>();
@@ -90,8 +90,8 @@ void DihedralCharmmKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
}
if (vflag_atom) {
//if(k_vatom.dimension_0()<maxvatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"dihedral:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"dihedral:vatom");
d_vatom = k_vatom.template view<DeviceType>();
k_vatom_pair = Kokkos::DualView<F_FLOAT*[6],Kokkos::LayoutRight,DeviceType>("dihedral:vatom_pair",maxvatom);
d_vatom_pair = k_vatom.template view<DeviceType>();
diff --git a/src/KOKKOS/dihedral_class2_kokkos.cpp b/src/KOKKOS/dihedral_class2_kokkos.cpp
index 89e42c6f83..d32ea4a461 100644
--- a/src/KOKKOS/dihedral_class2_kokkos.cpp
+++ b/src/KOKKOS/dihedral_class2_kokkos.cpp
@@ -24,7 +24,7 @@
#include "domain.h"
#include "force.h"
#include "update.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -56,8 +56,8 @@ template<class DeviceType>
DihedralClass2Kokkos<DeviceType>::~DihedralClass2Kokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -75,13 +75,13 @@ void DihedralClass2Kokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"dihedral:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"dihedral:eatom");
d_eatom = k_eatom.template view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"dihedral:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"dihedral:vatom");
d_vatom = k_vatom.template view<DeviceType>();
}
diff --git a/src/KOKKOS/dihedral_opls_kokkos.cpp b/src/KOKKOS/dihedral_opls_kokkos.cpp
index 0ee00ca8db..4349aff48d 100644
--- a/src/KOKKOS/dihedral_opls_kokkos.cpp
+++ b/src/KOKKOS/dihedral_opls_kokkos.cpp
@@ -24,7 +24,7 @@
#include "domain.h"
#include "force.h"
#include "update.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -56,8 +56,8 @@ template<class DeviceType>
DihedralOPLSKokkos<DeviceType>::~DihedralOPLSKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -75,13 +75,13 @@ void DihedralOPLSKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"dihedral:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"dihedral:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"dihedral:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"dihedral:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/fix_eos_table_rx_kokkos.cpp b/src/KOKKOS/fix_eos_table_rx_kokkos.cpp
index 552141ced2..5c106c19f3 100644
--- a/src/KOKKOS/fix_eos_table_rx_kokkos.cpp
+++ b/src/KOKKOS/fix_eos_table_rx_kokkos.cpp
@@ -21,7 +21,7 @@
#include "atom_kokkos.h"
#include "error.h"
#include "force.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "comm.h"
#include <math.h>
#include "modify.h"
@@ -517,14 +517,14 @@ void FixEOStableRXKokkos<DeviceType>::create_kokkos_tables()
{
const int tlm1 = tablength-1;
- memory->create_kokkos(d_table->lo,h_table->lo,ntables,"Table::lo");
- memory->create_kokkos(d_table->hi,h_table->hi,ntables,"Table::hi");
- memory->create_kokkos(d_table->invdelta,h_table->invdelta,ntables,"Table::invdelta");
+ memoryKK->create_kokkos(d_table->lo,h_table->lo,ntables,"Table::lo");
+ memoryKK->create_kokkos(d_table->hi,h_table->hi,ntables,"Table::hi");
+ memoryKK->create_kokkos(d_table->invdelta,h_table->invdelta,ntables,"Table::invdelta");
if(tabstyle == LINEAR) {
- memory->create_kokkos(d_table->r,h_table->r,ntables,tablength,"Table::r");
- memory->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
- memory->create_kokkos(d_table->de,h_table->de,ntables,tlm1,"Table::de");
+ memoryKK->create_kokkos(d_table->r,h_table->r,ntables,tablength,"Table::r");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
+ memoryKK->create_kokkos(d_table->de,h_table->de,ntables,tlm1,"Table::de");
}
for(int i=0; i < ntables; i++) {
diff --git a/src/KOKKOS/fix_langevin_kokkos.cpp b/src/KOKKOS/fix_langevin_kokkos.cpp
index fb0f329a91..108c3b692a 100644
--- a/src/KOKKOS/fix_langevin_kokkos.cpp
+++ b/src/KOKKOS/fix_langevin_kokkos.cpp
@@ -21,7 +21,7 @@
#include "update.h"
#include "respa.h"
#include "error.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "group.h"
#include "random_mars.h"
#include "compute.h"
@@ -49,9 +49,9 @@ FixLangevinKokkos<DeviceType>::FixLangevinKokkos(LAMMPS *lmp, int narg, char **a
int ntypes = atomKK->ntypes;
// allocate per-type arrays for force prefactors
- memory->create_kokkos(k_gfactor1,gfactor1,ntypes+1,"langevin:gfactor1");
- memory->create_kokkos(k_gfactor2,gfactor2,ntypes+1,"langevin:gfactor2");
- memory->create_kokkos(k_ratio,ratio,ntypes+1,"langevin:ratio");
+ memoryKK->create_kokkos(k_gfactor1,gfactor1,ntypes+1,"langevin:gfactor1");
+ memoryKK->create_kokkos(k_gfactor2,gfactor2,ntypes+1,"langevin:gfactor2");
+ memoryKK->create_kokkos(k_ratio,ratio,ntypes+1,"langevin:ratio");
d_gfactor1 = k_gfactor1.template view<DeviceType>();
h_gfactor1 = k_gfactor1.template view<LMPHostType>();
d_gfactor2 = k_gfactor2.template view<DeviceType>();
@@ -92,12 +92,12 @@ FixLangevinKokkos<DeviceType>::FixLangevinKokkos(LAMMPS *lmp, int narg, char **a
template<class DeviceType>
FixLangevinKokkos<DeviceType>::~FixLangevinKokkos()
{
- memory->destroy_kokkos(k_gfactor1,gfactor1);
- memory->destroy_kokkos(k_gfactor2,gfactor2);
- memory->destroy_kokkos(k_ratio,ratio);
- memory->destroy_kokkos(k_flangevin,flangevin);
- if(gjfflag) memory->destroy_kokkos(k_franprev,franprev);
- memory->destroy_kokkos(k_tforce,tforce);
+ memoryKK->destroy_kokkos(k_gfactor1,gfactor1);
+ memoryKK->destroy_kokkos(k_gfactor2,gfactor2);
+ memoryKK->destroy_kokkos(k_ratio,ratio);
+ memoryKK->destroy_kokkos(k_flangevin,flangevin);
+ if(gjfflag) memoryKK->destroy_kokkos(k_franprev,franprev);
+ memoryKK->destroy_kokkos(k_tforce,tforce);
}
/* ---------------------------------------------------------------------- */
@@ -121,7 +121,7 @@ void FixLangevinKokkos<DeviceType>::init()
template<class DeviceType>
void FixLangevinKokkos<DeviceType>::grow_arrays(int nmax)
{
- memory->grow_kokkos(k_franprev,franprev,nmax,3,"langevin:franprev");
+ memoryKK->grow_kokkos(k_franprev,franprev,nmax,3,"langevin:franprev");
d_franprev = k_franprev.template view<DeviceType>();
h_franprev = k_franprev.template view<LMPHostType>();
}
@@ -167,9 +167,9 @@ void FixLangevinKokkos<DeviceType>::post_force(int vflag)
// reallocate flangevin if necessary
if (tallyflag) {
if (nlocal > maxatom1) {
- memory->destroy_kokkos(k_flangevin,flangevin);
+ memoryKK->destroy_kokkos(k_flangevin,flangevin);
maxatom1 = atomKK->nmax;
- memory->create_kokkos(k_flangevin,flangevin,maxatom1,3,"langevin:flangevin");
+ memoryKK->create_kokkos(k_flangevin,flangevin,maxatom1,3,"langevin:flangevin");
d_flangevin = k_flangevin.template view<DeviceType>();
h_flangevin = k_flangevin.template view<LMPHostType>();
}
@@ -671,8 +671,8 @@ void FixLangevinKokkos<DeviceType>::compute_target()
} else {
if (atom->nmax > maxatom2) {
maxatom2 = atom->nmax;
- memory->destroy_kokkos(k_tforce,tforce);
- memory->create_kokkos(k_tforce,tforce,maxatom2,"langevin:tforce");
+ memoryKK->destroy_kokkos(k_tforce,tforce);
+ memoryKK->create_kokkos(k_tforce,tforce,maxatom2,"langevin:tforce");
d_tforce = k_tforce.template view<DeviceType>();
h_tforce = k_tforce.template view<LMPHostType>();
}
diff --git a/src/KOKKOS/fix_nh_kokkos.cpp b/src/KOKKOS/fix_nh_kokkos.cpp
index 7136c776a1..345259e355 100644
--- a/src/KOKKOS/fix_nh_kokkos.cpp
+++ b/src/KOKKOS/fix_nh_kokkos.cpp
@@ -33,7 +33,7 @@
#include "update.h"
#include "respa.h"
#include "domain_kokkos.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
#include "atom_kokkos.h"
diff --git a/src/KOKKOS/fix_property_atom_kokkos.cpp b/src/KOKKOS/fix_property_atom_kokkos.cpp
index cb52988c31..fe2f101e56 100644
--- a/src/KOKKOS/fix_property_atom_kokkos.cpp
+++ b/src/KOKKOS/fix_property_atom_kokkos.cpp
@@ -16,7 +16,7 @@
#include "fix_property_atom_kokkos.h"
#include "atom_kokkos.h"
#include "comm.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "update.h"
@@ -60,7 +60,7 @@ void FixPropertyAtomKokkos::grow_arrays(int nmax)
size_t nbytes = (nmax-nmax_old) * sizeof(int);
memset(&atom->ivector[index[m]][nmax_old],0,nbytes);
} else if (style[m] == DOUBLE) {
- memory->grow_kokkos(atomKK->k_dvector,atomKK->dvector,atomKK->k_dvector.dimension_0(),nmax,
+ memoryKK->grow_kokkos(atomKK->k_dvector,atomKK->dvector,atomKK->k_dvector.dimension_0(),nmax,
"atom:dvector");
//memory->grow(atom->dvector[index[m]],nmax,"atom:dvector");
//size_t nbytes = (nmax-nmax_old) * sizeof(double);
diff --git a/src/KOKKOS/fix_qeq_reax_kokkos.cpp b/src/KOKKOS/fix_qeq_reax_kokkos.cpp
index 5d2f6a0438..91a22361fc 100644
--- a/src/KOKKOS/fix_qeq_reax_kokkos.cpp
+++ b/src/KOKKOS/fix_qeq_reax_kokkos.cpp
@@ -35,7 +35,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "pair_reaxc_kokkos.h"
diff --git a/src/KOKKOS/fix_reaxc_bonds_kokkos.cpp b/src/KOKKOS/fix_reaxc_bonds_kokkos.cpp
index 0d74a49ed3..586daadd55 100644
--- a/src/KOKKOS/fix_reaxc_bonds_kokkos.cpp
+++ b/src/KOKKOS/fix_reaxc_bonds_kokkos.cpp
@@ -31,7 +31,7 @@
#include "compute.h"
#include "input.h"
#include "variable.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "reaxc_list.h"
#include "reaxc_types.h"
@@ -95,7 +95,7 @@ void FixReaxCBondsKokkos::Output_ReaxC_Bonds(bigint ntimestep, FILE *fp)
MPI_Allreduce(&nlocal,&nlocal_max,1,MPI_INT,MPI_MAX,world);
nbuf = 1+(numbonds_max*2+10)*nlocal_max;
- memory->create_kokkos(k_buf,buf,nbuf,"reax/c/bonds:buf");
+ memoryKK->create_kokkos(k_buf,buf,nbuf,"reax/c/bonds:buf");
// Pass information to buffer
if (reaxc->execution_space == Device)
@@ -107,7 +107,7 @@ void FixReaxCBondsKokkos::Output_ReaxC_Bonds(bigint ntimestep, FILE *fp)
// Receive information from buffer for output
RecvBuffer(buf, nbuf, nbuf_local, nlocal_tot, numbonds_max);
- memory->destroy_kokkos(k_buf,buf);
+ memoryKK->destroy_kokkos(k_buf,buf);
}
/* ---------------------------------------------------------------------- */
diff --git a/src/KOKKOS/fix_reaxc_species_kokkos.cpp b/src/KOKKOS/fix_reaxc_species_kokkos.cpp
index f2719f9f0e..a676c7ef27 100644
--- a/src/KOKKOS/fix_reaxc_species_kokkos.cpp
+++ b/src/KOKKOS/fix_reaxc_species_kokkos.cpp
@@ -33,7 +33,7 @@
#include "compute.h"
#include "input.h"
#include "variable.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "reaxc_list.h"
#include "atom_masks.h"
diff --git a/src/KOKKOS/fix_rx_kokkos.cpp b/src/KOKKOS/fix_rx_kokkos.cpp
index b1cfd20be2..d7b71a5fdd 100644
--- a/src/KOKKOS/fix_rx_kokkos.cpp
+++ b/src/KOKKOS/fix_rx_kokkos.cpp
@@ -17,7 +17,7 @@
#include "atom_masks.h"
#include "atom_kokkos.h"
#include "force.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "update.h"
#include "respa.h"
#include "modify.h"
@@ -81,15 +81,15 @@ FixRxKokkos<DeviceType>::~FixRxKokkos()
if (copymode) return;
if (localTempFlag)
- memory->destroy_kokkos(k_dpdThetaLocal, dpdThetaLocal);
+ memoryKK->destroy_kokkos(k_dpdThetaLocal, dpdThetaLocal);
- memory->destroy_kokkos(k_sumWeights, sumWeights);
- //memory->destroy_kokkos(k_sumWeights);
+ memoryKK->destroy_kokkos(k_sumWeights, sumWeights);
+ //memoryKK->destroy_kokkos(k_sumWeights);
//delete [] scratchSpace;
- memory->destroy_kokkos(d_scratchSpace);
+ memoryKK->destroy_kokkos(d_scratchSpace);
- memory->destroy_kokkos(k_cutsq);
+ memoryKK->destroy_kokkos(k_cutsq);
}
/* ---------------------------------------------------------------------- */
@@ -1233,9 +1233,9 @@ void FixRxKokkos<DeviceType>::create_kinetics_data(void)
{
//printf("Inside FixRxKokkos::create_kinetics_data\n");
- memory->create_kokkos( d_kineticsData.Arr, h_kineticsData.Arr, nreactions, "KineticsType::Arr");
- memory->create_kokkos( d_kineticsData.nArr, h_kineticsData.nArr, nreactions, "KineticsType::nArr");
- memory->create_kokkos( d_kineticsData.Ea, h_kineticsData.Ea, nreactions, "KineticsType::Ea");
+ memoryKK->create_kokkos( d_kineticsData.Arr, h_kineticsData.Arr, nreactions, "KineticsType::Arr");
+ memoryKK->create_kokkos( d_kineticsData.nArr, h_kineticsData.nArr, nreactions, "KineticsType::nArr");
+ memoryKK->create_kokkos( d_kineticsData.Ea, h_kineticsData.Ea, nreactions, "KineticsType::Ea");
for (int i = 0; i < nreactions; ++i)
{
@@ -1251,8 +1251,8 @@ void FixRxKokkos<DeviceType>::create_kinetics_data(void)
if (useSparseKinetics)
{
- memory->create_kokkos( d_kineticsData.nu , h_kineticsData.nu , nreactions, sparseKinetics_maxSpecies, "KineticsType::nu");
- memory->create_kokkos( d_kineticsData.nuk, h_kineticsData.nuk, nreactions, sparseKinetics_maxSpecies, "KineticsType::nuk");
+ memoryKK->create_kokkos( d_kineticsData.nu , h_kineticsData.nu , nreactions, sparseKinetics_maxSpecies, "KineticsType::nu");
+ memoryKK->create_kokkos( d_kineticsData.nuk, h_kineticsData.nuk, nreactions, sparseKinetics_maxSpecies, "KineticsType::nuk");
for (int i = 0; i < nreactions; ++i)
for (int k = 0; k < sparseKinetics_maxSpecies; ++k)
@@ -1266,8 +1266,8 @@ void FixRxKokkos<DeviceType>::create_kinetics_data(void)
if (SparseKinetics_enableIntegralReactions)
{
- memory->create_kokkos( d_kineticsData.inu, h_kineticsData.inu, nreactions, sparseKinetics_maxSpecies, "KineticsType::inu");
- memory->create_kokkos( d_kineticsData.isIntegral, h_kineticsData.isIntegral, nreactions, "KineticsType::isIntegral");
+ memoryKK->create_kokkos( d_kineticsData.inu, h_kineticsData.inu, nreactions, sparseKinetics_maxSpecies, "KineticsType::inu");
+ memoryKK->create_kokkos( d_kineticsData.isIntegral, h_kineticsData.isIntegral, nreactions, "KineticsType::isIntegral");
for (int i = 0; i < nreactions; ++i)
{
@@ -1286,9 +1286,9 @@ void FixRxKokkos<DeviceType>::create_kinetics_data(void)
//{
// Dense option
- memory->create_kokkos( d_kineticsData.stoich, h_kineticsData.stoich, nreactions, nspecies, "KineticsType::stoich");
- memory->create_kokkos( d_kineticsData.stoichReactants, h_kineticsData.stoichReactants, nreactions, nspecies, "KineticsType::stoichReactants");
- memory->create_kokkos( d_kineticsData.stoichProducts, h_kineticsData.stoichProducts, nreactions, nspecies, "KineticsType::stoichProducts");
+ memoryKK->create_kokkos( d_kineticsData.stoich, h_kineticsData.stoich, nreactions, nspecies, "KineticsType::stoich");
+ memoryKK->create_kokkos( d_kineticsData.stoichReactants, h_kineticsData.stoichReactants, nreactions, nspecies, "KineticsType::stoichReactants");
+ memoryKK->create_kokkos( d_kineticsData.stoichProducts, h_kineticsData.stoichProducts, nreactions, nspecies, "KineticsType::stoichProducts");
for (int i = 0; i < nreactions; ++i)
for (int k = 0; k < nspecies; ++k)
@@ -1445,8 +1445,8 @@ void FixRxKokkos<DeviceType>::solve_reactions(const int vflag, const bool isPreF
const int count = nlocal + (newton_pair ? nghost : 0);
if (count > k_dpdThetaLocal.template view<DeviceType>().dimension_0()) {
- memory->destroy_kokkos (k_dpdThetaLocal, dpdThetaLocal);
- memory->create_kokkos (k_dpdThetaLocal, dpdThetaLocal, count, "FixRxKokkos::dpdThetaLocal");
+ memoryKK->destroy_kokkos (k_dpdThetaLocal, dpdThetaLocal);
+ memoryKK->create_kokkos (k_dpdThetaLocal, dpdThetaLocal, count, "FixRxKokkos::dpdThetaLocal");
this->d_dpdThetaLocal = k_dpdThetaLocal.template view<DeviceType>();
this->h_dpdThetaLocal = k_dpdThetaLocal.h_view;
}
@@ -1511,8 +1511,8 @@ void FixRxKokkos<DeviceType>::solve_reactions(const int vflag, const bool isPreF
if (odeIntegrationFlag == ODE_LAMMPS_RKF45 && diagnosticFrequency == 1)
{
- memory->create_kokkos (k_diagnosticCounterPerODEnSteps, diagnosticCounterPerODEnSteps, nlocal, "FixRxKokkos::diagnosticCounterPerODEnSteps");
- memory->create_kokkos (k_diagnosticCounterPerODEnFuncs, diagnosticCounterPerODEnFuncs, nlocal, "FixRxKokkos::diagnosticCounterPerODEnFuncs");
+ memoryKK->create_kokkos (k_diagnosticCounterPerODEnSteps, diagnosticCounterPerODEnSteps, nlocal, "FixRxKokkos::diagnosticCounterPerODEnSteps");
+ memoryKK->create_kokkos (k_diagnosticCounterPerODEnFuncs, diagnosticCounterPerODEnFuncs, nlocal, "FixRxKokkos::diagnosticCounterPerODEnFuncs");
d_diagnosticCounterPerODEnSteps = k_diagnosticCounterPerODEnSteps.template view<DeviceType>();
d_diagnosticCounterPerODEnFuncs = k_diagnosticCounterPerODEnFuncs.template view<DeviceType>();
@@ -1542,8 +1542,8 @@ void FixRxKokkos<DeviceType>::solve_reactions(const int vflag, const bool isPreF
//typename ArrayTypes<DeviceType>::t_double_1d d_scratchSpace("d_scratchSpace", scratchSpaceSize * nlocal);
if (nlocal*scratchSpaceSize > d_scratchSpace.dimension_0()) {
- memory->destroy_kokkos (d_scratchSpace);
- memory->create_kokkos (d_scratchSpace, nlocal*scratchSpaceSize, "FixRxKokkos::d_scratchSpace");
+ memoryKK->destroy_kokkos (d_scratchSpace);
+ memoryKK->create_kokkos (d_scratchSpace, nlocal*scratchSpaceSize, "FixRxKokkos::d_scratchSpace");
}
#if 0
@@ -1811,8 +1811,8 @@ void FixRxKokkos<DeviceType>::odeDiagnostics(void)
my_min[FuncSum] = std::min( my_min[FuncSum], (double)nFuncs );
}
- memory->destroy_kokkos( k_diagnosticCounterPerODEnSteps, diagnosticCounterPerODEnSteps );
- memory->destroy_kokkos( k_diagnosticCounterPerODEnFuncs, diagnosticCounterPerODEnFuncs );
+ memoryKK->destroy_kokkos( k_diagnosticCounterPerODEnSteps, diagnosticCounterPerODEnSteps );
+ memoryKK->destroy_kokkos( k_diagnosticCounterPerODEnFuncs, diagnosticCounterPerODEnFuncs );
MPI_Reduce (my_sum_sq, sum_sq, 2*numCounters, MPI_DOUBLE, MPI_SUM, 0, world);
@@ -2022,10 +2022,10 @@ void FixRxKokkos<DeviceType>::computeLocalTemperature()
{
const int ntypes = atom->ntypes;
- //memory->create_kokkos (k_cutsq, h_cutsq, ntypes+1, ntypes+1, "pair:cutsq");
+ //memoryKK->create_kokkos (k_cutsq, h_cutsq, ntypes+1, ntypes+1, "pair:cutsq");
if (ntypes+1 > k_cutsq.dimension_0()) {
- memory->destroy_kokkos (k_cutsq);
- memory->create_kokkos (k_cutsq, ntypes+1, ntypes+1, "FixRxKokkos::k_cutsq");
+ memoryKK->destroy_kokkos (k_cutsq);
+ memoryKK->create_kokkos (k_cutsq, ntypes+1, ntypes+1, "FixRxKokkos::k_cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
}
@@ -2043,10 +2043,10 @@ void FixRxKokkos<DeviceType>::computeLocalTemperature()
// Initialize the local temperature weight array
int sumWeightsCt = nlocal + (NEWTON_PAIR ? nghost : 0);
- //memory->create_kokkos (k_sumWeights, sumWeights, sumWeightsCt, "FixRxKokkos::sumWeights");
+ //memoryKK->create_kokkos (k_sumWeights, sumWeights, sumWeightsCt, "FixRxKokkos::sumWeights");
if (sumWeightsCt > k_sumWeights.template view<DeviceType>().dimension_0()) {
- memory->destroy_kokkos(k_sumWeights, sumWeights);
- memory->create_kokkos (k_sumWeights, sumWeightsCt, "FixRxKokkos::sumWeights");
+ memoryKK->destroy_kokkos(k_sumWeights, sumWeights);
+ memoryKK->create_kokkos (k_sumWeights, sumWeightsCt, "FixRxKokkos::sumWeights");
d_sumWeights = k_sumWeights.template view<DeviceType>();
h_sumWeights = k_sumWeights.h_view;
}
diff --git a/src/KOKKOS/fix_setforce_kokkos.cpp b/src/KOKKOS/fix_setforce_kokkos.cpp
index 2ef04ad6ee..7c2c115f52 100644
--- a/src/KOKKOS/fix_setforce_kokkos.cpp
+++ b/src/KOKKOS/fix_setforce_kokkos.cpp
@@ -22,7 +22,7 @@
#include "respa.h"
#include "input.h"
#include "variable.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "force.h"
#include "atom_masks.h"
@@ -45,7 +45,7 @@ FixSetForceKokkos<DeviceType>::FixSetForceKokkos(LAMMPS *lmp, int narg, char **a
datamask_modify = EMPTY_MASK;
memory->destroy(sforce);
- memory->create_kokkos(k_sforce,sforce,maxatom,3,"setforce:sforce");
+ memoryKK->create_kokkos(k_sforce,sforce,maxatom,3,"setforce:sforce");
}
/* ---------------------------------------------------------------------- */
@@ -55,7 +55,7 @@ FixSetForceKokkos<DeviceType>::~FixSetForceKokkos()
{
if (copymode) return;
- memory->destroy_kokkos(k_sforce,sforce);
+ memoryKK->destroy_kokkos(k_sforce,sforce);
sforce = NULL;
}
@@ -99,8 +99,8 @@ void FixSetForceKokkos<DeviceType>::post_force(int vflag)
if (varflag == ATOM && atom->nmax > maxatom) {
maxatom = atom->nmax;
- memory->destroy_kokkos(k_sforce,sforce);
- memory->create_kokkos(k_sforce,sforce,maxatom,3,"setforce:sforce");
+ memoryKK->destroy_kokkos(k_sforce,sforce);
+ memoryKK->create_kokkos(k_sforce,sforce,maxatom,3,"setforce:sforce");
}
foriginal[0] = foriginal[1] = foriginal[2] = 0.0;
diff --git a/src/KOKKOS/fix_shardlow_kokkos.cpp b/src/KOKKOS/fix_shardlow_kokkos.cpp
index 98bbb02714..cc1bd6bede 100644
--- a/src/KOKKOS/fix_shardlow_kokkos.cpp
+++ b/src/KOKKOS/fix_shardlow_kokkos.cpp
@@ -50,7 +50,7 @@
#include "neighbor.h"
#include "neigh_list_kokkos.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "domain.h"
#include "modify.h"
// #include "pair_dpd_fdt.h"
diff --git a/src/KOKKOS/gridcomm_kokkos.cpp b/src/KOKKOS/gridcomm_kokkos.cpp
index 6871ef67ae..d262c51ff6 100644
--- a/src/KOKKOS/gridcomm_kokkos.cpp
+++ b/src/KOKKOS/gridcomm_kokkos.cpp
@@ -15,7 +15,7 @@
#include "gridcomm_kokkos.h"
#include "comm.h"
#include "kspace.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
using namespace LAMMPS_NS;
@@ -126,8 +126,8 @@ template<class DeviceType>
GridCommKokkos<DeviceType>::~GridCommKokkos()
{
for (int i = 0; i < nswap; i++) {
- //memory->destroy_kokkos(swap[i].k_packlist,swap[i].packlist);
- //memory->destroy_kokkos(swap[i].k_unpacklist,swap[i].unpacklist);
+ //memoryKK->destroy_kokkos(swap[i].k_packlist,swap[i].packlist);
+ //memoryKK->destroy_kokkos(swap[i].k_unpacklist,swap[i].unpacklist);
}
memory->sfree(swap);
diff --git a/src/KOKKOS/improper_class2_kokkos.cpp b/src/KOKKOS/improper_class2_kokkos.cpp
index c2cb7dfe2b..d2d465a250 100644
--- a/src/KOKKOS/improper_class2_kokkos.cpp
+++ b/src/KOKKOS/improper_class2_kokkos.cpp
@@ -26,7 +26,7 @@
#include "force.h"
#include "update.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -58,8 +58,8 @@ template<class DeviceType>
ImproperClass2Kokkos<DeviceType>::~ImproperClass2Kokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -78,15 +78,15 @@ void ImproperClass2Kokkos<DeviceType>::compute(int eflag_in, int vflag_in)
if (eflag_atom) {
//if(k_eatom.dimension_0()<maxeatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
d_eatom = k_eatom.template view<DeviceType>();
//}
}
if (vflag_atom) {
//if(k_vatom.dimension_0()<maxvatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
d_vatom = k_vatom.template view<DeviceType>();
//}
}
diff --git a/src/KOKKOS/improper_harmonic_kokkos.cpp b/src/KOKKOS/improper_harmonic_kokkos.cpp
index 73e105864f..49dd36ed19 100644
--- a/src/KOKKOS/improper_harmonic_kokkos.cpp
+++ b/src/KOKKOS/improper_harmonic_kokkos.cpp
@@ -26,7 +26,7 @@
#include "force.h"
#include "update.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -58,8 +58,8 @@ template<class DeviceType>
ImproperHarmonicKokkos<DeviceType>::~ImproperHarmonicKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -78,15 +78,15 @@ void ImproperHarmonicKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
if (eflag_atom) {
//if(k_eatom.dimension_0()<maxeatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"improper:eatom");
d_eatom = k_eatom.template view<DeviceType>();
//}
}
if (vflag_atom) {
//if(k_vatom.dimension_0()<maxvatom) { // won't work without adding zero functor
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"improper:vatom");
d_vatom = k_vatom.template view<DeviceType>();
//}
}
diff --git a/src/KOKKOS/kokkos.cpp b/src/KOKKOS/kokkos.cpp
index 2b02624dce..62f0df4181 100644
--- a/src/KOKKOS/kokkos.cpp
+++ b/src/KOKKOS/kokkos.cpp
@@ -23,6 +23,7 @@
#include "neighbor_kokkos.h"
#include "neigh_list_kokkos.h"
#include "error.h"
+#include "memory_kokkos.h"
using namespace LAMMPS_NS;
@@ -33,6 +34,8 @@ KokkosLMP::KokkosLMP(LAMMPS *lmp, int narg, char **arg) : Pointers(lmp)
kokkos_exists = 1;
lmp->kokkos = this;
+ memoryKK = (MemoryKokkos*) memory;
+
auto_sync = 1;
int me = 0;
diff --git a/src/KOKKOS/memory_kokkos.h b/src/KOKKOS/memory_kokkos.h
index 8ade198c40..2454b7c412 100644
--- a/src/KOKKOS/memory_kokkos.h
+++ b/src/KOKKOS/memory_kokkos.h
@@ -11,6 +11,19 @@
See the README file in the top-level LAMMPS directory.
------------------------------------------------------------------------- */
+#ifndef LMP_MEMORY_KOKKOS_H
+#define LMP_MEMORY_KOKKOS_H
+
+#ifdef LMP_KOKKOS
+#include "kokkos_type.h"
+#endif
+
+namespace LAMMPS_NS {
+
+class MemoryKokkos : public Memory {
+ public:
+ MemoryKokkos::MemoryKokkos(class LAMMPS *lmp) : Memory(lmp) {}
+
/* ----------------------------------------------------------------------
Kokkos versions of create/grow/destroy multi-dimensional arrays
------------------------------------------------------------------------- */
@@ -279,3 +292,9 @@ void destroy_kokkos(TYPE data, typename TYPE::value_type** &array)
sfree(array);
array = NULL;
}
+
+};
+
+}
+
+#endif
\ No newline at end of file
diff --git a/src/KOKKOS/nbin_ssa_kokkos.cpp b/src/KOKKOS/nbin_ssa_kokkos.cpp
index ab97cb5848..1a88165856 100644
--- a/src/KOKKOS/nbin_ssa_kokkos.cpp
+++ b/src/KOKKOS/nbin_ssa_kokkos.cpp
@@ -26,7 +26,7 @@
#include "error.h"
#include "atom_masks.h"
-// #include "memory.h"
+// #include "memory_kokkos.h"
using namespace LAMMPS_NS;
diff --git a/src/KOKKOS/neigh_bond_kokkos.cpp b/src/KOKKOS/neigh_bond_kokkos.cpp
index a674e7cec4..3ecc8b5e51 100644
--- a/src/KOKKOS/neigh_bond_kokkos.cpp
+++ b/src/KOKKOS/neigh_bond_kokkos.cpp
@@ -24,7 +24,7 @@
#include "domain_kokkos.h"
#include "output.h"
#include "thermo.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "modify.h"
#include "fix.h"
@@ -80,27 +80,27 @@ void NeighBondKokkos<DeviceType>::init_topology_kk() {
if (atom->molecular && atom->nbonds && maxbond == 0) {
if (nprocs == 1) maxbond = atom->nbonds;
else maxbond = static_cast<int> (LB_FACTOR * atom->nbonds / nprocs);
- memory->create_kokkos(k_bondlist,neighbor->bondlist,maxbond,3,"neigh:neighbor->bondlist");
+ memoryKK->create_kokkos(k_bondlist,neighbor->bondlist,maxbond,3,"neigh:neighbor->bondlist");
}
if (atom->molecular && atom->nangles && maxangle == 0) {
if (nprocs == 1) maxangle = atom->nangles;
else maxangle = static_cast<int> (LB_FACTOR * atom->nangles / nprocs);
- memory->create_kokkos(k_anglelist,neighbor->anglelist,maxangle,4,"neigh:neighbor->anglelist");
+ memoryKK->create_kokkos(k_anglelist,neighbor->anglelist,maxangle,4,"neigh:neighbor->anglelist");
}
if (atom->molecular && atom->ndihedrals && maxdihedral == 0) {
if (nprocs == 1) maxdihedral = atom->ndihedrals;
else maxdihedral = static_cast<int>
(LB_FACTOR * atom->ndihedrals / nprocs);
- memory->create_kokkos(k_dihedrallist,neighbor->dihedrallist,maxdihedral,5,"neigh:neighbor->dihedrallist");
+ memoryKK->create_kokkos(k_dihedrallist,neighbor->dihedrallist,maxdihedral,5,"neigh:neighbor->dihedrallist");
}
if (atom->molecular && atom->nimpropers && maximproper == 0) {
if (nprocs == 1) maximproper = atom->nimpropers;
else maximproper = static_cast<int>
(LB_FACTOR * atom->nimpropers / nprocs);
- memory->create_kokkos(k_improperlist,neighbor->improperlist,maximproper,5,"neigh:neighbor->improperlist");
+ memoryKK->create_kokkos(k_improperlist,neighbor->improperlist,maximproper,5,"neigh:neighbor->improperlist");
}
// set flags that determine which topology neighboring routines to use
@@ -283,7 +283,7 @@ void NeighBondKokkos<DeviceType>::bond_all()
k_fail_flag.template sync<LMPHostType>();
if (h_fail_flag()) {
maxbond = neighbor->nbondlist + BONDDELTA;
- memory->grow_kokkos(k_bondlist,neighbor->bondlist,maxbond,3,"neighbor:neighbor->bondlist");
+ memoryKK->grow_kokkos(k_bondlist,neighbor->bondlist,maxbond,3,"neighbor:neighbor->bondlist");
v_bondlist = k_bondlist.view<DeviceType>();
}
} while (h_fail_flag());
@@ -378,7 +378,7 @@ void NeighBondKokkos<DeviceType>::bond_partial()
k_fail_flag.template sync<LMPHostType>();
if (h_fail_flag()) {
maxbond = neighbor->nbondlist + BONDDELTA;
- memory->grow_kokkos(k_bondlist,neighbor->bondlist,maxbond,3,"neighbor:neighbor->bondlist");
+ memoryKK->grow_kokkos(k_bondlist,neighbor->bondlist,maxbond,3,"neighbor:neighbor->bondlist");
v_bondlist = k_bondlist.view<DeviceType>();
}
} while (h_fail_flag());
@@ -500,7 +500,7 @@ void NeighBondKokkos<DeviceType>::angle_all()
k_fail_flag.template sync<LMPHostType>();
if (h_fail_flag()) {
maxangle = neighbor->nanglelist + BONDDELTA;
- memory->grow_kokkos(k_anglelist,neighbor->anglelist,maxangle,4,"neighbor:neighbor->anglelist");
+ memoryKK->grow_kokkos(k_anglelist,neighbor->anglelist,maxangle,4,"neighbor:neighbor->anglelist");
v_anglelist = k_anglelist.view<DeviceType>();
}
} while (h_fail_flag());
@@ -602,7 +602,7 @@ void NeighBondKokkos<DeviceType>::angle_partial()
k_fail_flag.template sync<LMPHostType>();
if (h_fail_flag()) {
maxangle = neighbor->nanglelist + BONDDELTA;
- memory->grow_kokkos(k_anglelist,neighbor->anglelist,maxangle,4,"neighbor:neighbor->anglelist");
+ memoryKK->grow_kokkos(k_anglelist,neighbor->anglelist,maxangle,4,"neighbor:neighbor->anglelist");
v_anglelist = k_anglelist.view<DeviceType>();
}
} while (h_fail_flag());
@@ -744,7 +744,7 @@ void NeighBondKokkos<DeviceType>::dihedral_all()
k_fail_flag.template sync<LMPHostType>();
if (h_fail_flag()) {
maxdihedral = neighbor->ndihedrallist + BONDDELTA;
- memory->grow_kokkos(k_dihedrallist,neighbor->dihedrallist,maxdihedral,5,"neighbor:neighbor->dihedrallist");
+ memoryKK->grow_kokkos(k_dihedrallist,neighbor->dihedrallist,maxdihedral,5,"neighbor:neighbor->dihedrallist");
v_dihedrallist = k_dihedrallist.view<DeviceType>();
}
} while (h_fail_flag());
@@ -851,7 +851,7 @@ void NeighBondKokkos<DeviceType>::dihedral_partial()
k_fail_flag.template sync<LMPHostType>();
if (h_fail_flag()) {
maxdihedral = neighbor->ndihedrallist + BONDDELTA;
- memory->grow_kokkos(k_dihedrallist,neighbor->dihedrallist,maxdihedral,5,"neighbor:neighbor->dihedrallist");
+ memoryKK->grow_kokkos(k_dihedrallist,neighbor->dihedrallist,maxdihedral,5,"neighbor:neighbor->dihedrallist");
v_dihedrallist = k_dihedrallist.view<DeviceType>();
}
} while (h_fail_flag());
@@ -1015,7 +1015,7 @@ void NeighBondKokkos<DeviceType>::improper_all()
k_fail_flag.template sync<LMPHostType>();
if (h_fail_flag()) {
maximproper = neighbor->nimproperlist + BONDDELTA;
- memory->grow_kokkos(k_improperlist,neighbor->improperlist,maximproper,5,"neighbor:neighbor->improperlist");
+ memoryKK->grow_kokkos(k_improperlist,neighbor->improperlist,maximproper,5,"neighbor:neighbor->improperlist");
v_improperlist = k_improperlist.view<DeviceType>();
}
} while (h_fail_flag());
@@ -1122,7 +1122,7 @@ void NeighBondKokkos<DeviceType>::improper_partial()
k_fail_flag.template sync<LMPHostType>();
if (h_fail_flag()) {
maximproper = neighbor->nimproperlist + BONDDELTA;
- memory->grow_kokkos(k_improperlist,neighbor->improperlist,maximproper,5,"neighbor:neighbor->improperlist");
+ memoryKK->grow_kokkos(k_improperlist,neighbor->improperlist,maximproper,5,"neighbor:neighbor->improperlist");
v_improperlist = k_improperlist.view<DeviceType>();
}
} while (h_fail_flag());
diff --git a/src/KOKKOS/neigh_list_kokkos.cpp b/src/KOKKOS/neigh_list_kokkos.cpp
index 04454e53cb..98294a802a 100644
--- a/src/KOKKOS/neigh_list_kokkos.cpp
+++ b/src/KOKKOS/neigh_list_kokkos.cpp
@@ -13,7 +13,7 @@
#include "neigh_list_kokkos.h"
#include "atom.h"
-#include "memory.h"
+#include "memory_kokkos.h"
using namespace LAMMPS_NS;
diff --git a/src/KOKKOS/neighbor_kokkos.cpp b/src/KOKKOS/neighbor_kokkos.cpp
index f34b149864..cf1b318201 100644
--- a/src/KOKKOS/neighbor_kokkos.cpp
+++ b/src/KOKKOS/neighbor_kokkos.cpp
@@ -16,7 +16,7 @@
#include "pair.h"
#include "fix.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "update.h"
#include "atom_masks.h"
#include "error.h"
@@ -52,24 +52,24 @@ NeighborKokkos::NeighborKokkos(LAMMPS *lmp) : Neighbor(lmp),
NeighborKokkos::~NeighborKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_cutneighsq,cutneighsq);
+ memoryKK->destroy_kokkos(k_cutneighsq,cutneighsq);
cutneighsq = NULL;
- memory->destroy_kokkos(k_ex_type,ex_type);
- memory->destroy_kokkos(k_ex1_type,ex1_type);
- memory->destroy_kokkos(k_ex2_type,ex2_type);
- memory->destroy_kokkos(k_ex1_group,ex1_group);
- memory->destroy_kokkos(k_ex2_group,ex2_group);
- memory->destroy_kokkos(k_ex_mol_group,ex_mol_group);
- memory->destroy_kokkos(k_ex1_bit,ex1_bit);
- memory->destroy_kokkos(k_ex2_bit,ex2_bit);
- memory->destroy_kokkos(k_ex_mol_bit,ex_mol_bit);
- memory->destroy_kokkos(k_ex_mol_intra,ex_mol_intra);
-
- memory->destroy_kokkos(k_bondlist,bondlist);
- memory->destroy_kokkos(k_anglelist,anglelist);
- memory->destroy_kokkos(k_dihedrallist,dihedrallist);
- memory->destroy_kokkos(k_improperlist,improperlist);
+ memoryKK->destroy_kokkos(k_ex_type,ex_type);
+ memoryKK->destroy_kokkos(k_ex1_type,ex1_type);
+ memoryKK->destroy_kokkos(k_ex2_type,ex2_type);
+ memoryKK->destroy_kokkos(k_ex1_group,ex1_group);
+ memoryKK->destroy_kokkos(k_ex2_group,ex2_group);
+ memoryKK->destroy_kokkos(k_ex_mol_group,ex_mol_group);
+ memoryKK->destroy_kokkos(k_ex1_bit,ex1_bit);
+ memoryKK->destroy_kokkos(k_ex2_bit,ex2_bit);
+ memoryKK->destroy_kokkos(k_ex_mol_bit,ex_mol_bit);
+ memoryKK->destroy_kokkos(k_ex_mol_intra,ex_mol_intra);
+
+ memoryKK->destroy_kokkos(k_bondlist,bondlist);
+ memoryKK->destroy_kokkos(k_anglelist,anglelist);
+ memoryKK->destroy_kokkos(k_dihedrallist,dihedrallist);
+ memoryKK->destroy_kokkos(k_improperlist,improperlist);
}
}
@@ -90,7 +90,7 @@ void NeighborKokkos::init()
void NeighborKokkos::init_cutneighsq_kokkos(int n)
{
- memory->create_kokkos(k_cutneighsq,cutneighsq,n+1,n+1,"neigh:cutneighsq");
+ memoryKK->create_kokkos(k_cutneighsq,cutneighsq,n+1,n+1,"neigh:cutneighsq");
k_cutneighsq.modify<LMPHostType>();
}
@@ -112,7 +112,7 @@ void NeighborKokkos::create_kokkos_list(int i)
void NeighborKokkos::init_ex_type_kokkos(int n)
{
- memory->create_kokkos(k_ex_type,ex_type,n+1,n+1,"neigh:ex_type");
+ memoryKK->create_kokkos(k_ex_type,ex_type,n+1,n+1,"neigh:ex_type");
k_ex_type.modify<LMPHostType>();
}
@@ -120,9 +120,9 @@ void NeighborKokkos::init_ex_type_kokkos(int n)
void NeighborKokkos::init_ex_bit_kokkos()
{
- memory->create_kokkos(k_ex1_bit, ex1_bit, nex_group, "neigh:ex1_bit");
+ memoryKK->create_kokkos(k_ex1_bit, ex1_bit, nex_group, "neigh:ex1_bit");
k_ex1_bit.modify<LMPHostType>();
- memory->create_kokkos(k_ex2_bit, ex2_bit, nex_group, "neigh:ex2_bit");
+ memoryKK->create_kokkos(k_ex2_bit, ex2_bit, nex_group, "neigh:ex2_bit");
k_ex2_bit.modify<LMPHostType>();
}
@@ -130,7 +130,7 @@ void NeighborKokkos::init_ex_bit_kokkos()
void NeighborKokkos::init_ex_mol_bit_kokkos()
{
- memory->create_kokkos(k_ex_mol_bit, ex_mol_bit, nex_mol, "neigh:ex_mol_bit");
+ memoryKK->create_kokkos(k_ex_mol_bit, ex_mol_bit, nex_mol, "neigh:ex_mol_bit");
k_ex_mol_bit.modify<LMPHostType>();
}
@@ -138,7 +138,7 @@ void NeighborKokkos::init_ex_mol_bit_kokkos()
void NeighborKokkos::grow_ex_mol_intra_kokkos()
{
- memory->grow_kokkos(k_ex_mol_intra, ex_mol_intra, maxex_mol, "neigh:ex_mol_intra");
+ memoryKK->grow_kokkos(k_ex_mol_intra, ex_mol_intra, maxex_mol, "neigh:ex_mol_intra");
k_ex_mol_intra.modify<LMPHostType>();
}
@@ -335,29 +335,29 @@ void NeighborKokkos::operator()(TagNeighborXhold<DeviceType>, const int &i) cons
/* ---------------------------------------------------------------------- */
void NeighborKokkos::modify_ex_type_grow_kokkos(){
- memory->grow_kokkos(k_ex1_type,ex1_type,maxex_type,"neigh:ex1_type");
+ memoryKK->grow_kokkos(k_ex1_type,ex1_type,maxex_type,"neigh:ex1_type");
k_ex1_type.modify<LMPHostType>();
- memory->grow_kokkos(k_ex2_type,ex2_type,maxex_type,"neigh:ex2_type");
+ memoryKK->grow_kokkos(k_ex2_type,ex2_type,maxex_type,"neigh:ex2_type");
k_ex2_type.modify<LMPHostType>();
}
/* ---------------------------------------------------------------------- */
void NeighborKokkos::modify_ex_group_grow_kokkos(){
- memory->grow_kokkos(k_ex1_group,ex1_group,maxex_group,"neigh:ex1_group");
+ memoryKK->grow_kokkos(k_ex1_group,ex1_group,maxex_group,"neigh:ex1_group");
k_ex1_group.modify<LMPHostType>();
- memory->grow_kokkos(k_ex2_group,ex2_group,maxex_group,"neigh:ex2_group");
+ memoryKK->grow_kokkos(k_ex2_group,ex2_group,maxex_group,"neigh:ex2_group");
k_ex2_group.modify<LMPHostType>();
}
/* ---------------------------------------------------------------------- */
void NeighborKokkos::modify_mol_group_grow_kokkos(){
- memory->grow_kokkos(k_ex_mol_group,ex_mol_group,maxex_mol,"neigh:ex_mol_group");
+ memoryKK->grow_kokkos(k_ex_mol_group,ex_mol_group,maxex_mol,"neigh:ex_mol_group");
k_ex_mol_group.modify<LMPHostType>();
}
/* ---------------------------------------------------------------------- */
void NeighborKokkos::modify_mol_intra_grow_kokkos(){
- memory->grow_kokkos(k_ex_mol_intra,ex_mol_intra,maxex_mol,"neigh:ex_mol_intra");
+ memoryKK->grow_kokkos(k_ex_mol_intra,ex_mol_intra,maxex_mol,"neigh:ex_mol_intra");
k_ex_mol_intra.modify<LMPHostType>();
}
diff --git a/src/KOKKOS/pair_buck_coul_cut_kokkos.cpp b/src/KOKKOS/pair_buck_coul_cut_kokkos.cpp
index 0da8a0a3d6..ba3eda64dd 100644
--- a/src/KOKKOS/pair_buck_coul_cut_kokkos.cpp
+++ b/src/KOKKOS/pair_buck_coul_cut_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -65,8 +65,8 @@ PairBuckCoulCutKokkos<DeviceType>::~PairBuckCoulCutKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
k_cut_ljsq = DAT::tdual_ffloat_2d();
k_cut_coulsq = DAT::tdual_ffloat_2d();
@@ -98,13 +98,13 @@ void PairBuckCoulCutKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -252,13 +252,13 @@ void PairBuckCoulCutKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
memory->destroy(cut_coulsq);
- memory->create_kokkos(k_cut_coulsq,cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_buck_coul**,Kokkos::LayoutRight,DeviceType>("PairBuckCoulCut::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_buck_coul_long_kokkos.cpp b/src/KOKKOS/pair_buck_coul_long_kokkos.cpp
index 3a5cbd868f..19af349a63 100644
--- a/src/KOKKOS/pair_buck_coul_long_kokkos.cpp
+++ b/src/KOKKOS/pair_buck_coul_long_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -73,8 +73,8 @@ template<class DeviceType>
PairBuckCoulLongKokkos<DeviceType>::~PairBuckCoulLongKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
k_cut_ljsq = DAT::tdual_ffloat_2d();
k_cut_coulsq = DAT::tdual_ffloat_2d();
@@ -117,13 +117,13 @@ void PairBuckCoulLongKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -308,14 +308,14 @@ void PairBuckCoulLongKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_buck_coul**,Kokkos::LayoutRight,DeviceType>("PairBuckCoulLong::params",n+1,n+1);
diff --git a/src/KOKKOS/pair_buck_kokkos.cpp b/src/KOKKOS/pair_buck_kokkos.cpp
index e7640471d5..fcf14533dc 100644
--- a/src/KOKKOS/pair_buck_kokkos.cpp
+++ b/src/KOKKOS/pair_buck_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -61,8 +61,8 @@ template<class DeviceType>
PairBuckKokkos<DeviceType>::~PairBuckKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
memory->sfree(cutsq);
eatom = NULL;
@@ -87,13 +87,13 @@ void PairBuckKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -194,7 +194,7 @@ void PairBuckKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_buck**,Kokkos::LayoutRight,DeviceType>("PairBuck::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_coul_cut_kokkos.cpp b/src/KOKKOS/pair_coul_cut_kokkos.cpp
index 8edf093e2e..e20e243c09 100644
--- a/src/KOKKOS/pair_coul_cut_kokkos.cpp
+++ b/src/KOKKOS/pair_coul_cut_kokkos.cpp
@@ -27,7 +27,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -55,7 +55,7 @@ template<class DeviceType>
PairCoulCutKokkos<DeviceType>::~PairCoulCutKokkos()
{
if (allocated)
- memory->destroy_kokkos(k_cutsq, cutsq);
+ memoryKK->destroy_kokkos(k_cutsq, cutsq);
}
/* ---------------------------------------------------------------------- */
@@ -86,13 +86,13 @@ void PairCoulCutKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -190,7 +190,7 @@ void PairCoulCutKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_cut_ljsq = typename ArrayTypes<DeviceType>::tdual_ffloat_2d("pair:cut_ljsq",n+1,n+1);
diff --git a/src/KOKKOS/pair_coul_debye_kokkos.cpp b/src/KOKKOS/pair_coul_debye_kokkos.cpp
index c331ab8da8..4cac18cacf 100644
--- a/src/KOKKOS/pair_coul_debye_kokkos.cpp
+++ b/src/KOKKOS/pair_coul_debye_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -62,7 +62,7 @@ template<class DeviceType>
PairCoulDebyeKokkos<DeviceType>::~PairCoulDebyeKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_cutsq, cutsq);
+ memoryKK->destroy_kokkos(k_cutsq, cutsq);
}
}
@@ -93,13 +93,13 @@ void PairCoulDebyeKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -218,7 +218,7 @@ void PairCoulDebyeKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_coul**,Kokkos::LayoutRight,DeviceType>("PairCoulDebye::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_coul_dsf_kokkos.cpp b/src/KOKKOS/pair_coul_dsf_kokkos.cpp
index e6f5407f2d..f77e63bbf0 100644
--- a/src/KOKKOS/pair_coul_dsf_kokkos.cpp
+++ b/src/KOKKOS/pair_coul_dsf_kokkos.cpp
@@ -27,7 +27,7 @@
#include "neighbor.h"
#include "neigh_list_kokkos.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "update.h"
#include "integrate.h"
#include "respa.h"
@@ -65,8 +65,8 @@ template<class DeviceType>
PairCoulDSFKokkos<DeviceType>::~PairCoulDSFKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -86,13 +86,13 @@ void PairCoulDSFKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_coul_long_kokkos.cpp b/src/KOKKOS/pair_coul_long_kokkos.cpp
index 721e140e33..f2ade3f367 100644
--- a/src/KOKKOS/pair_coul_long_kokkos.cpp
+++ b/src/KOKKOS/pair_coul_long_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -72,8 +72,8 @@ template<class DeviceType>
PairCoulLongKokkos<DeviceType>::~PairCoulLongKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
k_cut_coulsq = DAT::tdual_ffloat_2d();
memory->sfree(cutsq);
@@ -110,13 +110,13 @@ void PairCoulLongKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -272,10 +272,10 @@ void PairCoulLongKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_coul**,Kokkos::LayoutRight,DeviceType>("PairCoulLong::params",n+1,n+1);
diff --git a/src/KOKKOS/pair_coul_wolf_kokkos.cpp b/src/KOKKOS/pair_coul_wolf_kokkos.cpp
index 75177e2d81..0f3e9b9429 100644
--- a/src/KOKKOS/pair_coul_wolf_kokkos.cpp
+++ b/src/KOKKOS/pair_coul_wolf_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -60,8 +60,8 @@ template<class DeviceType>
PairCoulWolfKokkos<DeviceType>::~PairCoulWolfKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -81,13 +81,13 @@ void PairCoulWolfKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp b/src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp
index c559ab412f..0fae6ffef9 100644
--- a/src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp
+++ b/src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp
@@ -28,7 +28,7 @@
#include "neighbor.h"
#include "neigh_list.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "modify.h"
#include "pair_dpd_fdt_energy_kokkos.h"
#include "error.h"
@@ -62,15 +62,15 @@ PairDPDfdtEnergyKokkos<DeviceType>::~PairDPDfdtEnergyKokkos()
{
if (copymode) return;
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
if (allocated) {
- memory->destroy_kokkos(k_duCond,duCond);
- memory->destroy_kokkos(k_duMech,duMech);
+ memoryKK->destroy_kokkos(k_duCond,duCond);
+ memoryKK->destroy_kokkos(k_duMech,duMech);
}
- memory->destroy_kokkos(k_cutsq,cutsq);
+ memoryKK->destroy_kokkos(k_cutsq,cutsq);
#ifdef DPD_USE_RAN_MARS
rand_pool.destroy();
@@ -167,13 +167,13 @@ void PairDPDfdtEnergyKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.template view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.template view<DeviceType>();
}
@@ -274,11 +274,11 @@ void PairDPDfdtEnergyKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// Allocate memory for duCond and duMech
if (allocated) {
- memory->destroy_kokkos(k_duCond,duCond);
- memory->destroy_kokkos(k_duMech,duMech);
+ memoryKK->destroy_kokkos(k_duCond,duCond);
+ memoryKK->destroy_kokkos(k_duMech,duMech);
}
- memory->create_kokkos(k_duCond,duCond,nlocal+nghost,"pair:duCond");
- memory->create_kokkos(k_duMech,duMech,nlocal+nghost,"pair:duMech");
+ memoryKK->create_kokkos(k_duCond,duCond,nlocal+nghost,"pair:duCond");
+ memoryKK->create_kokkos(k_duMech,duMech,nlocal+nghost,"pair:duMech");
d_duCond = k_duCond.view<DeviceType>();
d_duMech = k_duMech.view<DeviceType>();
h_duCond = k_duCond.h_view;
@@ -641,7 +641,7 @@ void PairDPDfdtEnergyKokkos<DeviceType>::allocate()
int nghost = atom->nghost;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_dpd**,Kokkos::LayoutRight,DeviceType>("PairDPDfdtEnergy::params",n+1,n+1);
@@ -650,8 +650,8 @@ void PairDPDfdtEnergyKokkos<DeviceType>::allocate()
if (!splitFDT_flag) {
memory->destroy(duCond);
memory->destroy(duMech);
- memory->create_kokkos(k_duCond,duCond,nlocal+nghost+1,"pair:duCond");
- memory->create_kokkos(k_duMech,duMech,nlocal+nghost+1,"pair:duMech");
+ memoryKK->create_kokkos(k_duCond,duCond,nlocal+nghost+1,"pair:duCond");
+ memoryKK->create_kokkos(k_duMech,duMech,nlocal+nghost+1,"pair:duMech");
d_duCond = k_duCond.view<DeviceType>();
d_duMech = k_duMech.view<DeviceType>();
h_duCond = k_duCond.h_view;
diff --git a/src/KOKKOS/pair_eam_alloy_kokkos.cpp b/src/KOKKOS/pair_eam_alloy_kokkos.cpp
index aa68d0a054..f263cbb6d9 100644
--- a/src/KOKKOS/pair_eam_alloy_kokkos.cpp
+++ b/src/KOKKOS/pair_eam_alloy_kokkos.cpp
@@ -28,7 +28,7 @@
#include "neighbor.h"
#include "neigh_list_kokkos.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -59,8 +59,8 @@ template<class DeviceType>
PairEAMAlloyKokkos<DeviceType>::~PairEAMAlloyKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -80,13 +80,13 @@ void PairEAMAlloyKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_eam_fs_kokkos.cpp b/src/KOKKOS/pair_eam_fs_kokkos.cpp
index a982f94ec4..05ec1a644b 100644
--- a/src/KOKKOS/pair_eam_fs_kokkos.cpp
+++ b/src/KOKKOS/pair_eam_fs_kokkos.cpp
@@ -28,7 +28,7 @@
#include "neighbor.h"
#include "neigh_list_kokkos.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -59,8 +59,8 @@ template<class DeviceType>
PairEAMFSKokkos<DeviceType>::~PairEAMFSKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -80,13 +80,13 @@ void PairEAMFSKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_eam_kokkos.cpp b/src/KOKKOS/pair_eam_kokkos.cpp
index 8ac92a1766..0aa43b26b2 100644
--- a/src/KOKKOS/pair_eam_kokkos.cpp
+++ b/src/KOKKOS/pair_eam_kokkos.cpp
@@ -28,7 +28,7 @@
#include "neighbor.h"
#include "neigh_list_kokkos.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -54,8 +54,8 @@ template<class DeviceType>
PairEAMKokkos<DeviceType>::~PairEAMKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -75,13 +75,13 @@ void PairEAMKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_exp6_rx_kokkos.cpp b/src/KOKKOS/pair_exp6_rx_kokkos.cpp
index 8d65be23af..6eb15f9b94 100644
--- a/src/KOKKOS/pair_exp6_rx_kokkos.cpp
+++ b/src/KOKKOS/pair_exp6_rx_kokkos.cpp
@@ -26,7 +26,7 @@
#include "neigh_list.h"
#include "math_const.h"
#include "math_special_kokkos.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "modify.h"
#include "fix.h"
@@ -89,18 +89,18 @@ PairExp6rxKokkos<DeviceType>::~PairExp6rxKokkos()
{
if (copymode) return;
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
- memory->destroy_kokkos(k_cutsq,cutsq);
+ memoryKK->destroy_kokkos(k_cutsq,cutsq);
for (int i=0; i < nparams; ++i) {
delete[] params[i].name;
delete[] params[i].potential;
}
- memory->destroy_kokkos(k_params,params);
+ memoryKK->destroy_kokkos(k_params,params);
- memory->destroy_kokkos(k_mol2param,mol2param);
+ memoryKK->destroy_kokkos(k_mol2param,mol2param);
}
/* ---------------------------------------------------------------------- */
@@ -151,13 +151,13 @@ void PairExp6rxKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.template view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.template view<DeviceType>();
}
@@ -1660,7 +1660,7 @@ void PairExp6rxKokkos<DeviceType>::allocate()
for (int j = i; j <= ntypes; j++)
setflag[i][j] = 0;
- memory->create_kokkos(k_cutsq,cutsq,ntypes+1,ntypes+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,ntypes+1,ntypes+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_cutsq.template modify<LMPHostType>();
@@ -1697,7 +1697,7 @@ void PairExp6rxKokkos<DeviceType>::read_file(char *file)
int params_per_line = 5;
char **words = new char*[params_per_line+1];
- memory->destroy_kokkos(k_params,params);
+ memoryKK->destroy_kokkos(k_params,params);
params = NULL;
nparams = maxparam = 0;
@@ -1777,7 +1777,7 @@ void PairExp6rxKokkos<DeviceType>::read_file(char *file)
if (nparams == maxparam) {
k_params.template modify<LMPHostType>();
maxparam += DELTA;
- memory->grow_kokkos(k_params,params,maxparam,
+ memoryKK->grow_kokkos(k_params,params,maxparam,
"pair:params");
}
@@ -1816,8 +1816,8 @@ void PairExp6rxKokkos<DeviceType>::setup()
// set mol2param for all combinations
// must be a single exact match to lines read from file
- memory->destroy_kokkos(k_mol2param,mol2param);
- memory->create_kokkos(k_mol2param,mol2param,nspecies,"pair:mol2param");
+ memoryKK->destroy_kokkos(k_mol2param,mol2param);
+ memoryKK->create_kokkos(k_mol2param,mol2param,nspecies,"pair:mol2param");
for (i = 0; i < nspecies; i++) {
n = -1;
diff --git a/src/KOKKOS/pair_hybrid_kokkos.cpp b/src/KOKKOS/pair_hybrid_kokkos.cpp
index 337b56c6ce..48bdf3081d 100644
--- a/src/KOKKOS/pair_hybrid_kokkos.cpp
+++ b/src/KOKKOS/pair_hybrid_kokkos.cpp
@@ -23,7 +23,7 @@
#include "neigh_request.h"
#include "update.h"
#include "comm.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "respa.h"
#include "atom_masks.h"
diff --git a/src/KOKKOS/pair_lj_charmm_coul_charmm_implicit_kokkos.cpp b/src/KOKKOS/pair_lj_charmm_coul_charmm_implicit_kokkos.cpp
index e7cf7ba42a..7d395cb3b4 100644
--- a/src/KOKKOS/pair_lj_charmm_coul_charmm_implicit_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_charmm_coul_charmm_implicit_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -73,8 +73,8 @@ template<class DeviceType>
PairLJCharmmCoulCharmmImplicitKokkos<DeviceType>::~PairLJCharmmCoulCharmmImplicitKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
k_cut_ljsq = DAT::tdual_ffloat_2d();
k_cut_coulsq = DAT::tdual_ffloat_2d();
@@ -118,13 +118,13 @@ void PairLJCharmmCoulCharmmImplicitKokkos<DeviceType>::compute(int eflag_in, int
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -315,14 +315,14 @@ void PairLJCharmmCoulCharmmImplicitKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
//memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJCharmmCoulCharmmImplicit::params",n+1,n+1);
diff --git a/src/KOKKOS/pair_lj_charmm_coul_charmm_kokkos.cpp b/src/KOKKOS/pair_lj_charmm_coul_charmm_kokkos.cpp
index a456d6e276..2663d71a1f 100644
--- a/src/KOKKOS/pair_lj_charmm_coul_charmm_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_charmm_coul_charmm_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -73,8 +73,8 @@ template<class DeviceType>
PairLJCharmmCoulCharmmKokkos<DeviceType>::~PairLJCharmmCoulCharmmKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
k_cut_ljsq = DAT::tdual_ffloat_2d();
k_cut_coulsq = DAT::tdual_ffloat_2d();
@@ -118,13 +118,13 @@ void PairLJCharmmCoulCharmmKokkos<DeviceType>::compute(int eflag_in, int vflag_i
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -316,14 +316,14 @@ void PairLJCharmmCoulCharmmKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
//memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJCharmmCoulCharmm::params",n+1,n+1);
diff --git a/src/KOKKOS/pair_lj_charmm_coul_long_kokkos.cpp b/src/KOKKOS/pair_lj_charmm_coul_long_kokkos.cpp
index dffbbb638f..81271c7d8a 100644
--- a/src/KOKKOS/pair_lj_charmm_coul_long_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_charmm_coul_long_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -73,8 +73,8 @@ template<class DeviceType>
PairLJCharmmCoulLongKokkos<DeviceType>::~PairLJCharmmCoulLongKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
k_cut_ljsq = DAT::tdual_ffloat_2d();
k_cut_coulsq = DAT::tdual_ffloat_2d();
@@ -118,13 +118,13 @@ void PairLJCharmmCoulLongKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -336,14 +336,14 @@ void PairLJCharmmCoulLongKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
//memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJCharmmCoulLong::params",n+1,n+1);
diff --git a/src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp b/src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp
index 0081aca4f1..3f355b1b16 100644
--- a/src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp
@@ -27,7 +27,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -60,9 +60,9 @@ template<class DeviceType>
PairLJClass2CoulCutKokkos<DeviceType>::~PairLJClass2CoulCutKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_cutsq, cutsq);
- memory->destroy_kokkos(k_cut_ljsq, cut_ljsq);
- memory->destroy_kokkos(k_cut_coulsq, cut_coulsq);
+ memoryKK->destroy_kokkos(k_cutsq, cutsq);
+ memoryKK->destroy_kokkos(k_cut_ljsq, cut_ljsq);
+ memoryKK->destroy_kokkos(k_cut_coulsq, cut_coulsq);
}
}
@@ -95,13 +95,13 @@ void PairLJClass2CoulCutKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -255,13 +255,13 @@ void PairLJClass2CoulCutKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
memory->destroy(cut_coulsq);
- memory->create_kokkos(k_cut_coulsq,cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJClass2CoulCut::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_class2_coul_long_kokkos.cpp b/src/KOKKOS/pair_lj_class2_coul_long_kokkos.cpp
index b5dc358feb..47976ec610 100644
--- a/src/KOKKOS/pair_lj_class2_coul_long_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_class2_coul_long_kokkos.cpp
@@ -27,7 +27,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -69,8 +69,8 @@ template<class DeviceType>
PairLJClass2CoulLongKokkos<DeviceType>::~PairLJClass2CoulLongKokkos()
{
if (!copymode){
- memory->destroy_kokkos(k_cutsq, cutsq);
- memory->destroy_kokkos(k_cut_ljsq, cut_ljsq);
+ memoryKK->destroy_kokkos(k_cutsq, cutsq);
+ memoryKK->destroy_kokkos(k_cut_ljsq, cut_ljsq);
}
}
@@ -103,13 +103,13 @@ void PairLJClass2CoulLongKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -307,13 +307,13 @@ void PairLJClass2CoulLongKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJClass2CoulLong::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_class2_kokkos.cpp b/src/KOKKOS/pair_lj_class2_kokkos.cpp
index 34cc15279b..5beb520c00 100644
--- a/src/KOKKOS/pair_lj_class2_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_class2_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -95,13 +95,13 @@ void PairLJClass2Kokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -200,7 +200,7 @@ void PairLJClass2Kokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj**,Kokkos::LayoutRight,DeviceType>("PairLJClass2::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp b/src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp
index c3fda01ce6..86e37bea77 100644
--- a/src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp
@@ -27,7 +27,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -60,9 +60,9 @@ template<class DeviceType>
PairLJCutCoulCutKokkos<DeviceType>::~PairLJCutCoulCutKokkos()
{
if (allocated){
- memory->destroy_kokkos(k_cutsq, cutsq);
- memory->destroy_kokkos(k_cut_ljsq, cut_ljsq);
- memory->destroy_kokkos(k_cut_coulsq, cut_coulsq);
+ memoryKK->destroy_kokkos(k_cutsq, cutsq);
+ memoryKK->destroy_kokkos(k_cut_ljsq, cut_ljsq);
+ memoryKK->destroy_kokkos(k_cut_coulsq, cut_coulsq);
}
}
@@ -95,13 +95,13 @@ void PairLJCutCoulCutKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -247,13 +247,13 @@ void PairLJCutCoulCutKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
memory->destroy(cut_coulsq);
- memory->create_kokkos(k_cut_coulsq,cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJCutCoulCut::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp b/src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp
index b7a71cb99a..76e488c686 100644
--- a/src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -64,9 +64,9 @@ template<class DeviceType>
PairLJCutCoulDebyeKokkos<DeviceType>::~PairLJCutCoulDebyeKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_cutsq, cutsq);
- memory->destroy_kokkos(k_cut_ljsq, cut_ljsq);
- memory->destroy_kokkos(k_cut_coulsq, cut_coulsq);
+ memoryKK->destroy_kokkos(k_cutsq, cutsq);
+ memoryKK->destroy_kokkos(k_cut_ljsq, cut_ljsq);
+ memoryKK->destroy_kokkos(k_cut_coulsq, cut_coulsq);
}
}
@@ -99,13 +99,13 @@ void PairLJCutCoulDebyeKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -261,13 +261,13 @@ void PairLJCutCoulDebyeKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
memory->destroy(cut_coulsq);
- memory->create_kokkos(k_cut_coulsq,cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJCutCoulDebye::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp b/src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp
index 9df5963676..0da5e7f5f5 100644
--- a/src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -72,9 +72,9 @@ template<class DeviceType>
PairLJCutCoulDSFKokkos<DeviceType>::~PairLJCutCoulDSFKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_cutsq, cutsq);
- memory->destroy_kokkos(k_cut_ljsq, cut_ljsq);
- //memory->destroy_kokkos(k_cut_coulsq, cut_coulsq);
+ memoryKK->destroy_kokkos(k_cutsq, cutsq);
+ memoryKK->destroy_kokkos(k_cut_ljsq, cut_ljsq);
+ //memoryKK->destroy_kokkos(k_cut_coulsq, cut_coulsq);
}
}
@@ -107,13 +107,13 @@ void PairLJCutCoulDSFKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -279,13 +279,13 @@ void PairLJCutCoulDSFKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
//memory->destroy(cut_coulsq);
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJCutCoulDSF::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp b/src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp
index 9bd79c7341..77a604534e 100644
--- a/src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp
@@ -27,7 +27,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -68,13 +68,13 @@ PairLJCutCoulLongKokkos<DeviceType>::PairLJCutCoulLongKokkos(LAMMPS *lmp):PairLJ
template<class DeviceType>
PairLJCutCoulLongKokkos<DeviceType>::~PairLJCutCoulLongKokkos()
{
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
eatom = NULL;
vatom = NULL;
if (allocated){
- memory->destroy_kokkos(k_cutsq, cutsq);
- memory->destroy_kokkos(k_cut_ljsq, cut_ljsq);
+ memoryKK->destroy_kokkos(k_cutsq, cutsq);
+ memoryKK->destroy_kokkos(k_cut_ljsq, cut_ljsq);
}
}
@@ -108,13 +108,13 @@ void PairLJCutCoulLongKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -304,13 +304,13 @@ void PairLJCutCoulLongKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul**,Kokkos::LayoutRight,DeviceType>("PairLJCutCoulLong::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_cut_kokkos.cpp b/src/KOKKOS/pair_lj_cut_kokkos.cpp
index c90d5ad11c..2a228fb168 100644
--- a/src/KOKKOS/pair_lj_cut_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_cut_kokkos.cpp
@@ -27,7 +27,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -57,8 +57,8 @@ template<class DeviceType>
PairLJCutKokkos<DeviceType>::~PairLJCutKokkos()
{
if (allocated) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
memory->sfree(cutsq);
eatom = NULL;
@@ -95,13 +95,13 @@ void PairLJCutKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -195,7 +195,7 @@ void PairLJCutKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj**,Kokkos::LayoutRight,DeviceType>("PairLJCut::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_expand_kokkos.cpp b/src/KOKKOS/pair_lj_expand_kokkos.cpp
index 95ec252ad5..aca7202b78 100644
--- a/src/KOKKOS/pair_lj_expand_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_expand_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -94,13 +94,13 @@ void PairLJExpandKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -202,7 +202,7 @@ void PairLJExpandKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj**,Kokkos::LayoutRight,DeviceType>("PairLJExpand::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_lj_gromacs_coul_gromacs_kokkos.cpp b/src/KOKKOS/pair_lj_gromacs_coul_gromacs_kokkos.cpp
index 4b21b08eb3..bf269288e0 100644
--- a/src/KOKKOS/pair_lj_gromacs_coul_gromacs_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_gromacs_coul_gromacs_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -64,8 +64,8 @@ template<class DeviceType>
PairLJGromacsCoulGromacsKokkos<DeviceType>::~PairLJGromacsCoulGromacsKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
k_cut_ljsq = DAT::tdual_ffloat_2d();
k_cut_coulsq = DAT::tdual_ffloat_2d();
@@ -109,13 +109,13 @@ void PairLJGromacsCoulGromacsKokkos<DeviceType>::compute(int eflag_in, int vflag
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -299,14 +299,14 @@ void PairLJGromacsCoulGromacsKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
//memory->destroy(cut_ljsq);
- memory->create_kokkos(k_cut_ljsq,n+1,n+1,"pair:cut_ljsq");
+ memoryKK->create_kokkos(k_cut_ljsq,n+1,n+1,"pair:cut_ljsq");
d_cut_ljsq = k_cut_ljsq.template view<DeviceType>();
- memory->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
+ memoryKK->create_kokkos(k_cut_coulsq,n+1,n+1,"pair:cut_coulsq");
d_cut_coulsq = k_cut_coulsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj_coul_gromacs**,Kokkos::LayoutRight,DeviceType>("PairLJGromacsCoulGromacs::params",n+1,n+1);
diff --git a/src/KOKKOS/pair_lj_gromacs_kokkos.cpp b/src/KOKKOS/pair_lj_gromacs_kokkos.cpp
index 73a17d7b22..f24ff718c1 100644
--- a/src/KOKKOS/pair_lj_gromacs_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_gromacs_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -64,8 +64,8 @@ template<class DeviceType>
PairLJGromacsKokkos<DeviceType>::~PairLJGromacsKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
k_cut_inner_sq = DAT::tdual_ffloat_2d();
memory->sfree(cutsq);
@@ -106,13 +106,13 @@ void PairLJGromacsKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -240,15 +240,15 @@ void PairLJGromacsKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
memory->destroy(cut_inner);
- memory->create_kokkos(k_cut_inner,cut_inner,n+1,n+1,"pair:cut_inner");
+ memoryKK->create_kokkos(k_cut_inner,cut_inner,n+1,n+1,"pair:cut_inner");
d_cut_inner = k_cut_inner.template view<DeviceType>();
memory->destroy(cut_inner_sq);
- memory->create_kokkos(k_cut_inner_sq,cut_inner_sq,n+1,n+1,"pair:cut_inner_sq");
+ memoryKK->create_kokkos(k_cut_inner_sq,cut_inner_sq,n+1,n+1,"pair:cut_inner_sq");
d_cut_inner_sq = k_cut_inner_sq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj**,Kokkos::LayoutRight,DeviceType>("PairLJGromacs::params",n+1,n+1);
diff --git a/src/KOKKOS/pair_lj_sdk_kokkos.cpp b/src/KOKKOS/pair_lj_sdk_kokkos.cpp
index 2063f62b20..aa579d5dc1 100644
--- a/src/KOKKOS/pair_lj_sdk_kokkos.cpp
+++ b/src/KOKKOS/pair_lj_sdk_kokkos.cpp
@@ -27,7 +27,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -94,13 +94,13 @@ void PairLJSDKKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -231,7 +231,7 @@ void PairLJSDKKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_lj**,Kokkos::LayoutRight,DeviceType>("PairLJSDK::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_morse_kokkos.cpp b/src/KOKKOS/pair_morse_kokkos.cpp
index 08a9b82640..5768d7e42c 100644
--- a/src/KOKKOS/pair_morse_kokkos.cpp
+++ b/src/KOKKOS/pair_morse_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -61,8 +61,8 @@ template<class DeviceType>
PairMorseKokkos<DeviceType>::~PairMorseKokkos()
{
if (allocated) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
k_cutsq = DAT::tdual_ffloat_2d();
memory->sfree(cutsq);
eatom = NULL;
@@ -99,13 +99,13 @@ void PairMorseKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -209,7 +209,7 @@ void PairMorseKokkos<DeviceType>::allocate()
int n = atom->ntypes;
memory->destroy(cutsq);
- memory->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,n+1,n+1,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_params = Kokkos::DualView<params_morse**,Kokkos::LayoutRight,DeviceType>("PairMorse::params",n+1,n+1);
params = k_params.template view<DeviceType>();
diff --git a/src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp b/src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp
index d9a4f1ab83..eaf28304ba 100644
--- a/src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp
+++ b/src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp
@@ -31,7 +31,7 @@
#include "force.h"
#include "comm.h"
#include "neigh_list.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "citeme.h"
#include "modify.h"
@@ -80,10 +80,10 @@ PairMultiLucyRXKokkos<DeviceType>::~PairMultiLucyRXKokkos()
{
if (copymode) return;
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
- memory->destroy_kokkos(k_cutsq,cutsq);
+ memoryKK->destroy_kokkos(k_cutsq,cutsq);
delete h_table;
delete d_table;
@@ -153,13 +153,13 @@ void PairMultiLucyRXKokkos<DeviceType>::compute_style(int eflag_in, int vflag_in
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.template view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.template view<DeviceType>();
}
@@ -864,20 +864,20 @@ void PairMultiLucyRXKokkos<DeviceType>::create_kokkos_tables()
{
const int tlm1 = tablength-1;
- memory->create_kokkos(d_table->innersq,h_table->innersq,ntables,"Table::innersq");
- memory->create_kokkos(d_table->invdelta,h_table->invdelta,ntables,"Table::invdelta");
+ memoryKK->create_kokkos(d_table->innersq,h_table->innersq,ntables,"Table::innersq");
+ memoryKK->create_kokkos(d_table->invdelta,h_table->invdelta,ntables,"Table::invdelta");
if(tabstyle == LOOKUP) {
- memory->create_kokkos(d_table->e,h_table->e,ntables,tlm1,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,tlm1,"Table::f");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tlm1,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,tlm1,"Table::f");
}
if(tabstyle == LINEAR) {
- memory->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
- memory->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
- memory->create_kokkos(d_table->de,h_table->de,ntables,tlm1,"Table::de");
- memory->create_kokkos(d_table->df,h_table->df,ntables,tlm1,"Table::df");
+ memoryKK->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
+ memoryKK->create_kokkos(d_table->de,h_table->de,ntables,tlm1,"Table::de");
+ memoryKK->create_kokkos(d_table->df,h_table->df,ntables,tlm1,"Table::df");
}
for(int i=0; i < ntables; i++) {
@@ -931,11 +931,11 @@ void PairMultiLucyRXKokkos<DeviceType>::allocate()
memory->create(setflag,nt,nt,"pair:setflag");
- memory->create_kokkos(k_cutsq,cutsq,nt,nt,"pair:cutsq");
+ memoryKK->create_kokkos(k_cutsq,cutsq,nt,nt,"pair:cutsq");
d_cutsq = k_cutsq.template view<DeviceType>();
k_cutsq.template modify<LMPHostType>();
- memory->create_kokkos(d_table->tabindex,h_table->tabindex,tabindex,nt,nt,"pair:tabindex");
+ memoryKK->create_kokkos(d_table->tabindex,h_table->tabindex,tabindex,nt,nt,"pair:tabindex");
d_table_const.tabindex = d_table->tabindex;
memset(&setflag[0][0],0,nt*nt*sizeof(int));
diff --git a/src/KOKKOS/pair_reaxc_kokkos.cpp b/src/KOKKOS/pair_reaxc_kokkos.cpp
index d5f83f4537..1407c464e3 100644
--- a/src/KOKKOS/pair_reaxc_kokkos.cpp
+++ b/src/KOKKOS/pair_reaxc_kokkos.cpp
@@ -32,7 +32,7 @@
#include "respa.h"
#include "math_const.h"
#include "math_special.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
#include "reaxc_defs.h"
@@ -81,12 +81,12 @@ PairReaxCKokkos<DeviceType>::~PairReaxCKokkos()
{
if (copymode) return;
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
- memory->destroy_kokkos(k_tmpid,tmpid);
+ memoryKK->destroy_kokkos(k_tmpid,tmpid);
tmpid = NULL;
- memory->destroy_kokkos(k_tmpbo,tmpbo);
+ memoryKK->destroy_kokkos(k_tmpbo,tmpbo);
tmpbo = NULL;
}
@@ -1339,10 +1339,10 @@ void PairReaxCKokkos<DeviceType>::allocate_array()
// FixReaxCSpecies
if (fixspecies_flag) {
- memory->destroy_kokkos(k_tmpid,tmpid);
- memory->destroy_kokkos(k_tmpbo,tmpbo);
- memory->create_kokkos(k_tmpid,tmpid,nmax,MAXSPECBOND,"pair:tmpid");
- memory->create_kokkos(k_tmpbo,tmpbo,nmax,MAXSPECBOND,"pair:tmpbo");
+ memoryKK->destroy_kokkos(k_tmpid,tmpid);
+ memoryKK->destroy_kokkos(k_tmpbo,tmpbo);
+ memoryKK->create_kokkos(k_tmpid,tmpid,nmax,MAXSPECBOND,"pair:tmpid");
+ memoryKK->create_kokkos(k_tmpbo,tmpbo,nmax,MAXSPECBOND,"pair:tmpbo");
}
// FixReaxCBonds
@@ -3905,14 +3905,14 @@ void PairReaxCKokkos<DeviceType>::ev_setup(int eflag, int vflag)
if (eflag_atom && atom->nmax > maxeatom) {
maxeatom = atom->nmax;
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
v_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom && atom->nmax > maxvatom) {
maxvatom = atom->nmax;
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
v_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_sw_kokkos.cpp b/src/KOKKOS/pair_sw_kokkos.cpp
index 3440f7c639..63b4c19ae1 100644
--- a/src/KOKKOS/pair_sw_kokkos.cpp
+++ b/src/KOKKOS/pair_sw_kokkos.cpp
@@ -27,10 +27,10 @@
#include "neigh_request.h"
#include "force.h"
#include "comm.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "neighbor.h"
#include "neigh_list_kokkos.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
#include "math_const.h"
@@ -63,8 +63,8 @@ template<class DeviceType>
PairSWKokkos<DeviceType>::~PairSWKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
eatom = NULL;
vatom = NULL;
}
@@ -86,13 +86,13 @@ void PairSWKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_table_kokkos.cpp b/src/KOKKOS/pair_table_kokkos.cpp
index 7f763baae6..b3e80d1f66 100644
--- a/src/KOKKOS/pair_table_kokkos.cpp
+++ b/src/KOKKOS/pair_table_kokkos.cpp
@@ -27,7 +27,7 @@
#include "neighbor.h"
#include "neigh_list.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -92,13 +92,13 @@ void PairTableKokkos<DeviceType>::compute_style(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -265,41 +265,41 @@ void PairTableKokkos<DeviceType>::create_kokkos_tables()
{
const int tlm1 = tablength-1;
- memory->create_kokkos(d_table->nshiftbits,h_table->nshiftbits,ntables,"Table::nshiftbits");
- memory->create_kokkos(d_table->nmask,h_table->nmask,ntables,"Table::nmask");
- memory->create_kokkos(d_table->innersq,h_table->innersq,ntables,"Table::innersq");
- memory->create_kokkos(d_table->invdelta,h_table->invdelta,ntables,"Table::invdelta");
- memory->create_kokkos(d_table->deltasq6,h_table->deltasq6,ntables,"Table::deltasq6");
+ memoryKK->create_kokkos(d_table->nshiftbits,h_table->nshiftbits,ntables,"Table::nshiftbits");
+ memoryKK->create_kokkos(d_table->nmask,h_table->nmask,ntables,"Table::nmask");
+ memoryKK->create_kokkos(d_table->innersq,h_table->innersq,ntables,"Table::innersq");
+ memoryKK->create_kokkos(d_table->invdelta,h_table->invdelta,ntables,"Table::invdelta");
+ memoryKK->create_kokkos(d_table->deltasq6,h_table->deltasq6,ntables,"Table::deltasq6");
if(tabstyle == LOOKUP) {
- memory->create_kokkos(d_table->e,h_table->e,ntables,tlm1,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,tlm1,"Table::f");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tlm1,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,tlm1,"Table::f");
}
if(tabstyle == LINEAR) {
- memory->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
- memory->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
- memory->create_kokkos(d_table->de,h_table->de,ntables,tlm1,"Table::de");
- memory->create_kokkos(d_table->df,h_table->df,ntables,tlm1,"Table::df");
+ memoryKK->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
+ memoryKK->create_kokkos(d_table->de,h_table->de,ntables,tlm1,"Table::de");
+ memoryKK->create_kokkos(d_table->df,h_table->df,ntables,tlm1,"Table::df");
}
if(tabstyle == SPLINE) {
- memory->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
- memory->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
- memory->create_kokkos(d_table->e2,h_table->e2,ntables,tablength,"Table::e2");
- memory->create_kokkos(d_table->f2,h_table->f2,ntables,tablength,"Table::f2");
+ memoryKK->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
+ memoryKK->create_kokkos(d_table->e2,h_table->e2,ntables,tablength,"Table::e2");
+ memoryKK->create_kokkos(d_table->f2,h_table->f2,ntables,tablength,"Table::f2");
}
if(tabstyle == BITMAP) {
int ntable = 1 << tablength;
- memory->create_kokkos(d_table->rsq,h_table->rsq,ntables,ntable,"Table::rsq");
- memory->create_kokkos(d_table->e,h_table->e,ntables,ntable,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,ntable,"Table::f");
- memory->create_kokkos(d_table->de,h_table->de,ntables,ntable,"Table::de");
- memory->create_kokkos(d_table->df,h_table->df,ntables,ntable,"Table::df");
- memory->create_kokkos(d_table->drsq,h_table->drsq,ntables,ntable,"Table::drsq");
+ memoryKK->create_kokkos(d_table->rsq,h_table->rsq,ntables,ntable,"Table::rsq");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,ntable,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,ntable,"Table::f");
+ memoryKK->create_kokkos(d_table->de,h_table->de,ntables,ntable,"Table::de");
+ memoryKK->create_kokkos(d_table->df,h_table->df,ntables,ntable,"Table::df");
+ memoryKK->create_kokkos(d_table->drsq,h_table->drsq,ntables,ntable,"Table::drsq");
}
@@ -410,8 +410,8 @@ void PairTableKokkos<DeviceType>::allocate()
const int nt = atom->ntypes + 1;
memory->create(setflag,nt,nt,"pair:setflag");
- memory->create_kokkos(d_table->cutsq,h_table->cutsq,cutsq,nt,nt,"pair:cutsq");
- memory->create_kokkos(d_table->tabindex,h_table->tabindex,tabindex,nt,nt,"pair:tabindex");
+ memoryKK->create_kokkos(d_table->cutsq,h_table->cutsq,cutsq,nt,nt,"pair:cutsq");
+ memoryKK->create_kokkos(d_table->tabindex,h_table->tabindex,tabindex,nt,nt,"pair:tabindex");
d_table_const.cutsq = d_table->cutsq;
d_table_const.tabindex = d_table->tabindex;
diff --git a/src/KOKKOS/pair_table_rx_kokkos.cpp b/src/KOKKOS/pair_table_rx_kokkos.cpp
index 2f5a670537..6b30d1439c 100644
--- a/src/KOKKOS/pair_table_rx_kokkos.cpp
+++ b/src/KOKKOS/pair_table_rx_kokkos.cpp
@@ -27,7 +27,7 @@
#include "neighbor.h"
#include "neigh_list.h"
#include "neigh_request.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
#include "fix.h"
@@ -162,12 +162,12 @@ PairTableRXKokkos<DeviceType>::~PairTableRXKokkos()
delete [] site1;
delete [] site2;
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
if (allocated) {
- memory->destroy_kokkos(d_table->cutsq, cutsq);
- memory->destroy_kokkos(d_table->tabindex, tabindex);
+ memoryKK->destroy_kokkos(d_table->cutsq, cutsq);
+ memoryKK->destroy_kokkos(d_table->tabindex, tabindex);
}
delete h_table;
@@ -621,13 +621,13 @@ void PairTableRXKokkos<DeviceType>::compute_style(int eflag_in, int vflag_in)
else evflag = vflag_fdotr = 0;
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.template view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.template view<DeviceType>();
}
@@ -798,41 +798,41 @@ void PairTableRXKokkos<DeviceType>::create_kokkos_tables()
{
const int tlm1 = tablength-1;
- memory->create_kokkos(d_table->nshiftbits,h_table->nshiftbits,ntables,"Table::nshiftbits");
- memory->create_kokkos(d_table->nmask,h_table->nmask,ntables,"Table::nmask");
- memory->create_kokkos(d_table->innersq,h_table->innersq,ntables,"Table::innersq");
- memory->create_kokkos(d_table->invdelta,h_table->invdelta,ntables,"Table::invdelta");
- memory->create_kokkos(d_table->deltasq6,h_table->deltasq6,ntables,"Table::deltasq6");
+ memoryKK->create_kokkos(d_table->nshiftbits,h_table->nshiftbits,ntables,"Table::nshiftbits");
+ memoryKK->create_kokkos(d_table->nmask,h_table->nmask,ntables,"Table::nmask");
+ memoryKK->create_kokkos(d_table->innersq,h_table->innersq,ntables,"Table::innersq");
+ memoryKK->create_kokkos(d_table->invdelta,h_table->invdelta,ntables,"Table::invdelta");
+ memoryKK->create_kokkos(d_table->deltasq6,h_table->deltasq6,ntables,"Table::deltasq6");
if(tabstyle == LOOKUP) {
- memory->create_kokkos(d_table->e,h_table->e,ntables,tlm1,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,tlm1,"Table::f");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tlm1,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,tlm1,"Table::f");
}
if(tabstyle == LINEAR) {
- memory->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
- memory->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
- memory->create_kokkos(d_table->de,h_table->de,ntables,tlm1,"Table::de");
- memory->create_kokkos(d_table->df,h_table->df,ntables,tlm1,"Table::df");
+ memoryKK->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
+ memoryKK->create_kokkos(d_table->de,h_table->de,ntables,tlm1,"Table::de");
+ memoryKK->create_kokkos(d_table->df,h_table->df,ntables,tlm1,"Table::df");
}
if(tabstyle == SPLINE) {
- memory->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
- memory->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
- memory->create_kokkos(d_table->e2,h_table->e2,ntables,tablength,"Table::e2");
- memory->create_kokkos(d_table->f2,h_table->f2,ntables,tablength,"Table::f2");
+ memoryKK->create_kokkos(d_table->rsq,h_table->rsq,ntables,tablength,"Table::rsq");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,tablength,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,tablength,"Table::f");
+ memoryKK->create_kokkos(d_table->e2,h_table->e2,ntables,tablength,"Table::e2");
+ memoryKK->create_kokkos(d_table->f2,h_table->f2,ntables,tablength,"Table::f2");
}
if(tabstyle == BITMAP) {
int ntable = 1 << tablength;
- memory->create_kokkos(d_table->rsq,h_table->rsq,ntables,ntable,"Table::rsq");
- memory->create_kokkos(d_table->e,h_table->e,ntables,ntable,"Table::e");
- memory->create_kokkos(d_table->f,h_table->f,ntables,ntable,"Table::f");
- memory->create_kokkos(d_table->de,h_table->de,ntables,ntable,"Table::de");
- memory->create_kokkos(d_table->df,h_table->df,ntables,ntable,"Table::df");
- memory->create_kokkos(d_table->drsq,h_table->drsq,ntables,ntable,"Table::drsq");
+ memoryKK->create_kokkos(d_table->rsq,h_table->rsq,ntables,ntable,"Table::rsq");
+ memoryKK->create_kokkos(d_table->e,h_table->e,ntables,ntable,"Table::e");
+ memoryKK->create_kokkos(d_table->f,h_table->f,ntables,ntable,"Table::f");
+ memoryKK->create_kokkos(d_table->de,h_table->de,ntables,ntable,"Table::de");
+ memoryKK->create_kokkos(d_table->df,h_table->df,ntables,ntable,"Table::df");
+ memoryKK->create_kokkos(d_table->drsq,h_table->drsq,ntables,ntable,"Table::drsq");
}
@@ -943,8 +943,8 @@ void PairTableRXKokkos<DeviceType>::allocate()
const int nt = atom->ntypes + 1;
memory->create(setflag,nt,nt,"pair:setflag");
- memory->create_kokkos(d_table->cutsq,h_table->cutsq,cutsq,nt,nt,"pair:cutsq");
- memory->create_kokkos(d_table->tabindex,h_table->tabindex,tabindex,nt,nt,"pair:tabindex");
+ memoryKK->create_kokkos(d_table->cutsq,h_table->cutsq,cutsq,nt,nt,"pair:cutsq");
+ memoryKK->create_kokkos(d_table->tabindex,h_table->tabindex,tabindex,nt,nt,"pair:tabindex");
d_table_const.cutsq = d_table->cutsq;
d_table_const.tabindex = d_table->tabindex;
diff --git a/src/KOKKOS/pair_tersoff_kokkos.cpp b/src/KOKKOS/pair_tersoff_kokkos.cpp
index 3a5c2227ef..c585da6029 100644
--- a/src/KOKKOS/pair_tersoff_kokkos.cpp
+++ b/src/KOKKOS/pair_tersoff_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -60,8 +60,8 @@ template<class DeviceType>
PairTersoffKokkos<DeviceType>::~PairTersoffKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -170,13 +170,13 @@ void PairTersoffKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_tersoff_mod_kokkos.cpp b/src/KOKKOS/pair_tersoff_mod_kokkos.cpp
index 9a59979f4c..8e718fbf6f 100644
--- a/src/KOKKOS/pair_tersoff_mod_kokkos.cpp
+++ b/src/KOKKOS/pair_tersoff_mod_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -60,8 +60,8 @@ template<class DeviceType>
PairTersoffMODKokkos<DeviceType>::~PairTersoffMODKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -170,13 +170,13 @@ void PairTersoffMODKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_tersoff_zbl_kokkos.cpp b/src/KOKKOS/pair_tersoff_zbl_kokkos.cpp
index 8468bb01f5..11a5ff1007 100644
--- a/src/KOKKOS/pair_tersoff_zbl_kokkos.cpp
+++ b/src/KOKKOS/pair_tersoff_zbl_kokkos.cpp
@@ -31,7 +31,7 @@
#include "integrate.h"
#include "respa.h"
#include "math_const.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -71,8 +71,8 @@ template<class DeviceType>
PairTersoffZBLKokkos<DeviceType>::~PairTersoffZBLKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
}
}
@@ -184,13 +184,13 @@ void PairTersoffZBLKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pair_vashishta_kokkos.cpp b/src/KOKKOS/pair_vashishta_kokkos.cpp
index fe2394ae84..e7dd01bfe4 100644
--- a/src/KOKKOS/pair_vashishta_kokkos.cpp
+++ b/src/KOKKOS/pair_vashishta_kokkos.cpp
@@ -27,10 +27,10 @@
#include "neigh_request.h"
#include "force.h"
#include "comm.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "neighbor.h"
#include "neigh_list_kokkos.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
#include "math_const.h"
@@ -62,8 +62,8 @@ template<class DeviceType>
PairVashishtaKokkos<DeviceType>::~PairVashishtaKokkos()
{
if (!copymode) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
eatom = NULL;
vatom = NULL;
}
@@ -85,13 +85,13 @@ void PairVashishtaKokkos<DeviceType>::compute(int eflag_in, int vflag_in)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
diff --git a/src/KOKKOS/pppm_kokkos.cpp b/src/KOKKOS/pppm_kokkos.cpp
index bd3ed3644f..e4fede474f 100644
--- a/src/KOKKOS/pppm_kokkos.cpp
+++ b/src/KOKKOS/pppm_kokkos.cpp
@@ -32,7 +32,7 @@
#include "domain.h"
#include "fft3d_wrap.h"
#include "remap_wrap.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include "atom_masks.h"
@@ -162,8 +162,8 @@ PPPMKokkos<DeviceType>::~PPPMKokkos()
//memory->destroy(part2grid);
//memory->destroy(acons);
- memory->destroy_kokkos(k_eatom,eatom);
- memory->destroy_kokkos(k_vatom,vatom);
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->destroy_kokkos(k_vatom,vatom);
eatom = NULL;
vatom = NULL;
}
@@ -618,13 +618,13 @@ void PPPMKokkos<DeviceType>::compute(int eflag, int vflag)
// reallocate per-atom arrays if necessary
if (eflag_atom) {
- memory->destroy_kokkos(k_eatom,eatom);
- memory->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
+ memoryKK->destroy_kokkos(k_eatom,eatom);
+ memoryKK->create_kokkos(k_eatom,eatom,maxeatom,"pair:eatom");
d_eatom = k_eatom.view<DeviceType>();
}
if (vflag_atom) {
- memory->destroy_kokkos(k_vatom,vatom);
- memory->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
+ memoryKK->destroy_kokkos(k_vatom,vatom);
+ memoryKK->create_kokkos(k_vatom,vatom,maxvatom,6,"pair:vatom");
d_vatom = k_vatom.view<DeviceType>();
}
@@ -805,12 +805,12 @@ void PPPMKokkos<DeviceType>::allocate()
{
d_density_brick = typename AT::t_FFT_SCALAR_3d("pppm:density_brick",nzhi_out-nzlo_out+1,nyhi_out-nylo_out+1,nxhi_out-nxlo_out+1);
- memory->create_kokkos(k_density_fft,density_fft,nfft_both,"pppm:d_density_fft");
+ memoryKK->create_kokkos(k_density_fft,density_fft,nfft_both,"pppm:d_density_fft");
d_density_fft = k_density_fft.view<DeviceType>();
d_greensfn = typename AT::t_float_1d("pppm:greensfn",nfft_both);
- memory->create_kokkos(k_work1,work1,2*nfft_both,"pppm:work1");
- memory->create_kokkos(k_work2,work2,2*nfft_both,"pppm:work2");
+ memoryKK->create_kokkos(k_work1,work1,2*nfft_both,"pppm:work1");
+ memoryKK->create_kokkos(k_work2,work2,2*nfft_both,"pppm:work2");
d_work1 = k_work1.view<DeviceType>();
d_work2 = k_work2.view<DeviceType>();
d_vg = typename AT::t_virial_array("pppm:vg",nfft_both);
@@ -878,13 +878,13 @@ void PPPMKokkos<DeviceType>::allocate()
template<class DeviceType>
void PPPMKokkos<DeviceType>::deallocate()
{
- memory->destroy_kokkos(d_density_fft,density_fft);
+ memoryKK->destroy_kokkos(d_density_fft,density_fft);
density_fft = NULL;
- memory->destroy_kokkos(d_greensfn,greensfn);
+ memoryKK->destroy_kokkos(d_greensfn,greensfn);
greensfn = NULL;
- memory->destroy_kokkos(d_work1,work1);
+ memoryKK->destroy_kokkos(d_work1,work1);
work1 = NULL;
- memory->destroy_kokkos(d_work2,work2);
+ memoryKK->destroy_kokkos(d_work2,work2);
work2 = NULL;
delete fft1;
diff --git a/src/KOKKOS/verlet_kokkos.cpp b/src/KOKKOS/verlet_kokkos.cpp
index adec5ff1bd..cae17cd54f 100644
--- a/src/KOKKOS/verlet_kokkos.cpp
+++ b/src/KOKKOS/verlet_kokkos.cpp
@@ -32,7 +32,7 @@
#include "compute.h"
#include "fix.h"
#include "timer.h"
-#include "memory.h"
+#include "memory_kokkos.h"
#include "error.h"
#include <ctime>
diff --git a/src/accelerator_kokkos.h b/src/accelerator_kokkos.h
index 8ea5b9d4d2..f7781c822d 100644
--- a/src/accelerator_kokkos.h
+++ b/src/accelerator_kokkos.h
@@ -25,6 +25,7 @@
#include "comm_tiled_kokkos.h"
#include "domain_kokkos.h"
#include "neighbor_kokkos.h"
+#include "memory_kokkos.h"
#include "modify_kokkos.h"
#else
@@ -37,6 +38,7 @@
#include "comm_tiled.h"
#include "domain.h"
#include "neighbor.h"
+#include "memory.h"
#include "modify.h"
namespace LAMMPS_NS {
@@ -89,6 +91,13 @@ class NeighborKokkos : public Neighbor {
~NeighborKokkos() {}
};
+class MemoryKokkos : public MemoryBrick {
+ public:
+ MemoryKokkos(class LAMMPS *lmp) : MemoryBrick(lmp) {}
+ ~MemoryKokkos() {}
+ void grow_kokkos(tagint **, tagint **, int, int, const char*) {}
+};
+
class ModifyKokkos : public Modify {
public:
ModifyKokkos(class LAMMPS *lmp) : Modify(lmp) {}
diff --git a/src/lammps.cpp b/src/lammps.cpp
index bde7ca035d..65a833d8ee 100644
--- a/src/lammps.cpp
+++ b/src/lammps.cpp
@@ -472,6 +472,9 @@ LAMMPS::LAMMPS(int narg, char **arg, MPI_Comm communicator)
kokkos = NULL;
if (kokkosflag == 1) {
+ delete memory;
+ memory = new MemoryKokkos(this);
+
kokkos = new KokkosLMP(this,kklast-kkfirst,&arg[kkfirst]);
if (!kokkos->kokkos_exists)
error->all(FLERR,"Cannot use -kokkos on without KOKKOS installed");
diff --git a/src/lammps.h b/src/lammps.h
index c432784a0b..e3b815e623 100644
--- a/src/lammps.h
+++ b/src/lammps.h
@@ -53,6 +53,7 @@ class LAMMPS {
class KokkosLMP *kokkos; // KOKKOS accelerator class
class AtomKokkos *atomKK; // KOKKOS version of Atom class
+ class MemoryKokkos *memoryKK; // KOKKOS version of Memory class
class Python * python; // Python interface
diff --git a/src/memory.h b/src/memory.h
index b83482e4c2..f2faecf6e1 100644
--- a/src/memory.h
+++ b/src/memory.h
@@ -16,9 +16,6 @@
#include "lmptype.h"
#include "pointers.h"
-#ifdef LMP_KOKKOS
-#include "kokkos_type.h"
-#endif
namespace LAMMPS_NS {
@@ -31,16 +28,6 @@ class Memory : protected Pointers {
void sfree(void *);
void fail(const char *);
- // Kokkos memory allocation functions
- // provide a dummy prototpye for any Kokkos memory function
- // called in main LAMMPS even when not built with KOKKOS package
-
-#ifdef LMP_KOKKOS
-#include "memory_kokkos.h"
-#else
- void grow_kokkos(tagint **, tagint **, int, int, const char*) {}
-#endif
-
/* ----------------------------------------------------------------------
create/grow/destroy vecs and multidim arrays with contiguous memory blocks
only use with primitive data types, e.g. 1d vec of ints, 2d array of doubles
diff --git a/src/pointers.h b/src/pointers.h
index 82b49c1dad..44967f5135 100644
--- a/src/pointers.h
+++ b/src/pointers.h
@@ -57,6 +57,7 @@ class Pointers {
screen(ptr->screen),
logfile(ptr->logfile),
atomKK(ptr->atomKK),
+ memoryKK(ptr->memoryKK),
python(ptr->python) {}
virtual ~Pointers() {}
@@ -84,6 +85,7 @@ class Pointers {
FILE *&logfile;
class AtomKokkos *&atomKK;
+ class MemoryKokkos *&memoryKK;
class Python *&python;
};
diff --git a/src/special.cpp b/src/special.cpp
index 381a763dd2..56529e748b 100644
--- a/src/special.cpp
+++ b/src/special.cpp
@@ -591,7 +591,8 @@ void Special::combine()
AtomKokkos* atomKK = (AtomKokkos*) atom;
atomKK->modified(Host,SPECIAL_MASK);
atomKK->sync(Device,SPECIAL_MASK);
- memory->grow_kokkos(atomKK->k_special,atom->special,
+ MemoryKokkos* memoryKK = (MemoryKokkos*) memory;
+ memoryKK->grow_kokkos(atomKK->k_special,atom->special,
atom->nmax,atom->maxspecial,"atom:special");
atomKK->modified(Device,SPECIAL_MASK);
atomKK->sync(Host,SPECIAL_MASK);
--
GitLab