diff --git a/src/KOKKOS/angle_charmm_kokkos.cpp b/src/KOKKOS/angle_charmm_kokkos.cpp index 8d7a4d23d6..288e5a475b 100644 --- a/src/KOKKOS/angle_charmm_kokkos.cpp +++ b/src/KOKKOS/angle_charmm_kokkos.cpp @@ -446,7 +446,7 @@ void AngleCharmmKokkos::ev_tally(EV_FLOAT &ev, const int i, const in namespace LAMMPS_NS { template class AngleCharmmKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class AngleCharmmKokkos; #endif } diff --git a/src/KOKKOS/angle_class2_kokkos.cpp b/src/KOKKOS/angle_class2_kokkos.cpp index 50d7391ec5..ebb016b31b 100644 --- a/src/KOKKOS/angle_class2_kokkos.cpp +++ b/src/KOKKOS/angle_class2_kokkos.cpp @@ -597,7 +597,7 @@ void AngleClass2Kokkos::ev_tally(EV_FLOAT &ev, const int i, const in namespace LAMMPS_NS { template class AngleClass2Kokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class AngleClass2Kokkos; #endif } diff --git a/src/KOKKOS/angle_cosine_kokkos.cpp b/src/KOKKOS/angle_cosine_kokkos.cpp index afb955694d..960988b3b5 100644 --- a/src/KOKKOS/angle_cosine_kokkos.cpp +++ b/src/KOKKOS/angle_cosine_kokkos.cpp @@ -386,7 +386,7 @@ void AngleCosineKokkos::ev_tally(EV_FLOAT &ev, const int i, const in namespace LAMMPS_NS { template class AngleCosineKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class AngleCosineKokkos; #endif } diff --git a/src/KOKKOS/angle_harmonic_kokkos.cpp b/src/KOKKOS/angle_harmonic_kokkos.cpp index 4500e9b313..bc08ff4b65 100644 --- a/src/KOKKOS/angle_harmonic_kokkos.cpp +++ b/src/KOKKOS/angle_harmonic_kokkos.cpp @@ -404,7 +404,7 @@ void AngleHarmonicKokkos::ev_tally(EV_FLOAT &ev, const int i, const namespace LAMMPS_NS { template class AngleHarmonicKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class AngleHarmonicKokkos; #endif } diff --git a/src/KOKKOS/bond_class2_kokkos.cpp b/src/KOKKOS/bond_class2_kokkos.cpp index 514c80ddfb..88504608f6 100644 --- a/src/KOKKOS/bond_class2_kokkos.cpp +++ b/src/KOKKOS/bond_class2_kokkos.cpp @@ -365,7 +365,7 @@ void BondClass2Kokkos::ev_tally(EV_FLOAT &ev, const int &i, const in namespace LAMMPS_NS { template class BondClass2Kokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class BondClass2Kokkos; #endif } diff --git a/src/KOKKOS/bond_fene_kokkos.cpp b/src/KOKKOS/bond_fene_kokkos.cpp index 864873343d..e6aa27649f 100644 --- a/src/KOKKOS/bond_fene_kokkos.cpp +++ b/src/KOKKOS/bond_fene_kokkos.cpp @@ -400,7 +400,7 @@ void BondFENEKokkos::ev_tally(EV_FLOAT &ev, const int &i, const int namespace LAMMPS_NS { template class BondFENEKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class BondFENEKokkos; #endif } diff --git a/src/KOKKOS/bond_harmonic_kokkos.cpp b/src/KOKKOS/bond_harmonic_kokkos.cpp index 4e2a5aaa05..ee1b58c91f 100644 --- a/src/KOKKOS/bond_harmonic_kokkos.cpp +++ b/src/KOKKOS/bond_harmonic_kokkos.cpp @@ -339,7 +339,7 @@ void BondHarmonicKokkos::ev_tally(EV_FLOAT &ev, const int &i, const namespace LAMMPS_NS { template class BondHarmonicKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class BondHarmonicKokkos; #endif } diff --git a/src/KOKKOS/compute_coord_atom_kokkos.cpp b/src/KOKKOS/compute_coord_atom_kokkos.cpp index d91bf5a82a..90bd151b2d 100644 --- a/src/KOKKOS/compute_coord_atom_kokkos.cpp +++ b/src/KOKKOS/compute_coord_atom_kokkos.cpp @@ -243,7 +243,7 @@ void ComputeCoordAtomKokkos::operator()(TagComputeCoordAtom; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class ComputeCoordAtomKokkos; #endif } diff --git a/src/KOKKOS/compute_orientorder_atom_kokkos.cpp b/src/KOKKOS/compute_orientorder_atom_kokkos.cpp index fa613ad699..cda8f72250 100644 --- a/src/KOKKOS/compute_orientorder_atom_kokkos.cpp +++ b/src/KOKKOS/compute_orientorder_atom_kokkos.cpp @@ -737,7 +737,7 @@ void ComputeOrientOrderAtomKokkos::check_team_size_for(int inum, int namespace LAMMPS_NS { template class ComputeOrientOrderAtomKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class ComputeOrientOrderAtomKokkos; #endif } diff --git a/src/KOKKOS/compute_temp_kokkos.cpp b/src/KOKKOS/compute_temp_kokkos.cpp index f5547d0200..0e7fe25540 100644 --- a/src/KOKKOS/compute_temp_kokkos.cpp +++ b/src/KOKKOS/compute_temp_kokkos.cpp @@ -154,7 +154,7 @@ void ComputeTempKokkos::operator()(TagComputeTempVector, cons namespace LAMMPS_NS { template class ComputeTempKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class ComputeTempKokkos; #endif } diff --git a/src/KOKKOS/dihedral_charmm_kokkos.cpp b/src/KOKKOS/dihedral_charmm_kokkos.cpp index 64ab79c305..0d8256f306 100644 --- a/src/KOKKOS/dihedral_charmm_kokkos.cpp +++ b/src/KOKKOS/dihedral_charmm_kokkos.cpp @@ -789,7 +789,7 @@ void DihedralCharmmKokkos::ev_tally(EVM_FLOAT &evm, const int i, con namespace LAMMPS_NS { template class DihedralCharmmKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class DihedralCharmmKokkos; #endif } diff --git a/src/KOKKOS/dihedral_class2_kokkos.cpp b/src/KOKKOS/dihedral_class2_kokkos.cpp index adda9c56a9..371af44e9a 100644 --- a/src/KOKKOS/dihedral_class2_kokkos.cpp +++ b/src/KOKKOS/dihedral_class2_kokkos.cpp @@ -1131,7 +1131,7 @@ void DihedralClass2Kokkos::ev_tally(EV_FLOAT &ev, const int i1, cons namespace LAMMPS_NS { template class DihedralClass2Kokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class DihedralClass2Kokkos; #endif } diff --git a/src/KOKKOS/dihedral_harmonic_kokkos.cpp b/src/KOKKOS/dihedral_harmonic_kokkos.cpp index 15e459f0d5..408fa276e4 100644 --- a/src/KOKKOS/dihedral_harmonic_kokkos.cpp +++ b/src/KOKKOS/dihedral_harmonic_kokkos.cpp @@ -530,7 +530,7 @@ void DihedralHarmonicKokkos::ev_tally(EV_FLOAT &ev, const int i1, co namespace LAMMPS_NS { template class DihedralHarmonicKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class DihedralHarmonicKokkos; #endif } diff --git a/src/KOKKOS/dihedral_opls_kokkos.cpp b/src/KOKKOS/dihedral_opls_kokkos.cpp index cc3cb321c0..309d617f0a 100644 --- a/src/KOKKOS/dihedral_opls_kokkos.cpp +++ b/src/KOKKOS/dihedral_opls_kokkos.cpp @@ -535,7 +535,7 @@ void DihedralOPLSKokkos::ev_tally(EV_FLOAT &ev, const int i1, const namespace LAMMPS_NS { template class DihedralOPLSKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class DihedralOPLSKokkos; #endif } diff --git a/src/KOKKOS/fft3d_kokkos.cpp b/src/KOKKOS/fft3d_kokkos.cpp index 04dd343af6..e49d780040 100644 --- a/src/KOKKOS/fft3d_kokkos.cpp +++ b/src/KOKKOS/fft3d_kokkos.cpp @@ -905,7 +905,7 @@ void FFT3dKokkos::fft_3d_1d_only_kokkos(typename FFT_AT::t_FFT_DATA_ namespace LAMMPS_NS { template class FFT3dKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FFT3dKokkos; #endif } diff --git a/src/KOKKOS/fftdata_kokkos.h b/src/KOKKOS/fftdata_kokkos.h index e68e888805..51c444c2a3 100644 --- a/src/KOKKOS/fftdata_kokkos.h +++ b/src/KOKKOS/fftdata_kokkos.h @@ -52,7 +52,7 @@ typedef double FFT_SCALAR; // CUFFT or KISSFFT, thus undefine all other // FFTs here, since they may be valid in fft3d.cpp -#if defined(KOKKOS_ENABLE_CUDA) +#ifdef KOKKOS_ENABLE_CUDA # if defined(FFT_FFTW) # undef FFT_FFTW # endif @@ -164,7 +164,7 @@ typedef tdual_int_64::t_dev_um t_int_64_um; }; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template <> struct FFTArrayTypes { diff --git a/src/KOKKOS/fix_dpd_energy_kokkos.cpp b/src/KOKKOS/fix_dpd_energy_kokkos.cpp index 20579a6dc5..6549430131 100644 --- a/src/KOKKOS/fix_dpd_energy_kokkos.cpp +++ b/src/KOKKOS/fix_dpd_energy_kokkos.cpp @@ -85,7 +85,7 @@ void FixDPDenergyKokkos::final_integrate() namespace LAMMPS_NS { template class FixDPDenergyKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixDPDenergyKokkos; #endif } diff --git a/src/KOKKOS/fix_enforce2d_kokkos.cpp b/src/KOKKOS/fix_enforce2d_kokkos.cpp index fca3cedb13..cc8314a00f 100644 --- a/src/KOKKOS/fix_enforce2d_kokkos.cpp +++ b/src/KOKKOS/fix_enforce2d_kokkos.cpp @@ -163,7 +163,7 @@ void FixEnforce2DKokkos::post_force_item( int i ) const namespace LAMMPS_NS { template class FixEnforce2DKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixEnforce2DKokkos; #endif } diff --git a/src/KOKKOS/fix_eos_table_rx_kokkos.cpp b/src/KOKKOS/fix_eos_table_rx_kokkos.cpp index 89561c9a92..21ee54df28 100644 --- a/src/KOKKOS/fix_eos_table_rx_kokkos.cpp +++ b/src/KOKKOS/fix_eos_table_rx_kokkos.cpp @@ -560,7 +560,7 @@ void FixEOStableRXKokkos::create_kokkos_tables() namespace LAMMPS_NS { template class FixEOStableRXKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixEOStableRXKokkos; #endif } diff --git a/src/KOKKOS/fix_freeze_kokkos.cpp b/src/KOKKOS/fix_freeze_kokkos.cpp index fb0c3841e6..f09ef28813 100644 --- a/src/KOKKOS/fix_freeze_kokkos.cpp +++ b/src/KOKKOS/fix_freeze_kokkos.cpp @@ -118,7 +118,7 @@ void FixFreezeKokkos::operator()(const int i, OriginalForce &origina namespace LAMMPS_NS { template class FixFreezeKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixFreezeKokkos; #endif } diff --git a/src/KOKKOS/fix_gravity_kokkos.cpp b/src/KOKKOS/fix_gravity_kokkos.cpp index e74595e598..4f8b57365e 100644 --- a/src/KOKKOS/fix_gravity_kokkos.cpp +++ b/src/KOKKOS/fix_gravity_kokkos.cpp @@ -115,7 +115,7 @@ void FixGravityKokkos::operator()(TagFixGravityMass, const int i, do namespace LAMMPS_NS { template class FixGravityKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixGravityKokkos; #endif } diff --git a/src/KOKKOS/fix_langevin_kokkos.cpp b/src/KOKKOS/fix_langevin_kokkos.cpp index 244b20ea18..a8cc072ffc 100644 --- a/src/KOKKOS/fix_langevin_kokkos.cpp +++ b/src/KOKKOS/fix_langevin_kokkos.cpp @@ -909,7 +909,7 @@ void FixLangevinKokkos::cleanup_copy() namespace LAMMPS_NS { template class FixLangevinKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixLangevinKokkos; #endif } diff --git a/src/KOKKOS/fix_momentum_kokkos.cpp b/src/KOKKOS/fix_momentum_kokkos.cpp index 2d4911bfda..9c9d6cb1cd 100644 --- a/src/KOKKOS/fix_momentum_kokkos.cpp +++ b/src/KOKKOS/fix_momentum_kokkos.cpp @@ -199,7 +199,7 @@ void FixMomentumKokkos::end_of_step() namespace LAMMPS_NS { template class FixMomentumKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixMomentumKokkos; #endif } diff --git a/src/KOKKOS/fix_neigh_history_kokkos.cpp b/src/KOKKOS/fix_neigh_history_kokkos.cpp index 8c33f6f2c3..848940c4b7 100644 --- a/src/KOKKOS/fix_neigh_history_kokkos.cpp +++ b/src/KOKKOS/fix_neigh_history_kokkos.cpp @@ -343,7 +343,7 @@ int FixNeighHistoryKokkos::unpack_exchange(int nlocal, double *buf) namespace LAMMPS_NS { template class FixNeighHistoryKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixNeighHistoryKokkos; #endif } diff --git a/src/KOKKOS/fix_nh_kokkos.cpp b/src/KOKKOS/fix_nh_kokkos.cpp index 578f2f5c70..46993994b1 100644 --- a/src/KOKKOS/fix_nh_kokkos.cpp +++ b/src/KOKKOS/fix_nh_kokkos.cpp @@ -735,7 +735,7 @@ void FixNHKokkos::pre_exchange() namespace LAMMPS_NS { template class FixNHKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixNHKokkos; #endif } diff --git a/src/KOKKOS/fix_nph_kokkos.cpp b/src/KOKKOS/fix_nph_kokkos.cpp index 3830860fd7..d667e29221 100644 --- a/src/KOKKOS/fix_nph_kokkos.cpp +++ b/src/KOKKOS/fix_nph_kokkos.cpp @@ -71,7 +71,7 @@ FixNPHKokkos::FixNPHKokkos(LAMMPS *lmp, int narg, char **arg) : namespace LAMMPS_NS { template class FixNPHKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixNPHKokkos; #endif } diff --git a/src/KOKKOS/fix_npt_kokkos.cpp b/src/KOKKOS/fix_npt_kokkos.cpp index 5b751c9e4d..47b3da4efe 100644 --- a/src/KOKKOS/fix_npt_kokkos.cpp +++ b/src/KOKKOS/fix_npt_kokkos.cpp @@ -71,7 +71,7 @@ FixNPTKokkos::FixNPTKokkos(LAMMPS *lmp, int narg, char **arg) : namespace LAMMPS_NS { template class FixNPTKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixNPTKokkos; #endif } diff --git a/src/KOKKOS/fix_nve_kokkos.cpp b/src/KOKKOS/fix_nve_kokkos.cpp index f94b68dbce..3198557ee7 100644 --- a/src/KOKKOS/fix_nve_kokkos.cpp +++ b/src/KOKKOS/fix_nve_kokkos.cpp @@ -172,7 +172,7 @@ void FixNVEKokkos::cleanup_copy() namespace LAMMPS_NS { template class FixNVEKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixNVEKokkos; #endif } diff --git a/src/KOKKOS/fix_nve_sphere_kokkos.cpp b/src/KOKKOS/fix_nve_sphere_kokkos.cpp index 95945a858f..13d9fef69c 100644 --- a/src/KOKKOS/fix_nve_sphere_kokkos.cpp +++ b/src/KOKKOS/fix_nve_sphere_kokkos.cpp @@ -149,7 +149,7 @@ void FixNVESphereKokkos::final_integrate_item(const int i) const namespace LAMMPS_NS { template class FixNVESphereKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixNVESphereKokkos; #endif } diff --git a/src/KOKKOS/fix_nvt_kokkos.cpp b/src/KOKKOS/fix_nvt_kokkos.cpp index 66165dd7bc..013095abc0 100644 --- a/src/KOKKOS/fix_nvt_kokkos.cpp +++ b/src/KOKKOS/fix_nvt_kokkos.cpp @@ -52,7 +52,7 @@ FixNVTKokkos::FixNVTKokkos(LAMMPS *lmp, int narg, char **arg) : namespace LAMMPS_NS { template class FixNVTKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixNVTKokkos; #endif } diff --git a/src/KOKKOS/fix_qeq_reax_kokkos.cpp b/src/KOKKOS/fix_qeq_reax_kokkos.cpp index d0aedd5fc2..c7159d2637 100644 --- a/src/KOKKOS/fix_qeq_reax_kokkos.cpp +++ b/src/KOKKOS/fix_qeq_reax_kokkos.cpp @@ -1524,7 +1524,7 @@ int FixQEqReaxKokkos::unpack_exchange(int nlocal, double *buf) namespace LAMMPS_NS { template class FixQEqReaxKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixQEqReaxKokkos; #endif } diff --git a/src/KOKKOS/fix_rx_kokkos.cpp b/src/KOKKOS/fix_rx_kokkos.cpp index 2ec3e3a77c..0f7739cd67 100644 --- a/src/KOKKOS/fix_rx_kokkos.cpp +++ b/src/KOKKOS/fix_rx_kokkos.cpp @@ -2275,7 +2275,7 @@ void FixRxKokkos::unpack_reverse_comm(int n, int *list, double *buf) namespace LAMMPS_NS { template class FixRxKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixRxKokkos; #endif } diff --git a/src/KOKKOS/fix_setforce_kokkos.cpp b/src/KOKKOS/fix_setforce_kokkos.cpp index 4763f04e70..da516f5ba2 100644 --- a/src/KOKKOS/fix_setforce_kokkos.cpp +++ b/src/KOKKOS/fix_setforce_kokkos.cpp @@ -184,7 +184,7 @@ void FixSetForceKokkos::operator()(TagFixSetForceNonConstant, const namespace LAMMPS_NS { template class FixSetForceKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixSetForceKokkos; #endif } diff --git a/src/KOKKOS/fix_shardlow_kokkos.cpp b/src/KOKKOS/fix_shardlow_kokkos.cpp index c6ad47501a..c44294b2ff 100644 --- a/src/KOKKOS/fix_shardlow_kokkos.cpp +++ b/src/KOKKOS/fix_shardlow_kokkos.cpp @@ -800,7 +800,7 @@ double FixShardlowKokkos::memory_usage() namespace LAMMPS_NS { template class FixShardlowKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixShardlowKokkos; #endif } diff --git a/src/KOKKOS/fix_wall_lj93_kokkos.cpp b/src/KOKKOS/fix_wall_lj93_kokkos.cpp index 61346144c8..4acf16c15b 100644 --- a/src/KOKKOS/fix_wall_lj93_kokkos.cpp +++ b/src/KOKKOS/fix_wall_lj93_kokkos.cpp @@ -98,7 +98,7 @@ void FixWallLJ93Kokkos::wall_particle_item(int i, value_type ewall) namespace LAMMPS_NS { template class FixWallLJ93Kokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixWallLJ93Kokkos; #endif } diff --git a/src/KOKKOS/fix_wall_reflect_kokkos.cpp b/src/KOKKOS/fix_wall_reflect_kokkos.cpp index ba104d19a9..a5bd6e6e6c 100644 --- a/src/KOKKOS/fix_wall_reflect_kokkos.cpp +++ b/src/KOKKOS/fix_wall_reflect_kokkos.cpp @@ -107,7 +107,7 @@ void FixWallReflectKokkos::operator()(TagFixWallReflectPostIntegrate namespace LAMMPS_NS { template class FixWallReflectKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class FixWallReflectKokkos; #endif } diff --git a/src/KOKKOS/gridcomm_kokkos.cpp b/src/KOKKOS/gridcomm_kokkos.cpp index bdf816b647..024428366e 100644 --- a/src/KOKKOS/gridcomm_kokkos.cpp +++ b/src/KOKKOS/gridcomm_kokkos.cpp @@ -652,7 +652,7 @@ double GridCommKokkos::memory_usage() namespace LAMMPS_NS { template class GridCommKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class GridCommKokkos; #endif } diff --git a/src/KOKKOS/improper_class2_kokkos.cpp b/src/KOKKOS/improper_class2_kokkos.cpp index fcc9592b19..320caecc83 100644 --- a/src/KOKKOS/improper_class2_kokkos.cpp +++ b/src/KOKKOS/improper_class2_kokkos.cpp @@ -1126,7 +1126,7 @@ void ImproperClass2Kokkos::ev_tally(EV_FLOAT &ev, const int i1, cons namespace LAMMPS_NS { template class ImproperClass2Kokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class ImproperClass2Kokkos; #endif } diff --git a/src/KOKKOS/improper_harmonic_kokkos.cpp b/src/KOKKOS/improper_harmonic_kokkos.cpp index 6f2969722a..124e56120b 100644 --- a/src/KOKKOS/improper_harmonic_kokkos.cpp +++ b/src/KOKKOS/improper_harmonic_kokkos.cpp @@ -477,7 +477,7 @@ void ImproperHarmonicKokkos::ev_tally(EV_FLOAT &ev, const int i1, co namespace LAMMPS_NS { template class ImproperHarmonicKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class ImproperHarmonicKokkos; #endif } diff --git a/src/KOKKOS/kissfft_kokkos.cpp b/src/KOKKOS/kissfft_kokkos.cpp index 5cfb3aa64c..ac8b1ef9cc 100644 --- a/src/KOKKOS/kissfft_kokkos.cpp +++ b/src/KOKKOS/kissfft_kokkos.cpp @@ -14,7 +14,7 @@ namespace LAMMPS_NS { template class KissFFTKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class KissFFTKokkos; #endif } diff --git a/src/KOKKOS/kokkos.cpp b/src/KOKKOS/kokkos.cpp index f54df5cb76..a19b53a75b 100644 --- a/src/KOKKOS/kokkos.cpp +++ b/src/KOKKOS/kokkos.cpp @@ -100,8 +100,8 @@ KokkosLMP::KokkosLMP(LAMMPS *lmp, int narg, char **arg) : Pointers(lmp) } else if (strcmp(arg[iarg],"g") == 0 || strcmp(arg[iarg],"gpus") == 0) { -#ifndef KOKKOS_ENABLE_CUDA - error->all(FLERR,"GPUs are requested but Kokkos has not been compiled for CUDA"); +#if !defined(KOKKOS_ENABLE_CUDA) && !defined(KOKKOS_ENABLE_HIP) + error->all(FLERR,"GPUs are requested but Kokkos has not been compiled for CUDA or HIP"); #endif if (iarg+2 > narg) error->all(FLERR,"Invalid Kokkos command-line args"); ngpus = atoi(arg[iarg+1]); @@ -164,9 +164,9 @@ KokkosLMP::KokkosLMP(LAMMPS *lmp, int narg, char **arg) : Pointers(lmp) if (logfile) fprintf(logfile," will use up to %d GPU(s) per node\n",ngpus); } -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU if (ngpus <= 0) - error->all(FLERR,"Kokkos has been compiled for CUDA but no GPUs are requested"); + error->all(FLERR,"Kokkos has been compiled for CUDA or HIP but no GPUs are requested"); #endif #ifndef KOKKOS_ENABLE_SERIAL diff --git a/src/KOKKOS/kokkos_type.h b/src/KOKKOS/kokkos_type.h index 3bd8ac3f1e..54488f790d 100644 --- a/src/KOKKOS/kokkos_type.h +++ b/src/KOKKOS/kokkos_type.h @@ -29,6 +29,10 @@ enum{FULL=1u,HALFTHREAD=2u,HALF=4u}; #define ISFINITE(x) std::isfinite(x) #endif +#if defined(KOKKOS_ENABLE_CUDA) || defined(KOKKOS_ENABLE_HIP) +#define LMP_KOKKOS_GPU +#endif + #define MAX_TYPES_STACKPARAMS 12 #define NeighClusterSize 8 @@ -186,6 +190,7 @@ template<> struct ExecutionSpaceFromDevice { static const LAMMPS_NS::ExecutionSpace space = LAMMPS_NS::Host; }; + #ifdef KOKKOS_ENABLE_CUDA template<> struct ExecutionSpaceFromDevice { @@ -193,6 +198,13 @@ struct ExecutionSpaceFromDevice { }; #endif +#if defined(KOKKOS_ENABLE_HIP) +template<> +struct ExecutionSpaceFromDevice { + static const LAMMPS_NS::ExecutionSpace space = LAMMPS_NS::Device; +}; +#endif + // Determine memory traits for force array // Do atomic trait when running HALFTHREAD neighbor list style @@ -221,6 +233,13 @@ struct AtomicDup { }; #endif +#if defined(KOKKOS_ENABLE_HIP) +template<> +struct AtomicDup { + enum {value = Kokkos::Experimental::ScatterAtomic}; +}; +#endif + #ifdef LMP_KOKKOS_USE_ATOMICS #ifdef KOKKOS_ENABLE_OPENMP @@ -753,7 +772,7 @@ typedef tdual_neighbors_2d::t_dev_const_randomread t_neighbors_2d_randomread; }; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template <> struct ArrayTypes { @@ -1087,7 +1106,7 @@ typedef SNAComplex SNAcomplex; #define ISFINITE(x) std::isfinite(x) #endif -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU #define LAMMPS_LAMBDA [=] __device__ #else #define LAMMPS_LAMBDA [=] diff --git a/src/KOKKOS/nbin_kokkos.cpp b/src/KOKKOS/nbin_kokkos.cpp index 5fc3a89732..9b41dc1bff 100644 --- a/src/KOKKOS/nbin_kokkos.cpp +++ b/src/KOKKOS/nbin_kokkos.cpp @@ -147,7 +147,7 @@ void NBinKokkos::binatomsItem(const int &i) const namespace LAMMPS_NS { template class NBinKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NBinKokkos; #endif } diff --git a/src/KOKKOS/nbin_ssa_kokkos.cpp b/src/KOKKOS/nbin_ssa_kokkos.cpp index 0e46f7fc61..2d9411e707 100644 --- a/src/KOKKOS/nbin_ssa_kokkos.cpp +++ b/src/KOKKOS/nbin_ssa_kokkos.cpp @@ -298,7 +298,7 @@ void NBinSSAKokkos::sortBin( namespace LAMMPS_NS { template class NBinSSAKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NBinSSAKokkos; #endif } diff --git a/src/KOKKOS/neigh_bond_kokkos.cpp b/src/KOKKOS/neigh_bond_kokkos.cpp index 34858ff1e1..b815a084fb 100644 --- a/src/KOKKOS/neigh_bond_kokkos.cpp +++ b/src/KOKKOS/neigh_bond_kokkos.cpp @@ -1305,7 +1305,7 @@ void NeighBondKokkos::update_domain_variables() namespace LAMMPS_NS { template class NeighBondKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NeighBondKokkos; #endif } diff --git a/src/KOKKOS/neigh_list_kokkos.cpp b/src/KOKKOS/neigh_list_kokkos.cpp index afaf0dd1b8..ab0cec15aa 100644 --- a/src/KOKKOS/neigh_list_kokkos.cpp +++ b/src/KOKKOS/neigh_list_kokkos.cpp @@ -52,7 +52,7 @@ void NeighListKokkos::grow(int nmax) namespace LAMMPS_NS { template class NeighListKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NeighListKokkos; #endif } diff --git a/src/KOKKOS/npair_copy_kokkos.cpp b/src/KOKKOS/npair_copy_kokkos.cpp index 016d68400f..98f4b29c1c 100644 --- a/src/KOKKOS/npair_copy_kokkos.cpp +++ b/src/KOKKOS/npair_copy_kokkos.cpp @@ -122,7 +122,7 @@ void NPairCopyKokkos::copy_to_cpu(NeighList *list) namespace LAMMPS_NS { template class NPairCopyKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NPairCopyKokkos; #endif } diff --git a/src/KOKKOS/npair_halffull_kokkos.cpp b/src/KOKKOS/npair_halffull_kokkos.cpp index 754a5ca010..b109010935 100644 --- a/src/KOKKOS/npair_halffull_kokkos.cpp +++ b/src/KOKKOS/npair_halffull_kokkos.cpp @@ -123,7 +123,7 @@ void NPairHalffullKokkos::operator()(TagNPairHalffullCompute, namespace LAMMPS_NS { template class NPairHalffullKokkos; template class NPairHalffullKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NPairHalffullKokkos; template class NPairHalffullKokkos; #endif diff --git a/src/KOKKOS/npair_kokkos.cpp b/src/KOKKOS/npair_kokkos.cpp index 0a2d05096c..66daccc981 100644 --- a/src/KOKKOS/npair_kokkos.cpp +++ b/src/KOKKOS/npair_kokkos.cpp @@ -1154,7 +1154,7 @@ template class NPairKokkos; template class NPairKokkos; template class NPairKokkos; template class NPairKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NPairKokkos; template class NPairKokkos; template class NPairKokkos; diff --git a/src/KOKKOS/npair_skip_kokkos.cpp b/src/KOKKOS/npair_skip_kokkos.cpp index bb393c9b95..13080a9580 100644 --- a/src/KOKKOS/npair_skip_kokkos.cpp +++ b/src/KOKKOS/npair_skip_kokkos.cpp @@ -147,7 +147,7 @@ void NPairSkipKokkos::operator()(TagNPairSkipCountLocal, const int & namespace LAMMPS_NS { template class NPairSkipKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NPairSkipKokkos; #endif } diff --git a/src/KOKKOS/npair_ssa_kokkos.cpp b/src/KOKKOS/npair_ssa_kokkos.cpp index b6d03e1445..6f7342d77a 100644 --- a/src/KOKKOS/npair_ssa_kokkos.cpp +++ b/src/KOKKOS/npair_ssa_kokkos.cpp @@ -753,7 +753,7 @@ void NPairSSAKokkosExecute::build_ghosts_onePhase(int workPhase) con namespace LAMMPS_NS { template class NPairSSAKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class NPairSSAKokkos; #endif } diff --git a/src/KOKKOS/pair_buck_coul_cut_kokkos.cpp b/src/KOKKOS/pair_buck_coul_cut_kokkos.cpp index 46c5ee82cb..a6af4708d5 100644 --- a/src/KOKKOS/pair_buck_coul_cut_kokkos.cpp +++ b/src/KOKKOS/pair_buck_coul_cut_kokkos.cpp @@ -352,7 +352,7 @@ double PairBuckCoulCutKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairBuckCoulCutKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairBuckCoulCutKokkos; #endif } diff --git a/src/KOKKOS/pair_buck_coul_long_kokkos.cpp b/src/KOKKOS/pair_buck_coul_long_kokkos.cpp index 29aca38a77..cdf3b0a6c1 100644 --- a/src/KOKKOS/pair_buck_coul_long_kokkos.cpp +++ b/src/KOKKOS/pair_buck_coul_long_kokkos.cpp @@ -515,7 +515,7 @@ double PairBuckCoulLongKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairBuckCoulLongKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairBuckCoulLongKokkos; #endif } diff --git a/src/KOKKOS/pair_buck_kokkos.cpp b/src/KOKKOS/pair_buck_kokkos.cpp index 04ebd6608b..71f63a9f92 100644 --- a/src/KOKKOS/pair_buck_kokkos.cpp +++ b/src/KOKKOS/pair_buck_kokkos.cpp @@ -264,7 +264,7 @@ double PairBuckKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairBuckKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairBuckKokkos; #endif } diff --git a/src/KOKKOS/pair_coul_cut_kokkos.cpp b/src/KOKKOS/pair_coul_cut_kokkos.cpp index 2e3f76463d..2047286181 100644 --- a/src/KOKKOS/pair_coul_cut_kokkos.cpp +++ b/src/KOKKOS/pair_coul_cut_kokkos.cpp @@ -270,7 +270,7 @@ double PairCoulCutKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairCoulCutKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairCoulCutKokkos; #endif } diff --git a/src/KOKKOS/pair_coul_debye_kokkos.cpp b/src/KOKKOS/pair_coul_debye_kokkos.cpp index a544a5fb8d..cddfc63b4d 100644 --- a/src/KOKKOS/pair_coul_debye_kokkos.cpp +++ b/src/KOKKOS/pair_coul_debye_kokkos.cpp @@ -313,7 +313,7 @@ double PairCoulDebyeKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairCoulDebyeKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairCoulDebyeKokkos; #endif } diff --git a/src/KOKKOS/pair_coul_dsf_kokkos.cpp b/src/KOKKOS/pair_coul_dsf_kokkos.cpp index 62e205d8fd..11ca10c7be 100644 --- a/src/KOKKOS/pair_coul_dsf_kokkos.cpp +++ b/src/KOKKOS/pair_coul_dsf_kokkos.cpp @@ -414,7 +414,7 @@ int PairCoulDSFKokkos::sbmask(const int& j) const { namespace LAMMPS_NS { template class PairCoulDSFKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairCoulDSFKokkos; #endif } diff --git a/src/KOKKOS/pair_coul_long_kokkos.cpp b/src/KOKKOS/pair_coul_long_kokkos.cpp index c1f630a7a1..123763ea29 100644 --- a/src/KOKKOS/pair_coul_long_kokkos.cpp +++ b/src/KOKKOS/pair_coul_long_kokkos.cpp @@ -464,7 +464,7 @@ double PairCoulLongKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairCoulLongKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairCoulLongKokkos; #endif } diff --git a/src/KOKKOS/pair_coul_wolf_kokkos.cpp b/src/KOKKOS/pair_coul_wolf_kokkos.cpp index 907d088348..3bbb32a125 100644 --- a/src/KOKKOS/pair_coul_wolf_kokkos.cpp +++ b/src/KOKKOS/pair_coul_wolf_kokkos.cpp @@ -416,7 +416,7 @@ int PairCoulWolfKokkos::sbmask(const int& j) const { namespace LAMMPS_NS { template class PairCoulWolfKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairCoulWolfKokkos; #endif } diff --git a/src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp b/src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp index 61d2707888..f4dc0e867c 100644 --- a/src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp +++ b/src/KOKKOS/pair_dpd_fdt_energy_kokkos.cpp @@ -788,7 +788,7 @@ int PairDPDfdtEnergyKokkos::sbmask(const int& j) const { namespace LAMMPS_NS { template class PairDPDfdtEnergyKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairDPDfdtEnergyKokkos; #endif } diff --git a/src/KOKKOS/pair_eam_alloy_kokkos.cpp b/src/KOKKOS/pair_eam_alloy_kokkos.cpp index bea1b28544..96a07628e8 100644 --- a/src/KOKKOS/pair_eam_alloy_kokkos.cpp +++ b/src/KOKKOS/pair_eam_alloy_kokkos.cpp @@ -1223,7 +1223,7 @@ void PairEAMAlloyKokkos::file2array_alloy() namespace LAMMPS_NS { template class PairEAMAlloyKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairEAMAlloyKokkos; #endif } diff --git a/src/KOKKOS/pair_eam_fs_kokkos.cpp b/src/KOKKOS/pair_eam_fs_kokkos.cpp index 978883ab58..4f3302b455 100644 --- a/src/KOKKOS/pair_eam_fs_kokkos.cpp +++ b/src/KOKKOS/pair_eam_fs_kokkos.cpp @@ -1233,7 +1233,7 @@ void PairEAMFSKokkos::file2array_fs() namespace LAMMPS_NS { template class PairEAMFSKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairEAMFSKokkos; #endif } diff --git a/src/KOKKOS/pair_eam_kokkos.cpp b/src/KOKKOS/pair_eam_kokkos.cpp index 99a0756610..f0ac646b6b 100644 --- a/src/KOKKOS/pair_eam_kokkos.cpp +++ b/src/KOKKOS/pair_eam_kokkos.cpp @@ -896,7 +896,7 @@ void PairEAMKokkos::ev_tally(EV_FLOAT &ev, const int &i, const int & namespace LAMMPS_NS { template class PairEAMKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairEAMKokkos; #endif } diff --git a/src/KOKKOS/pair_exp6_rx_kokkos.cpp b/src/KOKKOS/pair_exp6_rx_kokkos.cpp index 8c98e32d5b..a41ef0e6d1 100644 --- a/src/KOKKOS/pair_exp6_rx_kokkos.cpp +++ b/src/KOKKOS/pair_exp6_rx_kokkos.cpp @@ -232,7 +232,7 @@ void PairExp6rxKokkos::compute(int eflag_in, int vflag_in) } else Kokkos::parallel_for(Kokkos::RangePolicy(0,np_total),*this); -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU Kokkos::parallel_for(Kokkos::RangePolicy(0,np_total),*this); #else int errorFlag = 0; @@ -277,7 +277,7 @@ void PairExp6rxKokkos::compute(int eflag_in, int vflag_in) EV_FLOAT ev; -#ifdef KOKKOS_ENABLE_CUDA // Use atomics +#ifdef LMP_KOKKOS_GPU // Use atomics if (neighflag == HALF) { if (newton_pair) { @@ -2654,7 +2654,7 @@ int PairExp6rxKokkos::sbmask(const int& j) const { namespace LAMMPS_NS { template class PairExp6rxKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairExp6rxKokkos; #endif } diff --git a/src/KOKKOS/pair_gran_hooke_history_kokkos.cpp b/src/KOKKOS/pair_gran_hooke_history_kokkos.cpp index bc5eb30845..07ef629e77 100644 --- a/src/KOKKOS/pair_gran_hooke_history_kokkos.cpp +++ b/src/KOKKOS/pair_gran_hooke_history_kokkos.cpp @@ -585,7 +585,7 @@ void PairGranHookeHistoryKokkos::ev_tally_xyz_atom(EV_FLOAT &ev, int namespace LAMMPS_NS { template class PairGranHookeHistoryKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairGranHookeHistoryKokkos; #endif } diff --git a/src/KOKKOS/pair_kokkos.h b/src/KOKKOS/pair_kokkos.h index 2934ddf621..06458948ef 100644 --- a/src/KOKKOS/pair_kokkos.h +++ b/src/KOKKOS/pair_kokkos.h @@ -720,7 +720,7 @@ int GetTeamSize(FunctorStyle& functor, int inum, int reduce_flag, int team_size, else team_size_max = Kokkos::TeamPolicy<>(inum,Kokkos::AUTO).team_size_max(functor,Kokkos::ParallelForTag()); -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU if(team_size*vector_length > team_size_max) team_size = team_size_max/vector_length; #else 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 3996f49c29..c8c6e54a5f 100644 --- a/src/KOKKOS/pair_lj_charmm_coul_charmm_implicit_kokkos.cpp +++ b/src/KOKKOS/pair_lj_charmm_coul_charmm_implicit_kokkos.cpp @@ -520,7 +520,7 @@ double PairLJCharmmCoulCharmmImplicitKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJCharmmCoulCharmmImplicitKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJCharmmCoulCharmmImplicitKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_charmm_coul_charmm_kokkos.cpp b/src/KOKKOS/pair_lj_charmm_coul_charmm_kokkos.cpp index 44ba78fa7d..62adc84ab0 100644 --- a/src/KOKKOS/pair_lj_charmm_coul_charmm_kokkos.cpp +++ b/src/KOKKOS/pair_lj_charmm_coul_charmm_kokkos.cpp @@ -522,7 +522,7 @@ double PairLJCharmmCoulCharmmKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJCharmmCoulCharmmKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJCharmmCoulCharmmKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_charmm_coul_long_kokkos.cpp b/src/KOKKOS/pair_lj_charmm_coul_long_kokkos.cpp index 35724c7061..4c122773d7 100644 --- a/src/KOKKOS/pair_lj_charmm_coul_long_kokkos.cpp +++ b/src/KOKKOS/pair_lj_charmm_coul_long_kokkos.cpp @@ -529,7 +529,7 @@ double PairLJCharmmCoulLongKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJCharmmCoulLongKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJCharmmCoulLongKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp b/src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp index b8c72d7b88..4317029ab9 100644 --- a/src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp +++ b/src/KOKKOS/pair_lj_class2_coul_cut_kokkos.cpp @@ -355,7 +355,7 @@ double PairLJClass2CoulCutKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJClass2CoulCutKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJClass2CoulCutKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_class2_coul_long_kokkos.cpp b/src/KOKKOS/pair_lj_class2_coul_long_kokkos.cpp index 1e01a275a4..51193bf692 100644 --- a/src/KOKKOS/pair_lj_class2_coul_long_kokkos.cpp +++ b/src/KOKKOS/pair_lj_class2_coul_long_kokkos.cpp @@ -510,7 +510,7 @@ double PairLJClass2CoulLongKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJClass2CoulLongKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJClass2CoulLongKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_class2_kokkos.cpp b/src/KOKKOS/pair_lj_class2_kokkos.cpp index 5bac4c8140..a0d2fba581 100644 --- a/src/KOKKOS/pair_lj_class2_kokkos.cpp +++ b/src/KOKKOS/pair_lj_class2_kokkos.cpp @@ -282,7 +282,7 @@ double PairLJClass2Kokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJClass2Kokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJClass2Kokkos; #endif } diff --git a/src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp b/src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp index e2b57e8df7..ebfb484836 100644 --- a/src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp +++ b/src/KOKKOS/pair_lj_cut_coul_cut_kokkos.cpp @@ -347,7 +347,7 @@ double PairLJCutCoulCutKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJCutCoulCutKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJCutCoulCutKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp b/src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp index bc8369fd65..0e5b5999cd 100644 --- a/src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp +++ b/src/KOKKOS/pair_lj_cut_coul_debye_kokkos.cpp @@ -376,7 +376,7 @@ double PairLJCutCoulDebyeKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJCutCoulDebyeKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJCutCoulDebyeKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp b/src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp index feb6a58136..8690859cd8 100644 --- a/src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp +++ b/src/KOKKOS/pair_lj_cut_coul_dsf_kokkos.cpp @@ -369,7 +369,7 @@ double PairLJCutCoulDSFKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJCutCoulDSFKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJCutCoulDSFKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp b/src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp index 595f018857..8cbfe03597 100644 --- a/src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp +++ b/src/KOKKOS/pair_lj_cut_coul_long_kokkos.cpp @@ -510,7 +510,7 @@ double PairLJCutCoulLongKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJCutCoulLongKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJCutCoulLongKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_cut_kokkos.cpp b/src/KOKKOS/pair_lj_cut_kokkos.cpp index 5eb0a5b91a..8294d62f68 100644 --- a/src/KOKKOS/pair_lj_cut_kokkos.cpp +++ b/src/KOKKOS/pair_lj_cut_kokkos.cpp @@ -278,7 +278,7 @@ double PairLJCutKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJCutKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJCutKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_expand_kokkos.cpp b/src/KOKKOS/pair_lj_expand_kokkos.cpp index 0409eda401..93385b7c70 100644 --- a/src/KOKKOS/pair_lj_expand_kokkos.cpp +++ b/src/KOKKOS/pair_lj_expand_kokkos.cpp @@ -287,7 +287,7 @@ double PairLJExpandKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJExpandKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJExpandKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_gromacs_coul_gromacs_kokkos.cpp b/src/KOKKOS/pair_lj_gromacs_coul_gromacs_kokkos.cpp index de475fbc99..fa9ca453a8 100644 --- a/src/KOKKOS/pair_lj_gromacs_coul_gromacs_kokkos.cpp +++ b/src/KOKKOS/pair_lj_gromacs_coul_gromacs_kokkos.cpp @@ -509,7 +509,7 @@ double PairLJGromacsCoulGromacsKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJGromacsCoulGromacsKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJGromacsCoulGromacsKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_gromacs_kokkos.cpp b/src/KOKKOS/pair_lj_gromacs_kokkos.cpp index fa2ab64ece..ed1e38a346 100644 --- a/src/KOKKOS/pair_lj_gromacs_kokkos.cpp +++ b/src/KOKKOS/pair_lj_gromacs_kokkos.cpp @@ -341,7 +341,7 @@ double PairLJGromacsKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJGromacsKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJGromacsKokkos; #endif } diff --git a/src/KOKKOS/pair_lj_sdk_kokkos.cpp b/src/KOKKOS/pair_lj_sdk_kokkos.cpp index 70d04385d0..8b228f2c01 100644 --- a/src/KOKKOS/pair_lj_sdk_kokkos.cpp +++ b/src/KOKKOS/pair_lj_sdk_kokkos.cpp @@ -317,7 +317,7 @@ double PairLJSDKKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairLJSDKKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairLJSDKKokkos; #endif } diff --git a/src/KOKKOS/pair_morse_kokkos.cpp b/src/KOKKOS/pair_morse_kokkos.cpp index 16d4d8a0a3..ba6e9ba97d 100644 --- a/src/KOKKOS/pair_morse_kokkos.cpp +++ b/src/KOKKOS/pair_morse_kokkos.cpp @@ -295,7 +295,7 @@ double PairMorseKokkos::init_one(int i, int j) namespace LAMMPS_NS { template class PairMorseKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairMorseKokkos; #endif } diff --git a/src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp b/src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp index edf8bb38ae..ca609f72b7 100644 --- a/src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp +++ b/src/KOKKOS/pair_multi_lucy_rx_kokkos.cpp @@ -987,7 +987,7 @@ void PairMultiLucyRXKokkos::settings(int narg, char **arg) namespace LAMMPS_NS { template class PairMultiLucyRXKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairMultiLucyRXKokkos; #endif } diff --git a/src/KOKKOS/pair_reaxc_kokkos.cpp b/src/KOKKOS/pair_reaxc_kokkos.cpp index 2fdaddcf8d..201167f992 100644 --- a/src/KOKKOS/pair_reaxc_kokkos.cpp +++ b/src/KOKKOS/pair_reaxc_kokkos.cpp @@ -3958,7 +3958,7 @@ void PairReaxCKokkos::operator()(PairReaxFindBondSpecies, const int } template class PairReaxCKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairReaxCKokkos; #endif } diff --git a/src/KOKKOS/pair_snap_kokkos.cpp b/src/KOKKOS/pair_snap_kokkos.cpp index 76d6aa5462..baccd0476a 100644 --- a/src/KOKKOS/pair_snap_kokkos.cpp +++ b/src/KOKKOS/pair_snap_kokkos.cpp @@ -16,7 +16,7 @@ namespace LAMMPS_NS { template class PairSNAPKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairSNAPKokkos; #endif } diff --git a/src/KOKKOS/pair_snap_kokkos_impl.h b/src/KOKKOS/pair_snap_kokkos_impl.h index 35da0e8502..74bc26ef17 100644 --- a/src/KOKKOS/pair_snap_kokkos_impl.h +++ b/src/KOKKOS/pair_snap_kokkos_impl.h @@ -296,7 +296,7 @@ void PairSNAPKokkos::compute(int eflag_in, int vflag_in) } } else { // GPU -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU //PreUi { int vector_length = vector_length_default; diff --git a/src/KOKKOS/pair_sw_kokkos.cpp b/src/KOKKOS/pair_sw_kokkos.cpp index 4ca75ebd2e..4b5e9eafe2 100644 --- a/src/KOKKOS/pair_sw_kokkos.cpp +++ b/src/KOKKOS/pair_sw_kokkos.cpp @@ -967,7 +967,7 @@ void PairSWKokkos::ev_tally3_atom(EV_FLOAT &ev, const int &i, namespace LAMMPS_NS { template class PairSWKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairSWKokkos; #endif } diff --git a/src/KOKKOS/pair_table_kokkos.cpp b/src/KOKKOS/pair_table_kokkos.cpp index b2aaaf1566..721deb84f3 100644 --- a/src/KOKKOS/pair_table_kokkos.cpp +++ b/src/KOKKOS/pair_table_kokkos.cpp @@ -530,7 +530,7 @@ void PairTableKokkos::cleanup_copy() { namespace LAMMPS_NS { template class PairTableKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairTableKokkos; #endif diff --git a/src/KOKKOS/pair_table_rx_kokkos.cpp b/src/KOKKOS/pair_table_rx_kokkos.cpp index 2b066d7633..0efef97eb5 100644 --- a/src/KOKKOS/pair_table_rx_kokkos.cpp +++ b/src/KOKKOS/pair_table_rx_kokkos.cpp @@ -1297,7 +1297,7 @@ void PairTableRXKokkos::cleanup_copy() { namespace LAMMPS_NS { template class PairTableRXKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairTableRXKokkos; #endif diff --git a/src/KOKKOS/pair_tersoff_kokkos.cpp b/src/KOKKOS/pair_tersoff_kokkos.cpp index 134431aa44..910e8da4f3 100644 --- a/src/KOKKOS/pair_tersoff_kokkos.cpp +++ b/src/KOKKOS/pair_tersoff_kokkos.cpp @@ -1282,7 +1282,7 @@ int PairTersoffKokkos::sbmask(const int& j) const { namespace LAMMPS_NS { template class PairTersoffKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairTersoffKokkos; #endif } diff --git a/src/KOKKOS/pair_tersoff_mod_kokkos.cpp b/src/KOKKOS/pair_tersoff_mod_kokkos.cpp index ceb1ed1383..d51983bfc8 100644 --- a/src/KOKKOS/pair_tersoff_mod_kokkos.cpp +++ b/src/KOKKOS/pair_tersoff_mod_kokkos.cpp @@ -1285,7 +1285,7 @@ int PairTersoffMODKokkos::sbmask(const int& j) const { namespace LAMMPS_NS { template class PairTersoffMODKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairTersoffMODKokkos; #endif } diff --git a/src/KOKKOS/pair_tersoff_zbl_kokkos.cpp b/src/KOKKOS/pair_tersoff_zbl_kokkos.cpp index c7a08772c8..9c8bb6e1a5 100644 --- a/src/KOKKOS/pair_tersoff_zbl_kokkos.cpp +++ b/src/KOKKOS/pair_tersoff_zbl_kokkos.cpp @@ -1381,7 +1381,7 @@ int PairTersoffZBLKokkos::sbmask(const int& j) const { namespace LAMMPS_NS { template class PairTersoffZBLKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairTersoffZBLKokkos; #endif } diff --git a/src/KOKKOS/pair_vashishta_kokkos.cpp b/src/KOKKOS/pair_vashishta_kokkos.cpp index 1d46fe3cd2..47ffe64105 100644 --- a/src/KOKKOS/pair_vashishta_kokkos.cpp +++ b/src/KOKKOS/pair_vashishta_kokkos.cpp @@ -942,7 +942,7 @@ void PairVashishtaKokkos::ev_tally3_atom(EV_FLOAT &ev, const int &i, namespace LAMMPS_NS { template class PairVashishtaKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairVashishtaKokkos; #endif } diff --git a/src/KOKKOS/pair_yukawa_kokkos.cpp b/src/KOKKOS/pair_yukawa_kokkos.cpp index 5b641791d3..71236c807c 100644 --- a/src/KOKKOS/pair_yukawa_kokkos.cpp +++ b/src/KOKKOS/pair_yukawa_kokkos.cpp @@ -290,7 +290,7 @@ compute_evdwl(const F_FLOAT& rsq, const int& i, const int&j, namespace LAMMPS_NS { template class PairYukawaKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairYukawaKokkos; #endif } diff --git a/src/KOKKOS/pair_zbl_kokkos.cpp b/src/KOKKOS/pair_zbl_kokkos.cpp index 5e4ec398d9..8b583a75f7 100644 --- a/src/KOKKOS/pair_zbl_kokkos.cpp +++ b/src/KOKKOS/pair_zbl_kokkos.cpp @@ -423,7 +423,7 @@ void PairZBLKokkos::cleanup_copy() { namespace LAMMPS_NS { template class PairZBLKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PairZBLKokkos; #endif } diff --git a/src/KOKKOS/pppm_kokkos.cpp b/src/KOKKOS/pppm_kokkos.cpp index 8a87773816..743f2daec9 100644 --- a/src/KOKKOS/pppm_kokkos.cpp +++ b/src/KOKKOS/pppm_kokkos.cpp @@ -1627,7 +1627,7 @@ void PPPMKokkos::make_rho() nlocal = atomKK->nlocal; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU copymode = 1; Kokkos::parallel_for(Kokkos::RangePolicy(0,nlocal),*this); copymode = 0; @@ -3051,7 +3051,7 @@ double PPPMKokkos::memory_usage() namespace LAMMPS_NS { template class PPPMKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class PPPMKokkos; #endif } diff --git a/src/KOKKOS/pppm_kokkos.h b/src/KOKKOS/pppm_kokkos.h index 847a2f807f..56cd6f5140 100644 --- a/src/KOKKOS/pppm_kokkos.h +++ b/src/KOKKOS/pppm_kokkos.h @@ -31,7 +31,7 @@ KSpaceStyle(pppm/kk/host,PPPMKokkos) // fix up FFT defines for KOKKOS with CUDA -#if defined(KOKKOS_ENABLE_CUDA) +#ifdef KOKKOS_ENABLE_CUDA # if defined(FFT_FFTW) # undef FFT_FFTW # endif diff --git a/src/KOKKOS/rand_pool_wrap_kokkos.h b/src/KOKKOS/rand_pool_wrap_kokkos.h index c1461c5216..9124d8e9e7 100644 --- a/src/KOKKOS/rand_pool_wrap_kokkos.h +++ b/src/KOKKOS/rand_pool_wrap_kokkos.h @@ -50,7 +50,7 @@ class RandPoolWrap : protected Pointers { KOKKOS_INLINE_FUNCTION RandWrap get_state() const { -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU error->all(FLERR,"Cannot use Marsaglia RNG with GPUs"); #endif diff --git a/src/KOKKOS/region_block_kokkos.cpp b/src/KOKKOS/region_block_kokkos.cpp index b28230290e..da3122bd34 100644 --- a/src/KOKKOS/region_block_kokkos.cpp +++ b/src/KOKKOS/region_block_kokkos.cpp @@ -163,7 +163,7 @@ void RegBlockKokkos::rotate(double &x, double &y, double &z, double namespace LAMMPS_NS { template class RegBlockKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class RegBlockKokkos; #endif } diff --git a/src/KOKKOS/remap_kokkos.cpp b/src/KOKKOS/remap_kokkos.cpp index 604b438a89..2894009887 100644 --- a/src/KOKKOS/remap_kokkos.cpp +++ b/src/KOKKOS/remap_kokkos.cpp @@ -529,7 +529,7 @@ void RemapKokkos::remap_3d_destroy_plan_kokkos(struct remap_plan_3d_ namespace LAMMPS_NS { template class RemapKokkos; -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU template class RemapKokkos; #endif } diff --git a/src/KOKKOS/sna_kokkos_impl.h b/src/KOKKOS/sna_kokkos_impl.h index b017751c60..93d8347c5e 100644 --- a/src/KOKKOS/sna_kokkos_impl.h +++ b/src/KOKKOS/sna_kokkos_impl.h @@ -237,7 +237,7 @@ void SNAKokkos::grow_rij(int newnatom, int newnmax) element = t_sna_2i("sna:rcutij",natom,nmax); dedr = t_sna_3d("sna:dedr",natom,nmax,3); -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU if (std::is_same::value) { // dummy allocation ulisttot = t_sna_3c_ll("sna:ulisttot",1,1,1); @@ -269,7 +269,7 @@ void SNAKokkos::grow_rij(int newnatom, int newnmax) ylist_pack_im = t_sna_4d_ll("sna:ylist_pack_im",1,1,1,1); dulist = t_sna_4c3_ll("sna:dulist",idxu_max,natom,nmax); -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU } #endif } @@ -2029,7 +2029,7 @@ double SNAKokkos::memory_usage() bytes += jdimpq*jdimpq * sizeof(double); // pqarray bytes += idxcg_max * sizeof(double); // cglist -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU if (std::is_same::value) { int natom_pad = ((natom + 32 - 1) / 32) * 32; // for AoSoA layouts @@ -2055,7 +2055,7 @@ double SNAKokkos::memory_usage() bytes += natom * idxu_max * nelements * sizeof(double) * 2; // ylist bytes += natom * nmax * idxu_max * 3 * sizeof(double) * 2; // dulist -#ifdef KOKKOS_ENABLE_CUDA +#ifdef LMP_KOKKOS_GPU } #endif