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

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

Browse Source
This commit is contained in:
sjplimp
2010-04-28 21:43:36 +00:00
parent b65db86b68
commit 7900602be9
3 changed files with 40 additions and 45 deletions
Download Patch File Download Diff File Expand all files Collapse all files

71
lib/gpu/lj_gpu.cu
View File

@ -58,6 +58,29 @@ EXTERN void lj_gpu_name(const int id, const int max_nbors, char * name) {
strcpy(name,sname.c_str());
}
static bool _pc_cell_alloc;
inline void _lj_gpu_clear() {
if (_pc_cell_alloc) {
free(energy);
free(v_temp);
cudaFreeHost(f_temp);
cudaFree(d_force);
cudaFree(d_energy);
cudaFree(d_virial);
clear_cell_list(cell_list_gpu);
_pc_cell_alloc=false;
}
}
// ---------------------------------------------------------------------------
// Clear memory on host and device
// ---------------------------------------------------------------------------
EXTERN void lj_gpu_clear() {
_lj_gpu_clear();
LJMF.clear();
}
// ---------------------------------------------------------------------------
// Allocate memory on host and device and copy constants to device
// ---------------------------------------------------------------------------
@ -67,6 +90,11 @@ EXTERN bool lj_gpu_init(int &ij_size, const int ntypes, double **cutsq,double **
double *special_lj, double *boxlo, double *boxhi,
double cell_size, double skin,
const int max_nbors, const int gpu_id) {
if (LJMF.is_allocated())
lj_gpu_clear();
else
_pc_cell_alloc=false;
LJMF.gpu.init();
if (LJMF.gpu.num_devices()==0)
return false;
@ -86,22 +114,6 @@ EXTERN bool lj_gpu_init(int &ij_size, const int ntypes, double **cutsq,double **
return ret;
}
// ---------------------------------------------------------------------------
// Clear memory on host and device
// ---------------------------------------------------------------------------
EXTERN void lj_gpu_clear() {
free(energy);
free(v_temp);
cudaFreeHost(f_temp);
cudaFree(d_force);
cudaFree(d_energy);
cudaFree(d_virial);
clear_cell_list(cell_list_gpu);
LJMF.clear();
}
template <class numtyp, class acctyp>
double _lj_gpu_cell(LJMT &ljm, double **force, double *virial,
double **host_x, int *host_type, const int inum,
@ -123,29 +135,10 @@ double _lj_gpu_cell(LJMT &ljm, double **force, double *virial,
static int first_call = 1;
// allocate memory on CPU and GPU
if (first_call) {
energy = (float*) malloc(inum*sizeof(float));
v_temp = (float3*)malloc(inum*2*sizeof(float3));
cudaMallocHost((void**)&f_temp, inum*sizeof(float3));
cudaMalloc((void**)&d_force, inum*sizeof(float3));
cudaMalloc((void**)&d_energy, inum*sizeof(float));
cudaMalloc((void**)&d_virial, inum*3*sizeof(float3));
init_cell_list(cell_list_gpu, nall, ncell, blockSize);
if (first_call || ago == 0) {
first_call = 0;
}
if (!first_call && ago == 0) {
free(energy);
free(v_temp);
cudaFreeHost(f_temp);
cudaFree(d_force);
cudaFree(d_energy);
cudaFree(d_virial);
_lj_gpu_clear();
energy = (float*) malloc(inum*sizeof(float));
v_temp = (float3*)malloc(inum*2*sizeof(float3));
cudaMallocHost((void**)&f_temp, inum*sizeof(float3));
@ -154,8 +147,8 @@ double _lj_gpu_cell(LJMT &ljm, double **force, double *virial,
cudaMalloc((void**)&d_energy, inum*sizeof(float));
cudaMalloc((void**)&d_virial, inum*3*sizeof(float3));
clear_cell_list(cell_list_gpu);
init_cell_list(cell_list_gpu, nall, ncell, blockSize);
_pc_cell_alloc=true;
}
// build cell-list on GPU

2
lib/gpu/lj_gpu_memory.h
View File

@ -35,6 +35,8 @@ class LJ_GPU_Memory {
public:
LJ_GPU_Memory() : allocated(false) {}
~LJ_GPU_Memory() { clear(); }
inline bool is_allocated() { return allocated; }
/// Allocate memory on host and device
bool init(const int ij_size, const int ntypes, double **host_cutsq,

12
lib/gpu/nvc_timer.h
View File

@ -13,8 +13,6 @@
/* ----------------------------------------------------------------------
Contributing authors: Mike Brown (SNL), wmbrown@sandia.gov
Peng Wang (Nvidia), penwang@nvidia.com
Paul Crozier (SNL), pscrozi@sandia.gov
------------------------------------------------------------------------- */
#ifndef NVC_TIMER_H
@ -35,11 +33,13 @@ class NVCTimer {
}
inline void init() {
if (!initialized) {
initialized=true;
CUDA_SAFE_CALL( cudaEventCreate(&start_event) );
CUDA_SAFE_CALL( cudaEventCreate(&stop_event) );
if (initialized) {
cudaEventDestroy(start_event);
cudaEventDestroy(stop_event);
}
initialized=true;
CUDA_SAFE_CALL( cudaEventCreate(&start_event) );
CUDA_SAFE_CALL( cudaEventCreate(&stop_event) );
}
/// Start timing