All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Friends Macros Modules Pages
CuMatrix< Real > Class Template Reference

This class represents a matrix that's stored on the GPU if we have one, and in memory if not. More...

#include <matrix-common.h>

Inheritance diagram for CuMatrix< Real >:
Collaboration diagram for CuMatrix< Real >:

Public Member Functions

 CuMatrix ()
 
 CuMatrix (MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
 Constructor with memory initialisation. More...
 
 CuMatrix (const CuMatrix< Real > &other, MatrixTransposeType trans=kNoTrans)
 
 CuMatrix (const CuBlockMatrix< Real > &other, MatrixTransposeType trans=kNoTrans)
 
 CuMatrix (const CuMatrixBase< Real > &other, MatrixTransposeType trans=kNoTrans)
 
template<typename OtherReal >
 CuMatrix (const MatrixBase< OtherReal > &other, MatrixTransposeType trans=kNoTrans)
 
 CuMatrix (const CuSpMatrix< Real > &M)
 Copy constructor taking SpMatrix... More...
 
template<typename OtherReal >
 CuMatrix (const CuTpMatrix< OtherReal > &M, MatrixTransposeType trans=kNoTrans)
 Copy constructor taking TpMatrix... More...
 
template<typename OtherReal >
 CuMatrix (const CuMatrixBase< OtherReal > &M, MatrixTransposeType trans=kNoTrans)
 Copy constructor: as above, but from another type. More...
 
CuMatrix< Real > & operator= (const CuMatrixBase< Real > &other)
 
CuMatrix< Real > & operator= (const CuMatrix< Real > &other)
 
CuMatrix< Real > & operator= (const MatrixBase< Real > &other)
 
void Transpose ()
 
void Resize (MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
 Allocate the memory. More...
 
void Swap (Matrix< Real > *mat)
 
void Swap (CuMatrix< Real > *mat)
 
template<typename OtherReal >
void Swap (CuMatrix< OtherReal > *mat)
 
void Read (std::istream &is, bool binary)
 I/O functions. More...
 
 ~CuMatrix ()
 Destructor. More...
 
const Matrix< Real > & Mat () const
 
Matrix< Real > & Mat ()
 
void CompObjfAndDeriv (const std::vector< MatrixElement< Real > > &elements, const CuMatrix< Real > &A, Real *tot_objf, Real *tot_weight)
 Here, A is interpreted as a matrix of probabilities, and "elements" as a list of posteriors (possibly zero-one), and "*this" as a matrix of derivatives w.r.t. More...
 
- Public Member Functions inherited from CuMatrixBase< Real >
void CopyCols (const CuMatrixBase< Real > &src, const CuArray< MatrixIndexT > &indexes)
 Copies column r from column indexes[r] of src. More...
 
void AddCols (const CuMatrixBase< Real > &src, const CuArray< MatrixIndexT > &indices)
 Add column indices[r] of src to column r. More...
 
void CopyRows (const CuMatrixBase< Real > &src, const CuArray< MatrixIndexT > &indexes)
 Copies row r from row indexes[r] of src. More...
 
void CopyRows (const CuArray< const Real * > &src)
 Copies row r of this matrix from an array of floats at the location given by src[r], where src[r] is assumed to be obtained from the RowData() function of another CuMatrix, or from CuVector::Data() (the point is: the data it points to should be on the GPU if we're using a GPU, and on a CPU otherwise). More...
 
void CopyToRows (const CuArray< Real * > &dst) const
 For each row r of this matrix, copies it to the array of floats at the location given by dst[r], where dst[r] is assumed to be obtained from the RowData() function of another CuMatrix, or from CuVector::Data() (i.e. More...
 
void AddRows (Real alpha, const CuMatrixBase< Real > &src, const CuArray< MatrixIndexT > &indexes)
 Does for each row r, this.Row(r) += alpha * src.row(indexes[r]). More...
 
void AddRows (Real alpha, const CuArray< const Real * > &src)
 Does for each row r, this.Row(r) += alpha * src[r], treating src[r] as the beginning of a region of memory representing a vector of floats, of the same length as this.NumCols(). More...
 
void AddToRows (Real alpha, const CuArray< Real * > &dst) const
 For each row r of this matrix, adds it (times alpha) to the array of floats at the location given by dst[r], where dst[r] is assumed to be obtained from the RowData() function of another CuMatrix, or from CuVector::Data() (i.e. More...
 
void SumColumnRanges (const CuMatrixBase< Real > &src, const CuArray< Int32Pair > &indexes)
 For each row r of this and for each column c, sets (*this)(r, c) to the sum src(r, j), where j ranges from indexes[c].first through indexes[c].second - 1. More...
 
void AddRowRanges (const CuMatrixBase< Real > &src, const CuArray< Int32Pair > &indexes)
 For each row r of this and for each column c, do (*this)(r, c) += src(j, c), where j ranges from indexes[r].first through indexes[r].second - 1. More...
 
void AddToDiag (Real value)
 Adds "value" to the diagonal elements of the matrix. More...
 
MatrixIndexT NumRows () const
 Dimensions. More...
 
MatrixIndexT NumCols () const
 
MatrixIndexT Stride () const
 
::MatrixDim Dim () const
 
Real FrobeniusNorm () const
 
bool IsUnit (Real tol=0.001) const
 
bool ApproxEqual (const CuMatrixBase< Real > &other, float tol=0.01) const
 True if ((*this)-other).FrobeniusNorm() <= tol * this->FrobeniusNorm() More...
 
MatrixIndexT SizeInBytes () const
 Get size of matrix in bytes. More...
 
template<typename OtherReal >
void CopyFromMat (const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
 
void CopyFromGeneralMat (const GeneralMatrix &src, MatrixTransposeType trans=kNoTrans)
 
void CopyFromMat (const MatrixBase< Real > &src, MatrixTransposeType trans=kNoTrans)
 
void CopyFromSp (const CuSpMatrix< Real > &M)
 
template<typename OtherReal >
void CopyFromTp (const CuTpMatrix< OtherReal > &M, MatrixTransposeType trans=kNoTrans)
 
template<typename OtherReal >
void CopyFromMat (const CuMatrixBase< OtherReal > &M, MatrixTransposeType trans=kNoTrans)
 
template<typename OtherReal >
void CopyToMat (MatrixBase< OtherReal > *dst, MatrixTransposeType trans=kNoTrans) const
 
void CopyRowsFromVec (const CuVectorBase< Real > &v)
 This function has two modes of operation. More...
 
void CopyRowsFromVec (const VectorBase< Real > &v)
 Version of CopyRowsFromVec() that takes a CPU-based vector. More...
 
void CopyColsFromVec (const CuVectorBase< Real > &v)
 Copies vector into matrix, column-by-column. More...
 
void CopyColFromVec (const CuVectorBase< Real > &v, const MatrixIndexT col)
 Copy vector into specific column of matrix. More...
 
void Sigmoid (const CuMatrixBase< Real > &src)
 Set each element to the sigmoid of the corresponding element of "src": element by element, x = 1 / (1 + exp(-x)) More...
 
void Heaviside (const CuMatrixBase< Real > &src)
 Set each element to the Heaviside function of the corresponding element of "src", which we define as the function (x > 0 ? 1.0 : 0.0) [note: in general, there are different ways to deal with the situation when x==0. More...
 
void SoftHinge (const CuMatrixBase< Real > &src)
 Apply the function y = log(1 + exp(x)), to each element. More...
 
void GroupPnorm (const CuMatrixBase< Real > &src, Real pow)
 Apply the function y(i) = (sum_{j = i*G}^{(i+1)*G-1} x_j ^ (power)) ^ (1 / p) where G = x.NumCols() / y.NumCols() must be an integer. More...
 
void DiffGroupPnorm (const CuMatrixBase< Real > &in_value, const CuMatrixBase< Real > &out_value, const CuMatrixBase< Real > &out_deriv, Real power)
 Differentiate backward through the GroupPnorm function. More...
 
void GroupMax (const CuMatrixBase< Real > &src)
 Apply the function y(i) = (max_{j = i*G}^{(i+1)*G-1} x_j where G = x.NumCols() / y.NumCols() must be an integer. More...
 
void GroupMaxDeriv (const CuMatrixBase< Real > &input, const CuMatrixBase< Real > &output)
 Calculate derivatives for the GroupMax function above, where "input" is the input to the GroupMax function above (i.e. More...
 
void ParametricRelu (const CuMatrixBase< Real > &src, const CuVectorBase< Real > &alpha, const CuVectorBase< Real > &beta)
 Compute the parametric rectified linear unit function; element by element, *this = src * (src > 0 ? alpha : beta) More...
 
void DiffParametricRelu (const CuMatrixBase< Real > &value, const CuMatrixBase< Real > &diff, const CuVectorBase< Real > &alpha, const CuVectorBase< Real > &beta)
 Differentiate backward through the parametric relu function. More...
 
void Tanh (const CuMatrixBase< Real > &src)
 Compute the hyperbolic tangent (tanh) function; element by element, *this = tanh(src). More...
 
void DiffSigmoid (const CuMatrixBase< Real > &value, const CuMatrixBase< Real > &diff)
 Differentiate backward through the sigmoid function. More...
 
void DiffTanh (const CuMatrixBase< Real > &value, const CuMatrixBase< Real > &diff)
 Differentiate backward through the tanh function. More...
 
void DiffSoftmaxPerRow (const CuMatrixBase< Real > &value, const CuMatrixBase< Real > &diff)
 Differentiate backward through the softmax function. More...
 
void DiffLogSoftmaxPerRow (const CuMatrixBase< Real > &out_value, const CuMatrixBase< Real > &out_deriv)
 Differentiate backward through the log softmax function. More...
 
void DiffXent (const CuArray< int32 > &tgt, CuVector< Real > *log_post_tgt)
 Differentiate the block [softmax+cross-entropy] : dE/da = posterior_mat - target_mat, 'E' is error function, 'a' is activation on softmax input. More...
 
void Cholesky (CuMatrixBase< Real > *inv_cholesky=NULL)
 This function does sets *this to the Cholesky factor of *this (i.e. More...
 
void SymInvertPosDef ()
 Inversion for positive definite symmetric matrices. More...
 
void ApplyPow (Real power)
 Apply power to the absolute value of each element. More...
 
void ApplyPowAbs (Real power, bool include_sign=false)
 
void ApplyHeaviside ()
 For each element, sets x = (x > 0 ? 1.0 : 0.0). More...
 
void ApplyFloor (Real floor_val)
 
void ApplyCeiling (Real ceiling_val)
 
void ApplyExp ()
 
void ApplySoftMaxPerRow (const CuMatrixBase< Real > &src)
 Softmax nonlinearity Y = Softmax(X) : Yij = e^Xij / sum_k(e^Xik), done to each row for each row, the max value is first subtracted for good numerical stability. More...
 
void ApplyLogSoftMaxPerRow (const CuMatrixBase< Real > &src)
 LogSoftmax nonlinearity Y = LogSoftmax(X) : Yij = Xij - log(sum_k(e^Xik)), done to each row for each row, the max value is first subtracted for good numerical stability. More...
 
void FindRowMaxId (CuArray< int32 > *id) const
 Find the id of the maximal element for each row. More...
 
void SetZero ()
 Math operations, some calling kernels. More...
 
void Set (Real value)
 
void Add (Real value)
 
void SetZeroAboveDiag ()
 Zeroes all elements for which col > row. More...
 
void Scale (Real value)
 
void ApplyLog ()
 
void MulElements (const CuMatrixBase< Real > &A)
 Multiply two matrices elementwise: C = C .* A. More...
 
void DivElements (const CuMatrixBase< Real > &A)
 Divide two matrices elementwise: C = A ./ A. More...
 
void Max (const CuMatrixBase< Real > &A)
 Do, elementwise, *this = max(*this, A). More...
 
void Min (const CuMatrixBase< Real > &A)
 Do, elementwise, *this = min(*this, A). More...
 
void MulColsVec (const CuVectorBase< Real > &scale)
 scale i'th column by scale[i] More...
 
void MulRowsVec (const CuVectorBase< Real > &scale)
 scale i'th row by scale[i] More...
 
void MulRowsGroupMat (const CuMatrixBase< Real > &src)
 divide each row into src.NumCols() groups, and then scale i'th row's jth group of elements by src[i, j]. More...
 
void DivRowsVec (const CuVectorBase< Real > &div)
 divide i'th row by scale[i] More...
 
void InvertElements ()
 invert the matrix by elements. More...
 
void AddMat (Real alpha, const CuMatrixBase< Real > &A, MatrixTransposeType trans=kNoTrans)
 *this += alpha * A More...
 
void AddMatBlocks (Real alpha, const CuMatrixBase< Real > &A, MatrixTransposeType trans=kNoTrans)
 This function is like AddMat (it does *this += alpha * src), except that it supports cases where *this and src have different dimension. More...
 
void AddVecToCols (Real alpha, const CuVectorBase< Real > &col, Real beta=1.0)
 (for each column c of *this), c = alpha * col + beta * c More...
 
void AddVecToRows (Real alpha, const CuVectorBase< Real > &row, Real beta=1.0)
 (for each row r of *this), r = alpha * row + beta * r More...
 
void AddMatMat (Real alpha, const CuMatrixBase< Real > &A, MatrixTransposeType transA, const CuMatrixBase< Real > &B, MatrixTransposeType transB, Real beta)
 C = alpha * A(^T)*B(^T) + beta * C. More...
 
void AddVecVec (Real alpha, const CuVectorBase< Real > &x, const CuVectorBase< Real > &y)
 A = alpha * x * y^T + A . More...
 
void SetMatMatDivMat (const CuMatrixBase< Real > &A, const CuMatrixBase< Real > &B, const CuMatrixBase< Real > &C)
 *this = a * b / c (by element; when c = 0, *this = a) *this can be an alias of a, b or c safely and get expected result. More...
 
void SymAddMat2 (const Real alpha, const CuMatrixBase< Real > &M, MatrixTransposeType transA, Real beta)
 *this = beta * *this + alpha * M M^T, for symmetric matrices. More...
 
void AddMatBlock (Real alpha, const CuMatrixBase< Real > &A, MatrixTransposeType transA, const CuBlockMatrix< Real > &B, MatrixTransposeType transB, Real beta)
 This function is like AddMatMat but for where the second argument is of type CuBlockMatrix (a block-diagonal matrix of blocks). More...
 
void AddDiagVecMat (const Real alpha, const CuVectorBase< Real > &v, const CuMatrixBase< Real > &M, MatrixTransposeType transM, Real beta=1.0)
 *this = beta * *this + alpha * diag(v) * M [or M^T]. More...
 
void AddMatDiagVec (const Real alpha, const CuMatrixBase< Real > &M, MatrixTransposeType transM, CuVectorBase< Real > &v, Real beta=1.0)
 
void AddMatMatElements (const Real alpha, const CuMatrixBase< Real > &A, const CuMatrixBase< Real > &B, const Real beta)
 *this = beta * *this + alpha * A .* B (.* element by element multiplication) More...
 
void AddMatSp (const Real alpha, const CuMatrixBase< Real > &A, MatrixTransposeType transA, const CuSpMatrix< Real > &B, const Real beta)
 this <– beta*this + alpha*A*B More...
 
void AddSpMat (const Real alpha, const CuSpMatrix< Real > &A, const CuMatrixBase< Real > &B, MatrixTransposeType transB, const Real beta)
 this <– beta*this + alpha*SpA*B More...
 
void AddTpMat (const Real alpha, const CuTpMatrix< Real > &A, MatrixTransposeType transA, const CuMatrixBase< Real > &B, MatrixTransposeType transB, const Real beta)
 this <– beta*this + alpha*A*B. More...
 
void AddMatTp (const Real alpha, const CuMatrixBase< Real > &A, MatrixTransposeType transA, const CuTpMatrix< Real > &B, MatrixTransposeType transB, const Real beta)
 this <– beta*this + alpha*A*B. More...
 
void CopyFromBlock (const CuBlockMatrix< Real > &B, MatrixTransposeType trans=kNoTrans)
 
void CopyLowerToUpper ()
 
void CopyUpperToLower ()
 
CuSubMatrix< Real > Range (const MatrixIndexT row_offset, const MatrixIndexT num_rows, const MatrixIndexT col_offset, const MatrixIndexT num_cols) const
 
CuSubMatrix< Real > RowRange (const MatrixIndexT row_offset, const MatrixIndexT num_rows) const
 
CuSubMatrix< Real > ColRange (const MatrixIndexT col_offset, const MatrixIndexT num_cols) const
 
const CuSubVector< Real > Row (MatrixIndexT i) const
 
CuSubVector< Real > Row (MatrixIndexT i)
 
CuValue< Real > operator() (MatrixIndexT r, MatrixIndexT c)
 
Real operator() (MatrixIndexT r, MatrixIndexT c) const
 
Real Sum () const
 
Real Max () const
 
Real Min () const
 
Real Trace (bool check_square=true) const
 Return the trace. If check_square = true, will crash if matrix is not square. More...
 
void SetRandn ()
 
void SetRandUniform ()
 
void Write (std::ostream &os, bool binary) const
 
void AddElements (Real alpha, const std::vector< MatrixElement< Real > > &input)
 
void AddElements (Real alpha, const CuArray< Int32Pair > &indexes, const Real *input)
 
void Lookup (const std::vector< Int32Pair > &indexes, Real *output) const
 
void Lookup (const CuArray< Int32Pair > &indexes, Real *output) const
 
void EqualElementMask (const CuMatrixBase< Real > &mat, CuMatrix< Real > *mask) const
 
const Real * RowData (MatrixIndexT r) const
 Get raw row pointer (const). More...
 
Real * RowData (MatrixIndexT r)
 Get raw row pointer. More...
 
const Real * Data () const
 Return data pointer (const). More...
 
Real * Data ()
 Return data pointer. More...
 
const MatrixBase< Real > & Mat () const
 
MatrixBase< Real > & Mat ()
 

Private Member Functions

void Destroy ()
 

Additional Inherited Members

- Protected Member Functions inherited from CuMatrixBase< Real >
 CuMatrixBase ()
 
 CuMatrixBase (Real *data, MatrixIndexT num_rows, MatrixIndexT num_cols, MatrixIndexT stride)
 This constructor takes the #rows, #cols and stride; it's called from the constructor of CuSubMatrix. More...
 
- Protected Attributes inherited from CuMatrixBase< Real >
Real * data_
 GPU data pointer (or regular matrix data pointer,. More...
 
MatrixIndexT num_cols_
 
MatrixIndexT num_rows_
 
MatrixIndexT stride_
 

Detailed Description

template<typename Real>
class kaldi::CuMatrix< Real >

This class represents a matrix that's stored on the GPU if we have one, and in memory if not.

Definition at line 69 of file matrix-common.h.

Constructor & Destructor Documentation

CuMatrix ( )
inline

Definition at line 674 of file cu-matrix.h.

674 { }
CuMatrix ( MatrixIndexT  rows,
MatrixIndexT  cols,
MatrixResizeType  resize_type = kSetZero,
MatrixStrideType  stride_type = kDefaultStride 
)
inline

Constructor with memory initialisation.

Definition at line 677 of file cu-matrix.h.

679  {
680  Resize(rows, cols, resize_type, stride_type);
681  }
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
CuMatrix ( const CuMatrix< Real > &  other,
MatrixTransposeType  trans = kNoTrans 
)

Definition at line 373 of file cu-matrix.cc.

References kaldi::kNoTrans, kaldi::kUndefined, CuMatrixBase< Real >::NumCols(), and CuMatrixBase< Real >::NumRows().

373  {
374  if (trans == kNoTrans)
375  this->Resize(other.NumRows(), other.NumCols(), kUndefined);
376  else
377  this->Resize(other.NumCols(), other.NumRows(), kUndefined);
378  this->CopyFromMat(other, trans);
379 }
void CopyFromMat(const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:337
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
CuMatrix ( const CuBlockMatrix< Real > &  other,
MatrixTransposeType  trans = kNoTrans 
)
explicit

Definition at line 192 of file cu-matrix.cc.

References CuMatrixBase< Real >::CopyFromBlock(), kaldi::kNoTrans, kaldi::kTrans, kaldi::kUndefined, CuBlockMatrix< Real >::NumCols(), CuBlockMatrix< Real >::NumRows(), and CuMatrix< Real >::Resize().

193  : CuMatrixBase<Real>() {
194  if (trans == kNoTrans) {
195  Resize(B.NumRows(), B.NumCols(), kUndefined);
196  this->CopyFromBlock(B);
197  } else {
198  Resize(B.NumCols(), B.NumRows(), kUndefined);
199  this->CopyFromBlock(B, kTrans);
200  }
201 }
void CopyFromBlock(const CuBlockMatrix< Real > &B, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:158
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
CuMatrix ( const CuMatrixBase< Real > &  other,
MatrixTransposeType  trans = kNoTrans 
)
explicit

Definition at line 382 of file cu-matrix.cc.

References kaldi::kNoTrans, kaldi::kUndefined, CuMatrixBase< Real >::NumCols(), and CuMatrixBase< Real >::NumRows().

382  {
383  if (trans == kNoTrans)
384  this->Resize(other.NumRows(), other.NumCols(), kUndefined);
385  else
386  this->Resize(other.NumCols(), other.NumRows(), kUndefined);
387  this->CopyFromMat(other, trans);
388 }
void CopyFromMat(const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:337
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
CuMatrix ( const MatrixBase< OtherReal > &  other,
MatrixTransposeType  trans = kNoTrans 
)
explicit

Definition at line 393 of file cu-matrix.cc.

References kaldi::kNoTrans, kaldi::kUndefined, MatrixBase< Real >::NumCols(), and MatrixBase< Real >::NumRows().

393  {
394  if (trans == kNoTrans)
395  this->Resize(other.NumRows(), other.NumCols(), kUndefined);
396  else
397  this->Resize(other.NumCols(), other.NumRows(), kUndefined);
398  this->CopyFromMat(other, trans);
399 }
void CopyFromMat(const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:337
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
CuMatrix ( const CuSpMatrix< Real > &  M)
inlineexplicit

Copy constructor taking SpMatrix...

Definition at line 699 of file cu-matrix.h.

699  : CuMatrixBase<Real>() {
700  Resize(M.NumRows(), M.NumRows(), kUndefined);
701  this->CopyFromSp(M);
702  }
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
void CopyFromSp(const CuSpMatrix< Real > &M)
Definition: cu-matrix.cc:353
CuMatrix ( const CuTpMatrix< OtherReal > &  M,
MatrixTransposeType  trans = kNoTrans 
)
inlineexplicit

Copy constructor taking TpMatrix...

Definition at line 706 of file cu-matrix.h.

707  : CuMatrixBase<Real>() {
708  Resize(M.NumCols(), M.NumRows(), kUndefined);
709  this->CopyFromTp(M, trans);
710  }
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
void CopyFromTp(const CuTpMatrix< OtherReal > &M, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:275
CuMatrix ( const CuMatrixBase< OtherReal > &  M,
MatrixTransposeType  trans = kNoTrans 
)
explicit

Copy constructor: as above, but from another type.

Copy constructor from another type.

Definition at line 2878 of file cu-matrix.cc.

References CuMatrixBase< Real >::CopyFromMat(), kaldi::kNoTrans, kaldi::kTrans, CuMatrixBase< Real >::NumCols(), CuMatrixBase< Real >::NumRows(), and CuMatrix< Real >::Resize().

2879  : CuMatrixBase<Real>() {
2880 
2881  if (trans == kNoTrans) {
2882  Resize(M.NumRows(), M.NumCols());
2883  this->CopyFromMat(M);
2884  } else {
2885  Resize(M.NumCols(), M.NumRows());
2886  this->CopyFromMat(M, kTrans);
2887  }
2888 }
void CopyFromMat(const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:337
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
~CuMatrix ( )
inline

Destructor.

Definition at line 752 of file cu-matrix.h.

752 { Destroy(); }
void Destroy()
Definition: cu-matrix.cc:91

Member Function Documentation

void CompObjfAndDeriv ( const std::vector< MatrixElement< Real > > &  elements,
const CuMatrix< Real > &  A,
Real *  tot_objf,
Real *  tot_weight 
)

Here, A is interpreted as a matrix of probabilities, and "elements" as a list of posteriors (possibly zero-one), and "*this" as a matrix of derivatives w.r.t.

the log-probs. This function does: for each element { row, column, weight } indexed i in the vector "elements", let x(i) = A(row(i), column(i)); then it does (*this)(row(i), column(i)) += weight(i) / x(i), and *tot_objf = weight(i) * log(x(i)), and *tot_weight = weight(i) Preconditions: A must be strictly positive, and no (row, column) pair may be repeated within "elements"

Definition at line 1510 of file cu-matrix.cc.

References CU1DBLOCK, CuVectorBase< Real >::Data(), CuMatrixBase< Real >::Data(), CuMatrixBase< Real >::Dim(), rnnlm::i, KALDI_ASSERT, KALDI_WARN, kaldi::kUndefined, and kaldi::Log().

Referenced by NnetUpdater::ComputeObjfAndDeriv(), NnetDiscriminativeUpdater::LatticeComputations(), kaldi::TestCuMatrixCompObjfAndDeriv(), and kaldi::UnitTestCuMatrixObjfDeriv().

1512  {
1513  { // check the input.
1514  typedef typename std::vector<MatrixElement<Real> >::const_iterator Iter;
1515  MatrixIndexT num_rows = this->num_rows_, num_cols = this->num_cols_;
1516  for (Iter iter = sv_labels.begin(); iter != sv_labels.end(); ++iter) {
1517  KALDI_ASSERT(iter->row < num_rows && iter->row >= 0 &&
1518  iter->column < num_cols && iter->column >= 0);
1519  }
1520  }
1521 
1522 #if HAVE_CUDA == 1
1523  if (CuDevice::Instantiate().Enabled()) {
1524  if (sv_labels.empty()) {
1525  KALDI_WARN << "Empty supervision labels";
1526  *tot_objf = 0.0;
1527  *tot_weight = 0.0;
1528  return;
1529  }
1530  void *addr = CuDevice::Instantiate().Malloc(sv_labels.size() * sizeof(MatrixElement<Real>));
1531  CU_SAFE_CALL(cudaMemcpy(addr, sv_labels.data(), sv_labels.size() * sizeof(MatrixElement<Real>), cudaMemcpyHostToDevice));
1532  CuTimer tim;
1533  CuVector<Real> tmp(2, kUndefined);
1534  int dimBlock(CU1DBLOCK);
1535  int dimGrid = 1; // only 1 block here. we have loops in each thread.
1536  cuda_comp_obj_deriv(dimGrid, dimBlock, (MatrixElement<Real>*)addr,
1537  sv_labels.size(), output.Data(), output.Dim(),
1538  this->Data(), this->Dim(), tmp.Data());
1539  Vector<Real> tmp_cpu(tmp);
1540  *tot_objf = tmp_cpu(0);
1541  *tot_weight = tmp_cpu(1);
1542  CuDevice::Instantiate().Free(addr);
1543  CuDevice::Instantiate().AccuProfile(__func__, tim);
1544  } else
1545 #endif
1546  {
1547  *tot_objf = 0.0;
1548  *tot_weight = 0.0;
1549  for(int32 i = 0; i<sv_labels.size(); i++) {
1550  int32 m = sv_labels[i].row, label = sv_labels[i].column;
1551  Real weight = sv_labels[i].weight;
1552  //KALDI_ASSERT(label >= 0 && label < nnet_.OutputDim());
1553  Real this_prob = output(m, label);
1554  KALDI_ASSERT(this_prob >= 0.99e-20); // we floored to 1.0e-20 in SoftmaxLayer.
1555  *tot_objf += weight * Log(this_prob);
1556  *tot_weight += weight;
1557  (*this)(m, label) += weight / this_prob;
1558  }
1559  }
1560 }
MatrixIndexT num_cols_
Definition: cu-matrix.h:660
int32 MatrixIndexT
Definition: matrix-common.h:96
double Log(double x)
Definition: kaldi-math.h:100
#define CU1DBLOCK
Definition: cu-matrixdim.h:63
#define KALDI_WARN
Definition: kaldi-error.h:130
::MatrixDim Dim() const
Definition: cu-matrix.h:201
#define KALDI_ASSERT(cond)
Definition: kaldi-error.h:169
MatrixIndexT num_rows_
Definition: cu-matrix.h:661
const Real * Data() const
Return data pointer (const).
Definition: cu-matrix.h:621
void Destroy ( )
private

Definition at line 91 of file cu-matrix.cc.

References data_, and KALDI_MEMALIGN_FREE.

Referenced by CuMatrix< BaseFloat >::~CuMatrix().

91  {
92 #if HAVE_CUDA == 1
93  if (CuDevice::Instantiate().Enabled()) {
94  if (this->data_ != NULL) {
95  CuTimer tim;
96  CuDevice::Instantiate().Free(this->data_);
97  CuDevice::Instantiate().AccuProfile(__func__, tim);
98  }
99  } else
100 #endif
101  {
102  if (this->data_ != NULL) KALDI_MEMALIGN_FREE(this->data_);
103  }
104  this->data_ = NULL;
105  this->num_rows_ = 0;
106  this->num_cols_ = 0;
107  this->stride_ = 0;
108 }
MatrixIndexT num_cols_
Definition: cu-matrix.h:660
MatrixIndexT stride_
Definition: cu-matrix.h:662
Real * data_
GPU data pointer (or regular matrix data pointer,.
Definition: cu-matrix.h:652
#define KALDI_MEMALIGN_FREE(x)
Definition: kaldi-utils.h:56
MatrixIndexT num_rows_
Definition: cu-matrix.h:661
const Matrix<Real>& Mat ( ) const
inline

Definition at line 754 of file cu-matrix.h.

Referenced by CuVectorBase< Real >::CopyDiagFromMat(), CuRand< Real >::RandGaussian(), and CuRand< Real >::RandUniform().

754  {
755  return *(reinterpret_cast<const Matrix<Real>* >(this));
756  }
Matrix<Real>& Mat ( )
inline

Definition at line 757 of file cu-matrix.h.

757  {
758  return *(reinterpret_cast<Matrix<Real>* >(this));
759  }
CuMatrix<Real>& operator= ( const CuMatrixBase< Real > &  other)
inline

Definition at line 717 of file cu-matrix.h.

717  {
718  this->Resize(other.NumRows(), other.NumCols(), kUndefined);
719  this->CopyFromMat(other);
720  return *this;
721  }
void CopyFromMat(const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:337
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
CuMatrix<Real>& operator= ( const CuMatrix< Real > &  other)
inline

Definition at line 723 of file cu-matrix.h.

723  {
724  this->Resize(other.NumRows(), other.NumCols(), kUndefined);
725  this->CopyFromMat(other);
726  return *this;
727  }
void CopyFromMat(const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:337
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
CuMatrix<Real>& operator= ( const MatrixBase< Real > &  other)
inline

Definition at line 729 of file cu-matrix.h.

729  {
730  this->Resize(other.NumRows(), other.NumCols(), kUndefined);
731  this->CopyFromMat(other);
732  return *this;
733  }
void CopyFromMat(const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:337
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
void Resize ( MatrixIndexT  rows,
MatrixIndexT  cols,
MatrixResizeType  resize_type = kSetZero,
MatrixStrideType  stride_type = kDefaultStride 
)

Allocate the memory.

Definition at line 47 of file cu-matrix.cc.

References data_, KALDI_ASSERT, kaldi::kDefaultStride, kaldi::kSetZero, and kaldi::kUndefined.

Referenced by NnetComputer::AcceptInput(), MatrixRandomizer::AddData(), MaxoutComponent::Backprop(), MaxpoolingComponent::Backprop(), PnormComponent::Backprop(), NormalizeComponent::Backprop(), SigmoidComponent::Backprop(), TanhComponent::Backprop(), PowerComponent::Backprop(), RectifiedLinearComponent::Backprop(), SoftHingeComponent::Backprop(), ScaleComponent::Backprop(), SoftmaxComponent::Backprop(), LogSoftmaxComponent::Backprop(), AffineComponent::Backprop(), SpliceComponent::Backprop(), SpliceMaxComponent::Backprop(), BlockAffineComponent::Backprop(), SumGroupComponent::Backprop(), PermuteComponent::Backprop(), DctComponent::Backprop(), FixedLinearComponent::Backprop(), FixedAffineComponent::Backprop(), DropoutComponent::Backprop(), Convolutional1dComponent::Backprop(), Component::Backpropagate(), LstmProjected::BackpropagateFnc(), BlstmProjected::BackpropagateFnc(), AffineComponent::CollapseWithNext(), AffineComponent::CollapseWithPrevious(), NnetOnlineComputer::Compute(), DecodableAmNnetParallel::Compute(), NnetComputer::ComputeLastLayerDeriv(), NnetUpdater::ComputeObjfAndDeriv(), CuMatrix< BaseFloat >::CuMatrix(), CuMatrix< Real >::CuMatrix(), DecodableNnetSimple::DoNnetComputation(), CuMatrixBase< Real >::EqualElementMask(), MultiTaskLoss::Eval(), NnetOnlineComputer::Flush(), NnetRescaler::FormatInput(), RepeatedAffineComponent::Init(), NaturalGradientAffineComponent::Init(), AffineComponent::Init(), AffineComponentPreconditioned::Init(), AffineComponentPreconditionedOnline::Init(), BlockAffineComponent::Init(), DctComponent::Init(), Convolutional1dComponent::Init(), ConvolutionComponent::Init(), LstmNonlinearityComponent::Init(), AffineTransform::InitData(), LinearTransform::InitData(), RecurrentComponent::InitData(), LstmProjected::InitData(), BlstmProjected::InitData(), ConvolutionalComponent::InitData(), Convolutional2DComponent::InitData(), Rbm::InitData(), OnlineNaturalGradient::InitDefault(), OnlinePreconditioner::InitDefault(), TimeHeightConvolutionComponent::InitFromConfig(), NnetDiscriminativeUpdater::LatticeComputations(), main(), NnetComputer::NnetComputer(), NnetLogprobTask::operator()(), CuMatrix< BaseFloat >::operator=(), NnetDiscriminativeComputeObjf::ProcessOutputs(), NnetChainTrainer::ProcessOutputs(), NnetChainComputeProb::ProcessOutputs(), NnetDiscriminativeTrainer::ProcessOutputs(), NnetOnlineComputer::Propagate(), Component::Propagate(), BatchNormComponent::Propagate(), KlHmm::PropagateFnc(), ConvolutionalComponent::PropagateFnc(), Dropout::PropagateFnc(), LstmProjected::PropagateFnc(), BlstmProjected::PropagateFnc(), Rbm::RbmUpdate(), Rbm::Reconstruct(), LstmProjected::ResetStreams(), NaturalGradientAffineComponent::Resize(), AffineComponent::Resize(), AffineComponentPreconditionedOnline::Resize(), Convolutional1dComponent::Resize(), kaldi::UnitTestCuSparseMatrixSwap(), kaldi::nnet3::UnitTestNnetInputDerivatives(), kaldi::nnet3::UnitTestNnetModelDerivatives(), RecurrentComponent::Update(), ConvolutionalComponent::Update(), Convolutional2DComponent::Update(), NaturalGradientAffineComponent::Update(), AffineComponentPreconditioned::Update(), AffineComponentPreconditionedOnline::Update(), ConvolutionComponent::Update(), Convolutional1dComponent::Update(), MatrixRandomizer::Value(), and AffineComponent::Widen().

49  {
50  // This code does not currently support the other resize_type options.
51  KALDI_ASSERT(resize_type == kSetZero || resize_type == kUndefined);
52  if (rows * cols == 0) KALDI_ASSERT(rows == 0 && cols == 0);
53  if (this->num_rows_ == rows && this->num_cols_ == cols) {
54  if (resize_type == kSetZero) this->SetZero();
55  return;
56  }
57  if (this->num_rows_ != 0)
58  this->Destroy();
59  if (rows == 0) return;
60 #if HAVE_CUDA == 1
61  if (CuDevice::Instantiate().Enabled()) {
62  CuTimer tim;
63  MatrixIndexT row_bytes = cols * sizeof(Real);
64  size_t pitch;
65  if (stride_type == kDefaultStride) {
66  this->data_ = static_cast<Real*>(CuDevice::Instantiate().MallocPitch(
67  row_bytes, rows, &pitch));
68  this->num_rows_ = rows;
69  this->num_cols_ = cols;
70  this->stride_ = pitch / sizeof(Real);
71  } else { // kStrideEqualNumCols
72  size_t bytes = rows * cols * sizeof(Real);
73  this->data_ = static_cast<Real*>(CuDevice::Instantiate().Malloc(bytes));
74  this->num_rows_ = rows;
75  this->num_cols_ = cols;
76  this->stride_ = cols;
77  }
78  if (resize_type == kSetZero) this->SetZero();
79  CuDevice::Instantiate().AccuProfile("CuMatrix::Resize", tim);
80  } else
81 #endif
82  { // Let the initializer of Matrix<Real> handle the allocation,
83  // and then just do Swap which will switch the pointers.
84  // This wastes a few instructions but is simple to code.
85  Matrix<Real> mat(rows, cols, resize_type, stride_type);
86  this->Swap(&mat);
87  }
88 }
void Swap(Matrix< Real > *mat)
Definition: cu-matrix.cc:120
MatrixIndexT num_cols_
Definition: cu-matrix.h:660
int32 MatrixIndexT
Definition: matrix-common.h:96
MatrixIndexT stride_
Definition: cu-matrix.h:662
Real * data_
GPU data pointer (or regular matrix data pointer,.
Definition: cu-matrix.h:652
void SetZero()
Math operations, some calling kernels.
Definition: cu-matrix.cc:474
#define KALDI_ASSERT(cond)
Definition: kaldi-error.h:169
MatrixIndexT num_rows_
Definition: cu-matrix.h:661
void Destroy()
Definition: cu-matrix.cc:91
void Swap ( Matrix< Real > *  mat)

Definition at line 120 of file cu-matrix.cc.

References MatrixBase< Real >::data_, kaldi::kUndefined, MatrixBase< Real >::num_cols_, MatrixBase< Real >::num_rows_, Matrix< Real >::Resize(), MatrixBase< Real >::stride_, kaldi::swap(), and Matrix< Real >::Swap().

Referenced by DecodableNnetLoopedOnlineBase::AdvanceChunk(), NnetDiscriminativeUpdater::Backprop(), DecodableNnet2Online::ComputeForFrame(), kaldi::nnet3::ComputeObjectiveFunction(), GeneralMatrix::CopyToMat(), DecodableNnetSimple::DoNnetComputation(), Nnet::Feedforward(), OnlineNaturalGradient::Init(), OnlinePreconditioner::Init(), main(), OnlineNaturalGradient::PreconditionDirectionsInternal(), OnlinePreconditioner::PreconditionDirectionsInternal(), NnetRescaler::Rescale(), Matrix< Real >::Swap(), kaldi::UnitTestSwapCu2Cu(), and kaldi::UnitTestSwapCu2M().

120  {
121 #if HAVE_CUDA == 1
122  if (CuDevice::Instantiate().Enabled()) {
123  if (this->num_rows_ == 0) {
124  if (mat->num_rows_ != 0) {
125  // *this is empty, but mat is nonempty.
126  this->Resize(mat->num_rows_, mat->num_cols_, kUndefined);
127  this->CopyFromMat(*mat);
128  mat->Resize(0, 0);
129  }
130  // else both are empty.
131  } else { // *this is nonempty.
132  if (mat->num_rows_ != 0) {
133  // Both *this and *mat are nonempty. Recurse to simpler cases.
134  // this could be done more efficiently in the case where
135  // the size does not change.
136  Matrix<Real> temp;
137  this->Swap(&temp); // now temp is full, *this is empty.
138  mat->Swap(&temp); // now mat has data from *this, temp has
139  // data from mat.
140  this->Swap(&temp); // copy data in mat to *this, which is now empty.
141  } else { // *this is full but *mat is empty.
142  mat->Resize(this->num_rows_, this->num_cols_, kUndefined);
143  this->CopyToMat(mat);
144  this->Destroy();
145  }
146  }
147  } else
148 #endif
149  {
150  std::swap(mat->data_, this->data_);
151  std::swap(mat->num_cols_, this->num_cols_);
152  std::swap(mat->num_rows_, this->num_rows_);
153  std::swap(mat->stride_, this->stride_);
154  }
155 }
void Swap(Matrix< Real > *mat)
Definition: cu-matrix.cc:120
void swap(basic_filebuf< CharT, Traits > &x, basic_filebuf< CharT, Traits > &y)
void CopyFromMat(const MatrixBase< OtherReal > &src, MatrixTransposeType trans=kNoTrans)
Definition: cu-matrix.cc:337
MatrixIndexT num_cols_
Definition: cu-matrix.h:660
void Resize(MatrixIndexT rows, MatrixIndexT cols, MatrixResizeType resize_type=kSetZero, MatrixStrideType stride_type=kDefaultStride)
Allocate the memory.
Definition: cu-matrix.cc:47
void CopyToMat(MatrixBase< OtherReal > *dst, MatrixTransposeType trans=kNoTrans) const
Definition: cu-matrix.cc:413
MatrixIndexT num_rows_
Definition: cu-matrix.h:661
void Destroy()
Definition: cu-matrix.cc:91
void Swap ( CuMatrix< Real > *  mat)

Definition at line 111 of file cu-matrix.cc.

References CuMatrixBase< Real >::data_, CuMatrixBase< Real >::num_cols_, CuMatrixBase< Real >::num_rows_, CuMatrixBase< Real >::stride_, and kaldi::swap().

111  {
112  std::swap(mat->data_, this->data_);
113  std::swap(mat->num_cols_, this->num_cols_);
114  std::swap(mat->num_rows_, this->num_rows_);
115  std::swap(mat->stride_, this->stride_);
116 }
void swap(basic_filebuf< CharT, Traits > &x, basic_filebuf< CharT, Traits > &y)
void Swap ( CuMatrix< OtherReal > *  mat)
void Transpose ( )

Definition at line 2900 of file cu-matrix.cc.

References kaldi::kTrans.

Referenced by MultiBasisComponent::BackpropagateFnc(), MultiBasisComponent::PropagateFnc(), kaldi::TestCuMatrixTransposeNS(), kaldi::TestCuMatrixTransposeS(), kaldi::UnitTestCuMatrixAddDiagVecMat(), kaldi::UnitTestCuMatrixSymAddMat2(), and kaldi::UnitTestCuMatrixSymInvertPosDef().

2900  {
2901  if (this->num_rows_ == 0)
2902  return;
2903  // Copy and swap for all cases.
2904  // No need for a separate kernel of squared matrix in-place transpose.
2905  // It has the same possible peak performance as copy_transpose,
2906  // if allocate/deallocate overhead can be ignored.
2907  CuMatrix<Real> tmp(*this, kTrans);
2908  this->Swap(&tmp);
2909 }
void Swap(Matrix< Real > *mat)
Definition: cu-matrix.cc:120
MatrixIndexT num_rows_
Definition: cu-matrix.h:661

The documentation for this class was generated from the following files: