cs_blas_hip.cpp File Reference

#include "base/cs_defs.h"
#include <algorithm>
#include <assert.h>
#include <hip/hip_runtime.h>
#include <hip/hip_runtime_api.h>
#include "bft/bft_error.h"
#include "base/cs_base.h"
#include "base/cs_mem.h"
#include "base/cs_base_accel.h"
#include "base/cs_base_hip.h"
#include "base/cs_dispatch.h"
#include "alge/cs_blas_hip.h"

Include dependency graph for cs_blas_hip.cpp:

Macros
#define	BLOCKSIZE   256

Functions
template<size_t blockSize, typename T >
static __global__ void	_asum_stage_1_of_2 (cs_lnum_t n, const T *x, double *b_res)
	Compute the absolute sum of vector values, summing result over all threads of a block. More...
template<size_t blockSize, typename T >
static __global__ void	_dot_xy_stage_1_of_2 (cs_lnum_t n, const T *x, const T *y, double *b_res)
	Compute dot product x.y, summing result over all threads of a block. More...
template<size_t blockSize, typename T >
static __global__ void	_dot_xx_stage_1_of_2 (cs_lnum_t n, const T *x, double *b_res)
	Compute dot product x.x, summing result over all threads of a block. More...
static __global__ void	_scal (cs_lnum_t n, const cs_real_t *alpha, cs_real_t *restrict x)
unsigned int	_grid_size (cs_lnum_t n, unsigned int block_size)
	Compute grid size for given array and block sizes. More...
void	cs_blas_hip_finalize (void)
	Finalize HIP BLAS API. More...

Variables
static double *	_r_reduce = nullptr
static double *	_r_grid = nullptr
static unsigned int	_r_grid_size = 0
static unsigned int	_r_tuple_size = 0

Macro Definition Documentation

BLOCKSIZE

#define BLOCKSIZE   256

Function Documentation

_asum_stage_1_of_2()

static __global__ void _asum_stage_1_of_2 ( cs_lnum_t  n, const T *  x, double *  b_res  )

static

Compute the absolute sum of vector values, summing result over all threads of a block.

blockSize must be a power of 2.

Parameters

[in]	n	array size
[in]	x	x vector
[out]	b_res	result of s = x.x

_dot_xx_stage_1_of_2()

static __global__ void _dot_xx_stage_1_of_2 ( cs_lnum_t  n, const T *  x, double *  b_res  )

static

Compute dot product x.x, summing result over all threads of a block.

blockSize must be a power of 2.

Parameters

[in]	n	array size
[in]	x	x vector
[in]	y	y vector
[out]	b_res	result of s = x.x

_dot_xy_stage_1_of_2()

static __global__ void _dot_xy_stage_1_of_2 ( cs_lnum_t  n, const T *  x, const T *  y, double *  b_res  )

static

Compute dot product x.y, summing result over all threads of a block.

blockSize must be a power of 2.

Parameters

[in]	n	array size
[in]	x	x vector
[in]	y	y vector
[out]	b_res	result of s = x.x

_grid_size()

unsigned int _grid_size	(	cs_lnum_t	n,
		unsigned int	block_size
	)

Compute grid size for given array and block sizes.

Also check initialization of work arrays.

Parameters

[in]	n	size of arrays
[in]	block_size	block size for kernels

Returns

grid size for kernels

_scal()

static __global__ void _scal ( cs_lnum_t  n, const cs_real_t *  alpha, cs_real_t *restrict  x  )

static

cs_blas_hip_finalize()

void cs_blas_hip_finalize

(

void

)

Finalize HIP BLAS API.

This frees resources such as the hipBLAS handle, if used.

Variable Documentation

_r_grid

double* _r_grid = nullptr

static

_r_grid_size

unsigned int _r_grid_size = 0

static

_r_reduce

double* _r_reduce = nullptr

static

_r_tuple_size

unsigned int _r_tuple_size = 0

static