Skip to content

GitLab

Commit 6333db3a authored by peastman's avatar peastman
Browse files

Added comments to BAOAB kernels

parent e1aec292
Show whitespace changes
Inline Side-by-side
Showing with 12 additions and 8 deletions
platforms/cuda/src/kernels/baoab.cu
View file @ 6333db3a
enum {VelScale, NoiseScale}; enum {VelScale, NoiseScale};
/** /**
* Perform the first step of BAOAB integration. * Perform the first part of BAOAB integration: velocity half step, then position half step.
*/ */
extern "C" __global__ void integrateBAOABPart1(int numAtoms, int paddedNumAtoms, mixed4* __restrict__ velm, const long long* __restrict__ force, mixed4* __restrict__ posDelta, extern "C" __global__ void integrateBAOABPart1(int numAtoms, int paddedNumAtoms, mixed4* __restrict__ velm, const long long* __restrict__ force, mixed4* __restrict__ posDelta,
...@@ -23,7 +23,8 @@ extern "C" __global__ void integrateBAOABPart1(int numAtoms, int paddedNumAtoms, ...@@ -23,7 +23,8 @@ extern "C" __global__ void integrateBAOABPart1(int numAtoms, int paddedNumAtoms,
} }
/** /**
* Perform the second step of BAOAB integration. * Perform the second part of BAOAB integration: apply constraint forces to velocities, then interact with heat bath,
* then position half step.
*/ */
extern "C" __global__ void integrateBAOABPart2(int numAtoms, real4* __restrict__ posq, real4* __restrict__ posqCorrection, mixed4* __restrict__ velm, mixed4* __restrict__ posDelta, extern "C" __global__ void integrateBAOABPart2(int numAtoms, real4* __restrict__ posq, real4* __restrict__ posqCorrection, mixed4* __restrict__ velm, mixed4* __restrict__ posDelta,
...@@ -72,7 +73,8 @@ extern "C" __global__ void integrateBAOABPart2(int numAtoms, real4* __restrict__ ...@@ -72,7 +73,8 @@ extern "C" __global__ void integrateBAOABPart2(int numAtoms, real4* __restrict__
} }
/** /**
* Perform the third step of BAOAB integration. * Perform the third part of BAOAB integration: apply constraint forces to velocities, then record
* the constrained positions in preparation for computing forces.
*/ */
extern "C" __global__ void integrateBAOABPart3(int numAtoms, real4* __restrict__ posq, real4* __restrict__ posqCorrection, mixed4* __restrict__ velm, extern "C" __global__ void integrateBAOABPart3(int numAtoms, real4* __restrict__ posq, real4* __restrict__ posqCorrection, mixed4* __restrict__ velm,
...@@ -108,7 +110,7 @@ extern "C" __global__ void integrateBAOABPart3(int numAtoms, real4* __restrict__ ...@@ -108,7 +110,7 @@ extern "C" __global__ void integrateBAOABPart3(int numAtoms, real4* __restrict__
} }
/** /**
* Perform the fourth step of BAOAB integration. * Perform the fourth part of BAOAB integration: velocity half step.
*/ */
extern "C" __global__ void integrateBAOABPart4(int numAtoms, int paddedNumAtoms, mixed4* __restrict__ velm, extern "C" __global__ void integrateBAOABPart4(int numAtoms, int paddedNumAtoms, mixed4* __restrict__ velm,
......
platforms/opencl/src/kernels/baoab.cl
View file @ 6333db3a
enum {VelScale, NoiseScale}; enum {VelScale, NoiseScale};
/** /**
* Perform the first step of BAOAB integration. * Perform the first part of BAOAB integration: velocity half step, then position half step.
*/ */
__kernel void integrateBAOABPart1(__global mixed4* restrict velm, __global const real4* restrict force, __global mixed4* restrict posDelta, __kernel void integrateBAOABPart1(__global mixed4* restrict velm, __global const real4* restrict force, __global mixed4* restrict posDelta,
...@@ -22,7 +22,8 @@ __kernel void integrateBAOABPart1(__global mixed4* restrict velm, __global const ...@@ -22,7 +22,8 @@ __kernel void integrateBAOABPart1(__global mixed4* restrict velm, __global const
} }
/** /**
* Perform the second step of BAOAB integration. * Perform the second part of BAOAB integration: apply constraint forces to velocities, then interact with heat bath,
* then position half step.
*/ */
__kernel void integrateBAOABPart2(__global real4* restrict posq, __global real4* restrict posqCorrection, __global mixed4* restrict velm, __global mixed4* restrict posDelta, __kernel void integrateBAOABPart2(__global real4* restrict posq, __global real4* restrict posqCorrection, __global mixed4* restrict velm, __global mixed4* restrict posDelta,
...@@ -67,7 +68,8 @@ __kernel void integrateBAOABPart2(__global real4* restrict posq, __global real4* ...@@ -67,7 +68,8 @@ __kernel void integrateBAOABPart2(__global real4* restrict posq, __global real4*
} }
/** /**
* Perform the third step of BAOAB integration. * Perform the third part of BAOAB integration: apply constraint forces to velocities, then record
* the constrained positions in preparation for computing forces.
*/ */
__kernel void integrateBAOABPart3(__global real4* restrict posq, __global real4* restrict posqCorrection, __global mixed4* restrict velm, __kernel void integrateBAOABPart3(__global real4* restrict posq, __global real4* restrict posqCorrection, __global mixed4* restrict velm,
...@@ -101,7 +103,7 @@ __kernel void integrateBAOABPart3(__global real4* restrict posq, __global real4* ...@@ -101,7 +103,7 @@ __kernel void integrateBAOABPart3(__global real4* restrict posq, __global real4*
} }
/** /**
* Perform the fourth step of BAOAB integration. * Perform the fourth part of BAOAB integration: velocity half step.
*/ */
__kernel void integrateBAOABPart4(__global mixed4* restrict velm, __global const real4* restrict force, __global const mixed2* restrict dt) { __kernel void integrateBAOABPart4(__global mixed4* restrict velm, __global const real4* restrict force, __global const mixed2* restrict dt) {
......
Markdown is supported
0% or .
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment