Helper functions for the CUDA linear algebra backend. More...

Namespaces
	fft

Classes
struct	spmv_alpha_beta

struct	spmv_pure

struct	type_to_type2

struct	type_to_type2< double >

struct	type_to_type2< float >

Functions
unsigned int	make_options (vcl_size_t length, bool reciprocal, bool flip_sign)

void	cuda_last_error_check (const char message, const char file, const int line)

template<typename NumericT , typename OtherT >
viennacl::backend::mem_handle::cuda_handle_type &	arg_reference (viennacl::scalar< NumericT > &s, OtherT)

template<typename NumericT , typename OtherT >
viennacl::backend::mem_handle::cuda_handle_type const &	arg_reference (viennacl::scalar< NumericT > const &s, OtherT)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, char const & > ::type	arg_reference (ArgT, char const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, unsigned char const & >::type	arg_reference (ArgT, unsigned char const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, short const & > ::type	arg_reference (ArgT, short const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, unsigned short const & >::type	arg_reference (ArgT, unsigned short const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, int const & > ::type	arg_reference (ArgT, int const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, unsigned int const & >::type	arg_reference (ArgT, unsigned int const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, long const & > ::type	arg_reference (ArgT, long const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, unsigned long const & >::type	arg_reference (ArgT, unsigned long const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, float const & > ::type	arg_reference (ArgT, float const &val)

template<typename ArgT >
viennacl::enable_if < viennacl::is_cpu_scalar < ArgT >::value, double const & > ::type	arg_reference (ArgT, double const &val)

template<typename TagT >
bool	is_unit_solve (TagT const &tag)

bool	is_unit_solve (viennacl::linalg::unit_lower_tag)

bool	is_unit_solve (viennacl::linalg::unit_upper_tag)

template<typename TagT >
bool	is_upper_solve (TagT const &tag)

bool	is_upper_solve (viennacl::linalg::upper_tag)

bool	is_upper_solve (viennacl::linalg::unit_upper_tag)

template<typename Matrix1T , typename Matrix2T , typename SolverTagT >
void	inplace_solve_impl (Matrix1T const &A, Matrix2T &B, SolverTagT const &tag)

unsigned int	get_option_for_solver_tag (viennacl::linalg::upper_tag)

unsigned int	get_option_for_solver_tag (viennacl::linalg::unit_upper_tag)

unsigned int	get_option_for_solver_tag (viennacl::linalg::lower_tag)

unsigned int	get_option_for_solver_tag (viennacl::linalg::unit_lower_tag)

template<typename MatrixT , typename VectorT >
void	inplace_solve_vector_impl (MatrixT const &mat, VectorT &vec, unsigned int options)

template<typename MatrixT1 , typename MatrixT2 , typename MatrixT3 , typename ScalarT >
void	prod_slow_kernel (const MatrixT1 &A, bool transposed_A, const MatrixT2 &B, bool transposed_B, MatrixT3 &C, ScalarT alpha, ScalarT beta)

template<typename MatrixT1 , typename MatrixT2 , typename MatrixT3 , typename ScalarT >
void	prod (const MatrixT1 &A, bool transposed_A, const MatrixT2 &B, bool transposed_B, MatrixT3 &C, ScalarT alpha, ScalarT beta)

template<typename NumericT >
__global__ void	level_scheduling_substitute_kernel (const unsigned int row_index_array, const unsigned int row_indices, const unsigned int column_indices, const NumericT elements, NumericT *vec, unsigned int size)

template<typename NumericT >
void	level_scheduling_substitute (vector< NumericT > &vec, viennacl::backend::mem_handle const &row_index_array, viennacl::backend::mem_handle const &row_buffer, viennacl::backend::mem_handle const &col_buffer, viennacl::backend::mem_handle const &element_buffer, vcl_size_t num_rows)

template<typename NumericT >
__global__ void	csr_row_info_extractor_kernel (const unsigned int row_indices, const unsigned int column_indices, const NumericT elements, NumericT result, unsigned int size, unsigned int option)

template<typename NumericT , unsigned int AligmentV>
void	row_info (compressed_matrix< NumericT, AligmentV > const &mat, vector_base< NumericT > &vec, viennacl::linalg::detail::row_info_types info_selector)

template<typename NumericT , unsigned int AlignmentV>
void	block_inplace_solve (const matrix_expression< const compressed_matrix< NumericT, AlignmentV >, const compressed_matrix< NumericT, AlignmentV >, op_trans > &L, viennacl::backend::mem_handle const &block_indices, vcl_size_t num_blocks, vector_base< NumericT > const &, vector_base< NumericT > &vec, viennacl::linalg::unit_lower_tag)

template<typename NumericT , unsigned int AlignmentV>
void	block_inplace_solve (const matrix_expression< const compressed_matrix< NumericT, AlignmentV >, const compressed_matrix< NumericT, AlignmentV >, op_trans > &U, viennacl::backend::mem_handle const &block_indices, vcl_size_t num_blocks, vector_base< NumericT > const &U_diagonal, vector_base< NumericT > &vec, viennacl::linalg::upper_tag)

template<typename NumericT >
__global__ void	coo_row_info_extractor (const unsigned int coords, const NumericT elements, const unsigned int group_boundaries, NumericT result, unsigned int option)

template<typename NumericT , unsigned int AlignmentV>
void	row_info (coordinate_matrix< NumericT, AlignmentV > const &mat, vector_base< NumericT > &vec, viennacl::linalg::detail::row_info_types info_selector)

template<typename NumericT >
void	scan_impl (vector_base< NumericT > const &input, vector_base< NumericT > &output, bool is_inclusive)
	Worker routine for scan routines. More...

Detailed Description

Helper functions for the CUDA linear algebra backend.

Function Documentation

template<typename NumericT , typename OtherT >

viennacl::backend::mem_handle::cuda_handle_type& viennacl::linalg::cuda::detail::arg_reference	(	viennacl::scalar< NumericT > &	s,
		OtherT
	)

Definition at line 188 of file common.hpp.

template<typename NumericT , typename OtherT >

viennacl::backend::mem_handle::cuda_handle_type const& viennacl::linalg::cuda::detail::arg_reference	(	viennacl::scalar< NumericT > const &	s,
		OtherT
	)

Definition at line 191 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, char const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		char const &	val
	)

Definition at line 197 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, unsigned char const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		unsigned char const &	val
	)

Definition at line 202 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, short const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		short const &	val
	)

Definition at line 207 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, unsigned short const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		unsigned short const &	val
	)

Definition at line 212 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, int const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		int const &	val
	)

Definition at line 217 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, unsigned int const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		unsigned int const &	val
	)

Definition at line 222 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, long const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		long const &	val
	)

Definition at line 227 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, unsigned long const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		unsigned long const &	val
	)

Definition at line 232 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, float const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		float const &	val
	)

Definition at line 237 of file common.hpp.

template<typename ArgT >

viennacl::enable_if< viennacl::is_cpu_scalar<ArgT>::value, double const &>::type viennacl::linalg::cuda::detail::arg_reference	(	ArgT	,
		double const &	val
	)

Definition at line 242 of file common.hpp.

template<typename NumericT , unsigned int AlignmentV>

void viennacl::linalg::cuda::detail::block_inplace_solve	(	const matrix_expression< const compressed_matrix< NumericT, AlignmentV >, const compressed_matrix< NumericT, AlignmentV >, op_trans > &	L,
		viennacl::backend::mem_handle const &	block_indices,
		vcl_size_t	num_blocks,
		vector_base< NumericT > const &	,
		vector_base< NumericT > &	vec,
		viennacl::linalg::unit_lower_tag
	)

Definition at line 970 of file sparse_matrix_operations.hpp.

template<typename NumericT , unsigned int AlignmentV>

void viennacl::linalg::cuda::detail::block_inplace_solve	(	const matrix_expression< const compressed_matrix< NumericT, AlignmentV >, const compressed_matrix< NumericT, AlignmentV >, op_trans > &	U,
		viennacl::backend::mem_handle const &	block_indices,
		vcl_size_t	num_blocks,
		vector_base< NumericT > const &	U_diagonal,
		vector_base< NumericT > &	vec,
		viennacl::linalg::upper_tag
	)

Definition at line 989 of file sparse_matrix_operations.hpp.

template<typename NumericT >

__global__ void viennacl::linalg::cuda::detail::coo_row_info_extractor	(	const unsigned int *	coords,
		const NumericT *	elements,
		const unsigned int *	group_boundaries,
		NumericT *	result,
		unsigned int	option
	)

Definition at line 1095 of file sparse_matrix_operations.hpp.

template<typename NumericT >

__global__ void viennacl::linalg::cuda::detail::csr_row_info_extractor_kernel	(	const unsigned int *	row_indices,
		const unsigned int *	column_indices,
		const NumericT *	elements,
		NumericT *	result,
		unsigned int	size,
		unsigned int	option
	)

Definition at line 54 of file sparse_matrix_operations.hpp.

void viennacl::linalg::cuda::detail::cuda_last_error_check	(	const char *	message,
		const char *	file,
		const int	line
	)

inline

Definition at line 165 of file common.hpp.

unsigned int viennacl::linalg::cuda::detail::get_option_for_solver_tag ( viennacl::linalg::upper_tag )

inline

Definition at line 348 of file direct_solve.hpp.

unsigned int viennacl::linalg::cuda::detail::get_option_for_solver_tag ( viennacl::linalg::unit_upper_tag )

inline

Definition at line 349 of file direct_solve.hpp.

unsigned int viennacl::linalg::cuda::detail::get_option_for_solver_tag ( viennacl::linalg::lower_tag )

inline

Definition at line 350 of file direct_solve.hpp.

