Main Page   Namespace List   Class Hierarchy   Alphabetical List   Compound List   File List   Namespace Members   Compound Members   File Members   Related Pages  

CUDAMeasureRDF.cu File Reference

CUDA kernels for calculating radial distribution function histograms. More...

#include <stdio.h>
#include <stdlib.h>
#include <string.h>
#include <math.h>
#include <cuda.h>
#include "Inform.h"
#include "utilities.h"
#include "WKFThreads.h"
#include "WKFUtils.h"
#include "CUDAKernels.h"

Go to the source code of this file.

Compounds

struct  rdfthrparms

Defines

#define NBLOCK   32
#define MAXBIN   1024
#define NCUDABLOCKS   256
#define NBLOCKHIST   64
#define NCONSTBLOCK   5440
#define THREADSPERWARP   32
#define WARP_LOG_SIZE   5
#define BIN_OVERFLOW_LIMIT   2147483648
#define ull2float   __uint2float_rn

Functions

__global__ void init_hist (unsigned int *llhistg, int maxbin)
__global__ void init_hist_f (float *histdev)
__global__ void reimage_xyz (float *xyz, int natomsi, int natomsipad, float3 celld, float rmax)
__global__ void phantom_xyz (float *xyz, int natomsi, int natomsipad, float minxyz, float rmax)
void copycoordstoconstbuff (int natoms, const float *xyzh)
__device__ void addData (volatile unsigned int *s_WarpHist, unsigned int data, unsigned int threadTag)
__global__ void calculate_rdf (int usepbc, float *xyz, int natomsi, int natomsj, float3 celld, unsigned int *llhistg, int nbins, float rmin, float delr_inv)
__global__ void calculate_histogram (float *histdev, unsigned int *llhistg, int nbins)
void calculate_histogram_block (int usepbc, float *xyz, int natomsi, int natomsj, float3 celld, unsigned int *llhistg, int nbins, float rmin, float delr_inv, float *histdev, int nblockhist)
void * rdf_thread (void *voidparms)
int rdf_gpu (wkf_threadpool_t *devpool, int usepbc, int natoms1, float *xyz, int natoms2, float *xyz2, float *cell, float *hist, int nbins, float rmin, float delr)

Variables

__constant__ float xyzj [3 *NCONSTBLOCK]

Detailed Description

CUDA kernels for calculating radial distribution function histograms.

This work is described in the following paper:

"Fast Analysis of Molecular Dynamics Trajectories with Graphics Processing Units - Radial Distribution Function Histogramming" Benjamin G. Levine, John E. Stone, and Axel Kohlmeyer. Journal of Computational Physics, 230(9):3556-3569, 2011. http://dx.doi.org/10.1016/j.jcp.2011.01.048

Definition in file CUDAMeasureRDF.cu.

Define Documentation

#define BIN_OVERFLOW_LIMIT   2147483648
 

Definition at line 87 of file CUDAMeasureRDF.cu.

Referenced by rdf_thread.

#define MAXBIN   1024
 

Definition at line 55 of file CUDAMeasureRDF.cu.

Referenced by calculate_histogram, calculate_histogram_block, calculate_rdf, rdf_gpu, and rdf_thread.

#define NBLOCK   32
 

Definition at line 54 of file CUDAMeasureRDF.cu.

Referenced by calculate_histogram, calculate_rdf, init_hist, phantom_xyz, rdf_gpu, rdf_thread, and reimage_xyz.

#define NBLOCKHIST   64
 

Definition at line 77 of file CUDAMeasureRDF.cu.

Referenced by calculate_histogram, and rdf_gpu.

#define NCONSTBLOCK   5440
 

Definition at line 81 of file CUDAMeasureRDF.cu.

Referenced by rdf_gpu, and rdf_thread.

#define NCUDABLOCKS   256
 

Definition at line 72 of file CUDAMeasureRDF.cu.

Referenced by calculate_histogram, calculate_rdf, phantom_xyz, rdf_thread, and reimage_xyz.

#define THREADSPERWARP   32
 

Definition at line 84 of file CUDAMeasureRDF.cu.

Referenced by calculate_histogram, calculate_rdf, init_hist, and rdf_thread.

#define ull2float   __uint2float_rn
 

Definition at line 578 of file CUDAMeasureRDF.cu.

Referenced by calculate_histogram.

#define WARP_LOG_SIZE   5
 

Definition at line 85 of file CUDAMeasureRDF.cu.

Referenced by calculate_rdf.

Function Documentation

__device__ void addData (  volatile unsigned int *    s_WarpHist,
unsigned int    data,
unsigned int    threadTag
) 
 

Definition at line 348 of file CUDAMeasureRDF.cu.

References data.

Referenced by calculate_rdf.

__global__ void calculate_histogram (  float *    histdev,
unsigned int *    llhistg,
int    nbins
) 
 

Definition at line 585 of file CUDAMeasureRDF.cu.

References MAXBIN, NBLOCK, NBLOCKHIST, NCUDABLOCKS, THREADSPERWARP, and ull2float.

void calculate_histogram_block (  int    usepbc,
float *    xyz,
int    natomsi,
int    natomsj,
float3    celld,
unsigned int *    llhistg,
int    nbins,
float    rmin,
float    delr_inv,
float *    histdev,
int    nblockhist
) 
 

Definition at line 661 of file CUDAMeasureRDF.cu.

References MAXBIN.

Referenced by rdf_thread.

__global__ void calculate_rdf (  int    usepbc,
float *    xyz,
int    natomsi,
int    natomsj,
float3    celld,
unsigned int *    llhistg,
int    nbins,
float    rmin,
float    delr_inv
) 
 

Definition at line 367 of file CUDAMeasureRDF.cu.

References addData, MAXBIN, NBLOCK, NCUDABLOCKS, THREADSPERWARP, WARP_LOG_SIZE, and xyzj.

void copycoordstoconstbuff (  int    natoms,
const float *    xyzh
) 
 

Definition at line 335 of file CUDAMeasureRDF.cu.

References xyzj.

Referenced by rdf_thread.

__global__ void init_hist (  unsigned int *    llhistg,
int    maxbin
) 
 

Definition at line 93 of file CUDAMeasureRDF.cu.

References NBLOCK, and THREADSPERWARP.

__global__ void init_hist_f (  float *    histdev ) 
 

Definition at line 154 of file CUDAMeasureRDF.cu.

__global__ void phantom_xyz (  float *    xyz,
int    natomsi,
int    natomsipad,
float    minxyz,
float    rmax
) 
 

Definition at line 252 of file CUDAMeasureRDF.cu.

References NBLOCK, and NCUDABLOCKS.

int rdf_gpu (  wkf_threadpool_t *    devpool,
int    usepbc,
int    natoms1,
float *    xyz,
int    natoms2,
float *    xyz2,
float *    cell,
float *    hist,
int    nbins,
float    rmin,
float    delr
) 
 

Definition at line 912 of file CUDAMeasureRDF.cu.

References rdfthrparms::cell, rdfthrparms::delr, wkf_tasktile_struct::end, rdfthrparms::histarray, MAXBIN, rdfthrparms::natoms1, rdfthrparms::natoms2, rdfthrparms::nbins, NBLOCK, NBLOCKHIST, rdfthrparms::nblockhist, rdfthrparms::nblocks, NCONSTBLOCK, rdfthrparms::nhblock, NULL, rdf_thread, rdfthrparms::rmin, wkf_tasktile_struct::start, rdfthrparms::usepbc, wkf_threadpool_get_workercount, wkf_threadpool_launch, wkf_threadpool_sched_dynamic, rdfthrparms::xyz, and rdfthrparms::xyz2array.

Referenced by measure_rdf.

void* rdf_thread (  void *    voidparms )  [static]
 

Definition at line 713 of file CUDAMeasureRDF.cu.

References BIN_OVERFLOW_LIMIT, calculate_histogram_block, rdfthrparms::cell, copycoordstoconstbuff, rdfthrparms::delr, wkf_tasktile_struct::end, rdfthrparms::histarray, MAXBIN, rdfthrparms::natoms1, rdfthrparms::natoms2, rdfthrparms::nbins, NBLOCK, rdfthrparms::nblockhist, NCONSTBLOCK, NCUDABLOCKS, rdfthrparms::nhblock, NULL, rdfthrparms::rmin, wkf_tasktile_struct::start, THREADSPERWARP, rdfthrparms::usepbc, WKF_SCHED_DONE, wkf_threadpool_next_tile, wkf_threadpool_worker_getdata, wkf_threadpool_worker_getid, rdfthrparms::xyz, and rdfthrparms::xyz2array.

Referenced by rdf_gpu.

__global__ void reimage_xyz (  float *    xyz,
int    natomsi,
int    natomsipad,
float3    celld,
float    rmax
) 
 

Definition at line 166 of file CUDAMeasureRDF.cu.

References NBLOCK, and NCUDABLOCKS.

Variable Documentation

__constant__ float xyzj[3*NCONSTBLOCK] [static]
 

Definition at line 331 of file CUDAMeasureRDF.cu.

Referenced by calculate_rdf, and copycoordstoconstbuff.

Generated on Wed Apr 17 02:47:02 2024 for VMD (current) by doxygen1.2.14 written by Dimitri van Heesch, © 1997-2002