/import/home/rettenbs/src/SWE/src/blocks/cuda/SWE_WavePropagationBlockCuda_kernels.hh

Go to the documentation of this file.

#ifndef SWEWAVEPROPAGATIONBLOCKCUDAKERNELS_HH_
#define SWEWAVEPROPAGATIONBLOCKCUDAKERNELS_HH_

// CUDA-kernel which computes the net-updates
__global__
void computeNetUpdatesKernel(
    const float* i_h, const float* i_hu, const float* i_hv, const float* i_b,
    float* o_hNetUpdatesLeftD,   float* o_hNetUpdatesRightD,
    float* o_huNetUpdatesLeftD,  float* o_huNetUpdatesRightD,
    float* o_hNetUpdatesBelowD,  float* o_hNetUpdatesAboveD,
    float* o_hvNetUpdatesBelowD, float* o_hvNetUpdatesAboveD,
    float* o_maximumWaveSpeeds,
    const int i_nx, const int i_ny,
    const int i_offsetX = 0, const int i_offsetY = 0,
    const int i_blockOffSetX = 0, const int i_blockOffSetY = 0
);

// CUDA-kernel which updates the unknowns
__global__
void updateUnknownsKernel(
    const float* i_hNetUpdatesLeftD,   const float* i_hNetUpdatesRightD,
    const float* i_huNetUpdatesLeftD,  const float* i_huNetUpdatesRightD,
    const float* i_hNetUpdatesBelowD,  const float* i_hNetUpdatesAboveD,
    const float* i_hvNetUpdatesBelowD, const float* i_hvNetUpdatesAboveD,
    float* io_h, float* io_hu, float* io_hv,
    const float i_updateWidthX, const float i_updateWidthY,
    const int i_nx, const int i_ny
);

// CUDA-kernel which computes the 1D position in an array from a given 2D index
__device__
inline int computeOneDPositionKernel(const int i_i, const int i_j, const int i_nx);

#endif /* SWEWAVEPROPAGATIONBLOCKCUDAKERNELS_HH_ */