unsigned int viennacl::linalg::cuda::detail::get_option_for_solver_tag ( viennacl::linalg::unit_lower_tag )

inline

Definition at line 351 of file direct_solve.hpp.

template<typename Matrix1T , typename Matrix2T , typename SolverTagT >

void viennacl::linalg::cuda::detail::inplace_solve_impl	(	Matrix1T const &	A,
		Matrix2T &	B,
		SolverTagT const &	tag
	)

Definition at line 189 of file direct_solve.hpp.

template<typename MatrixT , typename VectorT >

void viennacl::linalg::cuda::detail::inplace_solve_vector_impl	(	MatrixT const &	mat,
		VectorT &	vec,
		unsigned int	options
	)

Definition at line 354 of file direct_solve.hpp.

template<typename TagT >

bool viennacl::linalg::cuda::detail::is_unit_solve ( TagT const & tag )

Definition at line 177 of file direct_solve.hpp.

bool viennacl::linalg::cuda::detail::is_unit_solve ( viennacl::linalg::unit_lower_tag )

inline

Definition at line 179 of file direct_solve.hpp.

bool viennacl::linalg::cuda::detail::is_unit_solve ( viennacl::linalg::unit_upper_tag )

inline

Definition at line 180 of file direct_solve.hpp.

template<typename TagT >

bool viennacl::linalg::cuda::detail::is_upper_solve ( TagT const & tag )

Definition at line 183 of file direct_solve.hpp.

bool viennacl::linalg::cuda::detail::is_upper_solve ( viennacl::linalg::upper_tag )

inline

Definition at line 185 of file direct_solve.hpp.

bool viennacl::linalg::cuda::detail::is_upper_solve ( viennacl::linalg::unit_upper_tag )

inline

Definition at line 186 of file direct_solve.hpp.

template<typename NumericT >

void viennacl::linalg::cuda::detail::level_scheduling_substitute	(	vector< NumericT > &	vec,
		viennacl::backend::mem_handle const &	row_index_array,
		viennacl::backend::mem_handle const &	row_buffer,
		viennacl::backend::mem_handle const &	col_buffer,
		viennacl::backend::mem_handle const &	element_buffer,
		vcl_size_t	num_rows
	)

Definition at line 68 of file misc_operations.hpp.

template<typename NumericT >

__global__ void viennacl::linalg::cuda::detail::level_scheduling_substitute_kernel	(	const unsigned int *	row_index_array,
		const unsigned int *	row_indices,
		const unsigned int *	column_indices,
		const NumericT *	elements,
		NumericT *	vec,
		unsigned int	size
	)

Definition at line 42 of file misc_operations.hpp.

unsigned int viennacl::linalg::cuda::detail::make_options	(	vcl_size_t	length,
		bool	reciprocal,
		bool	flip_sign
	)

inline

Definition at line 160 of file common.hpp.

template<typename MatrixT1 , typename MatrixT2 , typename MatrixT3 , typename ScalarT >

void viennacl::linalg::cuda::detail::prod	(	const MatrixT1 &	A,
		bool	transposed_A,
		const MatrixT2 &	B,
		bool	transposed_B,
		MatrixT3 &	C,
		ScalarT	alpha,
		ScalarT	beta
	)

Definition at line 2346 of file matrix_operations.hpp.

template<typename MatrixT1 , typename MatrixT2 , typename MatrixT3 , typename ScalarT >

void viennacl::linalg::cuda::detail::prod_slow_kernel	(	const MatrixT1 &	A,
		bool	transposed_A,
		const MatrixT2 &	B,
		bool	transposed_B,
		MatrixT3 &	C,
		ScalarT	alpha,
		ScalarT	beta
	)

Definition at line 1569 of file matrix_operations.hpp.

template<typename NumericT , unsigned int AligmentV>

void viennacl::linalg::cuda::detail::row_info	(	compressed_matrix< NumericT, AligmentV > const &	mat,
		vector_base< NumericT > &	vec,
		viennacl::linalg::detail::row_info_types	info_selector
	)

Definition at line 107 of file sparse_matrix_operations.hpp.

template<typename NumericT , unsigned int AlignmentV>

void viennacl::linalg::cuda::detail::row_info	(	coordinate_matrix< NumericT, AlignmentV > const &	mat,
		vector_base< NumericT > &	vec,
		viennacl::linalg::detail::row_info_types	info_selector
	)

Definition at line 1222 of file sparse_matrix_operations.hpp.

template<typename NumericT >

void viennacl::linalg::cuda::detail::scan_impl	(	vector_base< NumericT > const &	input,
		vector_base< NumericT > &	output,
		bool	is_inclusive
	)

Worker routine for scan routines.

Note on performance: For non-in-place scans one could optimize away the temporary 'cuda_carries'-array. This, however, only provides small savings in the latency-dominated regime, yet would effectively double the amount of code to maintain.

Definition at line 3181 of file vector_operations.hpp.

Namespaces

Classes

Functions

Detailed Description

Function Documentation