From 81078bd69f824c6655bf8d8d30c22b71edaa1ca1 Mon Sep 17 00:00:00 2001
From: kradchen <chenjuijun@realcan.cn>
Date: Tue, 19 Dec 2023 13:12:20 +0800
Subject: [PATCH] Add std, and fix fft, ifft bug for cuda

---
 cmake/AuroraConfig.cmake      |   9 +-
 src/Function2D.cu             |  39 +-
 src/Function2D.cuh            |   8 +
 test/CudaMatrix_Test.cpp      | 727 ++++++++++++++++++++--------------
 test/Function2D_Cuda_Test.cpp |  82 +++-
 5 files changed, 552 insertions(+), 313 deletions(-)

diff --git a/cmake/AuroraConfig.cmake b/cmake/AuroraConfig.cmake
index 6bcbe4a..41ef2bd 100644
--- a/cmake/AuroraConfig.cmake
+++ b/cmake/AuroraConfig.cmake
@@ -1,6 +1,8 @@
 set(MKL_INTERFACE_FULL intel_lp64)
 find_package(OpenMP REQUIRED)
 find_package(MKL CONFIG REQUIRED)
+enable_language(CUDA)
+find_package(CUDAToolkit REQUIRED)
 
 set(Aurora_MAJOR_VERSION 1)
 set(Aurora_MINOR_VERSION 0)
@@ -9,12 +11,11 @@ set(Aurora_BUILD_VERSION 0)
 get_filename_component(Aurora_DIR "${CMAKE_CURRENT_LIST_DIR}/" PATH)
 
 message("Aurora_DIR: ${Aurora_DIR}")
-file(GLOB_RECURSE Aurora_Source "${Aurora_DIR}/src/*.cpp")
-
+file(GLOB_RECURSE Aurora_Source "${Aurora_DIR}/src/[AFSC]*.cpp" "${Aurora_DIR}/src/Matrix*.cpp" "${Aurora_DIR}/src/*.cu")
+message( ${Aurora_Source})
 set(Aurora_INCLUDE_DIRS "${Aurora_DIR}/src" "${Aurora_DIR}/thirdparty/include" $<TARGET_PROPERTY:MKL::MKL,INTERFACE_INCLUDE_DIRECTORIES>)
 
-
 set(Aurora_Complie_Options $<TARGET_PROPERTY:MKL::MKL,INTERFACE_COMPILE_OPTIONS> )
-set(Aurora_Libraries $<LINK_ONLY:MKL::MKL> OpenMP::OpenMP_CXX)
+set(Aurora_Libraries $<LINK_ONLY:MKL::MKL> OpenMP::OpenMP_CXX ${CUDA_cublas_LIBRARY} ${CUDA_cusolver_LIBRARY})
 set(Aurora_FOUND TRUE)
 message(Aurora Found)
\ No newline at end of file
diff --git a/src/Function2D.cu b/src/Function2D.cu
index 25e817f..2669038 100644
--- a/src/Function2D.cu
+++ b/src/Function2D.cu
@@ -156,7 +156,6 @@ CudaMatrix Aurora::max(const CudaMatrix &aMatrix, FunctionDirection direction, l
 CudaMatrix vxmMax(CudaMatrix aVec, CudaMatrix aMat) {
     //col-vec x mat
     if (aVec.getDimSize(1) == 1 && aVec.getDimSize(0) == aMat.getDimSize(0)) {
-        std::cout<<"max mat and col-vec "<<std::endl;
         size_t size = aMat.getDataSize();
         float* data = nullptr;
         cudaMalloc((void**)&data, sizeof(float) * size);
@@ -175,7 +174,6 @@ CudaMatrix vxmMax(CudaMatrix aVec, CudaMatrix aMat) {
     // row-vec x mat
     else if (aVec.getDimSize(0) == 1 && aVec.getDimSize(1) == aMat.getDimSize(1))
     {
-        std::cout<<"max mat and row-vec "<<std::endl;
         size_t size = aMat.getDataSize() ;
         float* data = nullptr;
         cudaMalloc((void**)&data, sizeof(float) * size);
@@ -376,7 +374,6 @@ CudaMatrix Aurora::min(const CudaMatrix &aMatrix, FunctionDirection direction, l
 CudaMatrix vxmMin(CudaMatrix aVec, CudaMatrix aMat) {
     //col-vec x mat
     if (aVec.getDimSize(1) == 1 && aVec.getDimSize(0) == aMat.getDimSize(0)) {
-        std::cout<<"min mat and col-vec "<<std::endl;
         size_t size = aMat.getDataSize();
         float* data = nullptr;
         cudaMalloc((void**)&data, sizeof(float) * size);
@@ -395,7 +392,6 @@ CudaMatrix vxmMin(CudaMatrix aVec, CudaMatrix aMat) {
     // row-vec x mat
     else if (aVec.getDimSize(0) == 1 && aVec.getDimSize(1) == aMat.getDimSize(1))
     {
-        std::cout<<"min mat and row-vec "<<std::endl;
         size_t size = aMat.getDataSize() ;
         float* data = nullptr;
         cudaMalloc((void**)&data, sizeof(float) * size);
@@ -682,7 +678,6 @@ CudaMatrix Aurora::sum(const CudaMatrix &aMatrix, FunctionDirection direction ){
             case Column:
             default:
             {
-                std::cout<<"Column sum"<<std::endl;
                 float* matData = aMatrix.getData();
                 float* retData = nullptr;
                 int colElementCount = aMatrix.getDimSize(0);
@@ -793,6 +788,21 @@ CudaMatrix Aurora::mean(const CudaMatrix &aMatrix, FunctionDirection direction )
         return CudaMatrix();
     }
 }
+
+CudaMatrix Aurora::std(const CudaMatrix &aMatrix){
+    if (aMatrix.getDimSize(2) > 1 || aMatrix.isComplex()) {
+        std::cerr
+                << (aMatrix.getDimSize(2) > 1 ? "std() not support 3D data!" : "std() not support complex value type!")
+                << std::endl;
+        return CudaMatrix();
+    }
+
+    auto src = aMatrix.isComplex() ? Aurora::abs(aMatrix) : aMatrix.deepCopy();
+    int calc_size = src.getDimSize(0) == 1 ? src.getDimSize(1) : src.getDimSize(0);
+    auto meanM = Aurora::mean(src);
+    return sqrt(Aurora::sum((src-meanM)^2.0)/((float)calc_size-1.0f));
+}
+
 template <typename ValueType>
 class RowElementIterator:public thrust::iterator_facade<
         RowElementIterator<ValueType>, 
@@ -1294,12 +1304,13 @@ __global__ void complexFillKernel(float* aInputData, float* aOutput,unsigned int
     for (int offset = 0; offset < aDesColEleCount; offset+=blockDim.x)
     {
         if(threadIdx.x + offset< aCopySize){
-            aOutput[2*idx_d] = aInputData[idx_s];
-            aOutput[2*idx_d + 1] = 0;
+            aOutput[2 * idx_d + offset * 2] = aInputData[idx_s + offset];
+            aOutput[2 * idx_d + offset * 2 + 1] = 0;
+            
         }
         else if(threadIdx.x + offset< aDesColEleCount){
-            aOutput[2*idx_d] = 0;
-            aOutput[2*idx_d + 1] = 0;
+            aOutput[2 * idx_d + offset * 2] = 0;
+            aOutput[2 * idx_d + offset * 2 + 1] = 0;
         }
         else{
             return;
@@ -1316,12 +1327,12 @@ __global__ void complexCopyKernel(float* aInputData, float* aOutput,unsigned int
     for (int offset = 0; offset < aDesColEleCount; offset+=blockDim.x)
     {
         if(threadIdx.x + offset< aCopySize){
-            aOutput[2*idx_d] = aInputData[idx_s*2];
-            aOutput[2*idx_d + 1] = aInputData[idx_s*2+1];
+            aOutput[2*idx_d + offset * 2 ] = aInputData[idx_s*2 + offset*2];
+            aOutput[2*idx_d + offset*2+ 1] = aInputData[idx_s*2+ offset*2+1];
         }
         else if(threadIdx.x + offset< aDesColEleCount){
-            aOutput[2*idx_d] = 0;
-            aOutput[2*idx_d + 1] = 0;
+            aOutput[2*idx_d + offset*2] = 0;
+            aOutput[2*idx_d + offset*2+ 1] = 0;
         }
         else{
             return;
@@ -1344,7 +1355,9 @@ if (aMatrix.isComplex()){
     complexFillKernel<<<aMatrix.getDimSize(1), 256>>>(aMatrix.getData(), data, needCopySize, aMatrix.getDimSize(0),ColEleCount);
 }
     auto ret = Aurora::CudaMatrix::fromRawData(data,ColEleCount,aMatrix.getDimSize(1),1,Complex);
+    auto mm = ret.toHostMatrix();
     ExecFFT(ret,0);
+    mm = ret.toHostMatrix();
     return ret;
 }
 
diff --git a/src/Function2D.cuh b/src/Function2D.cuh
index 2a3e502..26c8428 100644
--- a/src/Function2D.cuh
+++ b/src/Function2D.cuh
@@ -26,6 +26,14 @@ namespace Aurora
      */
     CudaMatrix mean(const CudaMatrix &aMatrix, FunctionDirection direction = Column);
 
+    /**
+     * @brief 标准差，只支持列方向
+     * 
+     * @param aMatrix 
+     * @return CudaMatrix 
+     */
+    CudaMatrix std(const CudaMatrix &aMatrix);
+
     CudaMatrix sort(const CudaMatrix &aMatrix,FunctionDirection direction = Column);
     CudaMatrix sort(CudaMatrix &&aMatrix,FunctionDirection direction = Column);
 
diff --git a/test/CudaMatrix_Test.cpp b/test/CudaMatrix_Test.cpp
index e84baf8..5bb8273 100644
--- a/test/CudaMatrix_Test.cpp
+++ b/test/CudaMatrix_Test.cpp
@@ -26,11 +26,12 @@ protected:
     }
 };
 
-TEST_F(CudaMatrix_Test, MatrixAdd) {
+TEST_F(CudaMatrix_Test, MatrixAddScalar) {
     {
-        auto A = Aurora::zeros(1000,1,1);
-        auto B = Aurora::zeros(1000,1,1);
-        for (size_t i = 0; i < 1000; i++)
+        auto A = Aurora::zeros(257,257,1);
+        auto B = Aurora::zeros(257,257,1);
+        size_t size = 257*257;
+        for (size_t i = 0; i < size; i++)
         {
             A[i] = -1;
             B[i] =  i;
@@ -42,35 +43,48 @@ TEST_F(CudaMatrix_Test, MatrixAdd) {
         auto dB = B.toDeviceMatrix();
         auto dC = (dA+dB);
         auto dhC = dC.toHostMatrix();
-        for (size_t i = 0; i < 1000; i++)
+        EXPECT_EQ(257, dC.getDimSize(0));
+        EXPECT_EQ(257, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+
+        for (size_t i = 0; i < size; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dhC[i]);
+            EXPECT_FLOAT_EQ(C[i],dhC[i]);
         }
         printf("Test CudaMatrix operator+(float aScalar) const \r\n");
         //CudaMatrix operator+(float aScalar) const
         auto D = C+0.5;
         auto dD = dC+0.5;
         auto dhD = dD.toHostMatrix();
-        for (size_t i = 0; i < 1000; i++)
+        EXPECT_EQ(257, dD.getDimSize(0));
+        EXPECT_EQ(257, dD.getDimSize(1));
+        EXPECT_EQ(1, dD.getDimSize(2));
+        for (size_t i = 0; i < size; i++)
         {
-            ASSERT_FLOAT_EQ(D[i],dhD[i]);
+            EXPECT_FLOAT_EQ(D[i],dhD[i]);
         }
         printf("Test CudaMatrix operator+(float aScalar, const CudaMatrix &aMatrix) \r\n");
         // CudaMatrix operator+(float aScalar, const CudaMatrix &aMatrix)
         dD = 0.5 + dC;
         dhD = dD.toHostMatrix();
-        for (size_t i = 0; i < 1000; i++)
+        EXPECT_EQ(257, dD.getDimSize(0));
+        EXPECT_EQ(257, dD.getDimSize(1));
+        EXPECT_EQ(1, dD.getDimSize(2));
+        for (size_t i = 0; i < size; i++)
         {
-            ASSERT_FLOAT_EQ(D[i],dhD[i]);
+            EXPECT_FLOAT_EQ(D[i],dhD[i]);
         }
         printf("Test CudaMatrix &operator+(float aScalar, CudaMatrix &&aMatrix) \r\n");
         // CudaMatrix &operator+(float aScalar, CudaMatrix &&aMatrix)
         {
             auto dD2 = 0.5 + (dA+dB);
             dhD = dD2.toHostMatrix();
-            for (size_t i = 0; i < 1000; i++)
+            EXPECT_EQ(257, dD2.getDimSize(0));
+            EXPECT_EQ(257, dD2.getDimSize(1));
+            EXPECT_EQ(1, dD2.getDimSize(2));
+            for (size_t i = 0; i < size; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         printf("Test CudaMatrix &operator+(CudaMatrix &&aMatrix, float aScalar) \r\n");
@@ -79,9 +93,12 @@ TEST_F(CudaMatrix_Test, MatrixAdd) {
             
             auto dD2 =  (dA+dB)+0.5;
             dhD = dD2.toHostMatrix();
-            for (size_t i = 0; i < 1000; i++)
+            EXPECT_EQ(257, dD2.getDimSize(0));
+            EXPECT_EQ(257, dD2.getDimSize(1));
+            EXPECT_EQ(1, dD2.getDimSize(2));
+            for (size_t i = 0; i < size; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         //CudaMatrix operator+(CudaMatrix &&aMatrix) const
@@ -90,9 +107,12 @@ TEST_F(CudaMatrix_Test, MatrixAdd) {
             auto D = A+C;
             auto dD2 = dA+(dA+dB);
             dhD = dD2.toHostMatrix();
-            for (size_t i = 0; i < 1000; i++)
+            EXPECT_EQ(257, dD2.getDimSize(0));
+            EXPECT_EQ(257, dD2.getDimSize(1));
+            EXPECT_EQ(1, dD2.getDimSize(2));
+            for (size_t i = 0; i < size; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         //CudaMatrix operator+(CudaMatrix &&aMatrix,CudaMatrix &aOther)
@@ -101,9 +121,12 @@ TEST_F(CudaMatrix_Test, MatrixAdd) {
             auto D = A+C;
             auto dD2 = (dA+dB)+dA;
             dhD = dD2.toHostMatrix();
-            for (size_t i = 0; i < 1000; i++)
+            EXPECT_EQ(257, dD2.getDimSize(0));
+            EXPECT_EQ(257, dD2.getDimSize(1));
+            EXPECT_EQ(1, dD2.getDimSize(2));
+            for (size_t i = 0; i < size; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
     }
@@ -120,41 +143,135 @@ TEST_F(CudaMatrix_Test, MatrixAdd) {
         auto C = A+B;
         std::complex<float> scalar(-1,-1);
         auto dA = A.toDeviceMatrix();
-
+        //complex matrix + complex scalar
         auto dC = (dA+scalar);
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
-
+        
+        //complex Matrix&& + complex scalar
         dC = A.toDeviceMatrix()+scalar;
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
-
+        //complex scalar + complex Matrix&&
         dC = scalar+A.toDeviceMatrix();
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
+        //complex scalar + complex matrix
+        dC = scalar+dA;
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        for (size_t i = 0; i < C.getDataSize()*2; i++)
+        {
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        }
+        //complex matrix + real scalar
         C = A+1;
         dC = (dA+1);
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index"<<i;
         }
-
+        //complex Matrix&& + real scalar 
         dC = A.toDeviceMatrix()+1;
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
+        //real scalar + complex Matrix &&
         dC = 1+A.toDeviceMatrix();
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        }
+    }
+    {
+        float *dataA = new float[5000];
+        float *dataB = new float[10000];
+        for (size_t i = 0; i < 5000; i++)
+        {
+            dataA[i] = i;
+            dataB[2*i] = -1;
+            dataB[2*i+1] = -1;
+        }
+        auto A = Aurora::Matrix::fromRawData(dataA, 50,100,1);
+        auto B = Aurora::Matrix::fromRawData(dataB, 50,100,1,Aurora::Complex);
+        auto C = A+B;
+        std::complex<float> scalar(-1,-1);
+        auto dA = A.toDeviceMatrix();
+        //real matrix + complex scalar
+        auto dC = (dA+scalar);
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        for (size_t i = 0; i < C.getDataSize()*2; i++)
+        {
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        }
+        
+        //real Matrix&& + complex scalar
+        dC = A.toDeviceMatrix()+scalar;
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        for (size_t i = 0; i < C.getDataSize()*2; i++)
+        {
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        }
+        //complex scalar + real Matrix&&
+        dC = scalar+A.toDeviceMatrix();
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        for (size_t i = 0; i < C.getDataSize()*2; i++)
+        {
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        }
+
+        //complex scalar + real Matrix
+        dC = scalar+dA;
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(50, dC.getDimSize(0));
+        EXPECT_EQ(100, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        for (size_t i = 0; i < C.getDataSize()*2; i++)
+        {
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -162,64 +279,89 @@ TEST_F(CudaMatrix_Test, MatrixAdd) {
 TEST_F(CudaMatrix_Test, MatrixAddmm) {
     //real
     {
-        auto A = Aurora::zeros(100,100,1);
-        auto B = Aurora::zeros(100,100,1);
-        for (size_t i = 0; i < 10000; i++)
-        {
-            A[i] = -1;
-            B[i] =  8;
-        }
+        auto A = -Aurora::ones(4096,23519,1);
+        auto B = Aurora::ones(4096,23519,1)*8;
         auto C = B-1;
         auto dA = A.toDeviceMatrix();
         auto dB = B.toDeviceMatrix();
-        //Matrix& + Matrix&, col mode
+        //Matrix& + Matrix&
         auto dC = (dA+dB);
+        EXPECT_TRUE(!dC.isComplex());
+        EXPECT_EQ(4096, dC.getDimSize(0));
+        EXPECT_EQ(23519, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        #pragma omp parallel for
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
-        //Matrix&& + Matrix&, col mode
+        //Matrix&& + Matrix&
         dC = A.toDeviceMatrix()+dB;
-        for (size_t i = 0; i < C.getDataSize(); i++)
-        {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
-        }
-        //Matrix& + Matrix&&, col mode
+        EXPECT_TRUE(!dC.isComplex());
+        EXPECT_EQ(4096, dC.getDimSize(0));
+        EXPECT_EQ(23519, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        // #pragma omp parallel for
+        // for (size_t i = 0; i < C.getDataSize(); i++)
+        // {
+        //     EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        // }
+        //Matrix& + Matrix&&
         dC = dB+A.toDeviceMatrix();
-        for (size_t i = 0; i < C.getDataSize(); i++)
-        {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
-        }
+        EXPECT_TRUE(!dC.isComplex());
+        EXPECT_EQ(4096, dC.getDimSize(0));
+        EXPECT_EQ(23519, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        // #pragma omp parallel for
+        // for (size_t i = 0; i < C.getDataSize(); i++)
+        // {
+        //     EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        // }
     }
     //complex
     {
-        float *dataA = Aurora::random(100*100*2);
-        float *dataB = Aurora::random(100*100*2);
+        float *dataA = Aurora::random(4096*23519*2);
+        float *dataB = Aurora::random(4096*23519*2);
 
-        auto A = Aurora::Matrix::fromRawData(dataA, 100,100,1,Aurora::Complex);
-        auto B = Aurora::Matrix::fromRawData(dataB, 100,100,1,Aurora::Complex);
+        auto A = Aurora::Matrix::fromRawData(dataA, 4096,23519,1,Aurora::Complex);
+        auto B = Aurora::Matrix::fromRawData(dataB, 4096,23519,1,Aurora::Complex);
 
         auto C = B+A;
         auto dA = A.toDeviceMatrix();
         auto dB = B.toDeviceMatrix();
         //Matrix& + Matrix&, 
         auto dC = (dA+dB);
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(4096, dC.getDimSize(0));
+        EXPECT_EQ(23519, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        #pragma omp parallel for
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&,
         dC = A.toDeviceMatrix()+dB;
-        for (size_t i = 0; i < C.getDataSize(); i++)
-        {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
-        }
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(4096, dC.getDimSize(0));
+        EXPECT_EQ(23519, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        // #pragma omp parallel for
+        // for (size_t i = 0; i < C.getDataSize(); i++)
+        // {
+        //     EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        // }
         //Matrix& + Matrix&&, 
         dC = dB+A.toDeviceMatrix();
-        for (size_t i = 0; i < C.getDataSize(); i++)
-        {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
-        }
+        EXPECT_TRUE(dC.isComplex());
+        EXPECT_EQ(4096, dC.getDimSize(0));
+        EXPECT_EQ(23519, dC.getDimSize(1));
+        EXPECT_EQ(1, dC.getDimSize(2));
+        // #pragma omp parallel for
+        // for (size_t i = 0; i < C.getDataSize(); i++)
+        // {
+        //     EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
+        // }
     }
 }
 
@@ -235,31 +377,31 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         auto dC = (dA+dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()+dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, col mode
         dC = dB+A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA+B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, col mode
         dC = B.toDeviceMatrix()+dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 100, 1);
 
@@ -267,13 +409,13 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         dC = (dA+dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix& + Matrix&&, row mode
         dC = dB+A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, row mode
         dC = A.toDeviceMatrix()+dB;
@@ -285,13 +427,13 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         dC = dA+B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, row mode
         dC = B.toDeviceMatrix()+dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complex
@@ -315,31 +457,31 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         auto dC = (dA+dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()+dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, col mode
         dC = dB+A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA+B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, col mode
         dC = B.toDeviceMatrix()+dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 50, 1);
         B.forceReshape(1, 50, 1);
@@ -350,13 +492,13 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         dC = (dA+dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix& + Matrix&&, row mode
         dC = dB+A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, row mode
         dC = A.toDeviceMatrix()+dB;
@@ -368,13 +510,13 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         dC = dA+B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, row mode
         dC = B.toDeviceMatrix()+dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complexM - realv
@@ -398,31 +540,31 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         auto dC = (dA+dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + vec&, col mode
         dC = A.toDeviceMatrix()+dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec& + Matrix&&, col mode
         dC = dB+A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA+B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&, col mode
         dC = B.toDeviceMatrix()+dA;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 100, 1);
         dA.forceReshape(50, 100, 1);
@@ -433,32 +575,32 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         dC = (dA+dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
 
         //Matrix&& + vec&, row mode
         dC = A.toDeviceMatrix()+dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec& + Matrix&&, row mode
         dC = dB+A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, row mode
         dC = dA+B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&, row mode
         dC = B.toDeviceMatrix()+dA;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //realM - complexV
@@ -485,19 +627,19 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         auto dC = (dA+dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()+dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, col mode
         dC = dB+A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 50, 1);
         dA.forceReshape(100, 50, 1);
@@ -507,20 +649,20 @@ TEST_F(CudaMatrix_Test, MatrixAddmv) {
         dC = (dA+dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
 
         //Matrix&& + Matrix&, row mode
         dC = A.toDeviceMatrix()+dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, row mode
         dC = dB+A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -543,7 +685,7 @@ TEST_F(CudaMatrix_Test, MatrixMul) {
         auto dhC = dC.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dhC[i]);
+            EXPECT_FLOAT_EQ(C[i],dhC[i]);
         }
         printf("Test CudaMatrix operator*(float aScalar) const \r\n");
         //CudaMatrix operator+(float aScalar) const
@@ -552,7 +694,7 @@ TEST_F(CudaMatrix_Test, MatrixMul) {
         auto dhD = dD.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(D[i],dhD[i]);
+            EXPECT_FLOAT_EQ(D[i],dhD[i]);
         }
         printf("Test CudaMatrix operator*(float aScalar, const CudaMatrix &aMatrix) \r\n");
         // CudaMatrix operator+(float aScalar, const CudaMatrix &aMatrix)
@@ -560,7 +702,7 @@ TEST_F(CudaMatrix_Test, MatrixMul) {
         dhD = dD.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(D[i],dhD[i]);
+            EXPECT_FLOAT_EQ(D[i],dhD[i]);
         }
         printf("Test CudaMatrix &operator*(float aScalar, CudaMatrix &&aMatrix) \r\n");
         // CudaMatrix &operator+(float aScalar, CudaMatrix &&aMatrix)
@@ -569,7 +711,7 @@ TEST_F(CudaMatrix_Test, MatrixMul) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         printf("Test CudaMatrix &operator*(CudaMatrix &&aMatrix, float aScalar) \r\n");
@@ -580,7 +722,7 @@ TEST_F(CudaMatrix_Test, MatrixMul) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         //CudaMatrix operator+(CudaMatrix &&aMatrix) const
@@ -591,7 +733,7 @@ TEST_F(CudaMatrix_Test, MatrixMul) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         //CudaMatrix operator*(CudaMatrix &&aMatrix,CudaMatrix &aOther)
@@ -602,7 +744,7 @@ TEST_F(CudaMatrix_Test, MatrixMul) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
     }
@@ -623,37 +765,37 @@ TEST_F(CudaMatrix_Test, MatrixMul) {
         auto dC = (dA*scalar);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = A.toDeviceMatrix()*scalar;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = scalar*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = A*2;
         dC = (dA*2);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index"<<i;
         }
 
         dC = A.toDeviceMatrix()*2;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = 2*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -675,19 +817,19 @@ TEST_F(CudaMatrix_Test, MatrixMulmm) {
         auto dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& * Matrix&, col mode
         dC = A.toDeviceMatrix()*dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& * Matrix&&, col mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complex
@@ -710,19 +852,19 @@ TEST_F(CudaMatrix_Test, MatrixMulmm) {
         auto dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()*dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, col mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -747,31 +889,31 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         auto dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()*dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, col mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA*B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, col mode
         dC = B.toDeviceMatrix()*dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 100, 1);
 
@@ -779,13 +921,13 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix& + Matrix&&, row mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, row mode
         dC = A.toDeviceMatrix()*dB;
@@ -797,13 +939,13 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         dC = dA*B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, row mode
         dC = B.toDeviceMatrix()*dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complex
@@ -827,31 +969,31 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         auto dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()*dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, col mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA*B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, col mode
         dC = B.toDeviceMatrix()*dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 50, 1);
         B.forceReshape(1, 50, 1);
@@ -862,13 +1004,13 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix& + Matrix&&, row mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, row mode
         dC = A.toDeviceMatrix()*dB;
@@ -880,13 +1022,13 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         dC = dA*B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, row mode
         dC = B.toDeviceMatrix()*dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complexM - realv
@@ -910,31 +1052,31 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         auto dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + vec&, col mode
         dC = A.toDeviceMatrix()*dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec& * Matrix&&, col mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA*B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&, col mode
         dC = B.toDeviceMatrix()*dA;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 100, 1);
         dA.forceReshape(50, 100, 1);
@@ -945,32 +1087,32 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
 
         //Matrix&& + vec&, row mode
         dC = A.toDeviceMatrix()*dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec& + Matrix&&, row mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, row mode
         dC = dA*B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&, row mode
         dC = B.toDeviceMatrix()*dA;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //realM - complexV
@@ -997,19 +1139,19 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         auto dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()*dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, col mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 50, 1);
         dA.forceReshape(100, 50, 1);
@@ -1019,20 +1161,20 @@ TEST_F(CudaMatrix_Test, MatrixMulmv) {
         dC = (dA*dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
 
         //Matrix&& + Matrix&, row mode
         dC = A.toDeviceMatrix()*dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + Matrix&&, row mode
         dC = dB*A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -1055,7 +1197,7 @@ TEST_F(CudaMatrix_Test, MatrixSub) {
         auto dhC = dC.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dhC[i]);
+            EXPECT_FLOAT_EQ(C[i],dhC[i]);
         }
         printf("Test CudaMatrix operator-(float aScalar) const \r\n");
         //CudaMatrix operator+(float aScalar) const
@@ -1064,7 +1206,7 @@ TEST_F(CudaMatrix_Test, MatrixSub) {
         auto dhD = dD.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(D[i],dhD[i]);
+            EXPECT_FLOAT_EQ(D[i],dhD[i]);
         }
         printf("Test CudaMatrix operator-(float aScalar, const CudaMatrix &aMatrix) \r\n");
         // CudaMatrix operator+(float aScalar, const CudaMatrix &aMatrix)
@@ -1073,7 +1215,7 @@ TEST_F(CudaMatrix_Test, MatrixSub) {
         dhD = dD.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(D[i],dhD[i]);
+            EXPECT_FLOAT_EQ(D[i],dhD[i]);
         }
         printf("Test CudaMatrix &operator-(float aScalar, CudaMatrix &&aMatrix) \r\n");
         // CudaMatrix &operator+(float aScalar, CudaMatrix &&aMatrix)
@@ -1082,7 +1224,7 @@ TEST_F(CudaMatrix_Test, MatrixSub) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         D =  C - 0.5;
@@ -1094,7 +1236,7 @@ TEST_F(CudaMatrix_Test, MatrixSub) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         //CudaMatrix operator+(CudaMatrix &&aMatrix) const
@@ -1105,7 +1247,7 @@ TEST_F(CudaMatrix_Test, MatrixSub) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         //CudaMatrix operator+(CudaMatrix &&aMatrix,CudaMatrix &aOther)
@@ -1116,7 +1258,7 @@ TEST_F(CudaMatrix_Test, MatrixSub) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
     }
@@ -1137,37 +1279,37 @@ TEST_F(CudaMatrix_Test, MatrixSub) {
         auto dC = (dA-scalar);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = A.toDeviceMatrix()-scalar;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = B-A;
         dC = scalar-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = A-1;
         dC = (dA-1);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index"<<i;
         }
 
         dC = A.toDeviceMatrix()-1;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = 1-A;
         dC = 1-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -1189,19 +1331,19 @@ TEST_F(CudaMatrix_Test, MatrixSubmm) {
         auto dC = (dB-dA);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = dB - A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = B.toDeviceMatrix()-dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complex
@@ -1223,20 +1365,20 @@ TEST_F(CudaMatrix_Test, MatrixSubmm) {
         auto dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&
         dC = A.toDeviceMatrix()-dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = B-A;
         //Matrix& + Matrix&&,
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //real & complex
@@ -1259,38 +1401,38 @@ TEST_F(CudaMatrix_Test, MatrixSubmm) {
         auto dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i<<", A:"<<A[i]<<", B:"<<B[i/2];
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i<<", A:"<<A[i]<<", B:"<<B[i/2];
         }
         //Matrix&&(real) - Matrix&(complex)
         dC = A.toDeviceMatrix()-dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&(real) - Matrix&&(complex)
         dC = dA-B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = B-A;
         //Matrix&(complex) - Matrix&&(real)
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&&(complex) - Matrix&(real)
         dC = B.toDeviceMatrix()-dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&(complex) - Matrix&(real)
         dC = dB-dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -1307,32 +1449,32 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         auto dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()-dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA-B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = 1- A;
         //Matrix& + Matrix&&, col mode
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, col mode
         dC = B.toDeviceMatrix()-dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 100, 1);
         C = A-1;
@@ -1340,7 +1482,7 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& - Matrix&, row mode
         dC = A.toDeviceMatrix()-dB;
@@ -1352,20 +1494,20 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         dC = dA-B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = 1- A;
         //Matrix& + Matrix&&, row mode
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, row mode
         dC = B.toDeviceMatrix()-dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complex
@@ -1389,32 +1531,32 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         auto dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()-dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA-B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = D-A;
         //Matrix& + Matrix&&, col mode
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //vec&& + Matrix&&, col mode
         dC = B.toDeviceMatrix()-dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 50, 1);
         B.forceReshape(1, 50, 1);
@@ -1427,7 +1569,7 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& - Matrix&, row mode
         dC = A.toDeviceMatrix()-dB;
@@ -1439,7 +1581,7 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         dC = dA-B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         C= D - A;
@@ -1447,14 +1589,14 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         //vec&& + Matrix&&, row mode
         dC = B.toDeviceMatrix()-dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complexM - realv
@@ -1479,33 +1621,33 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         auto dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + vec&, col mode
         dC = A.toDeviceMatrix()-dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA-B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C=1-A;
         //vec& + Matrix&&, col mode
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         //vec&& + Matrix&, col mode
         dC = B.toDeviceMatrix()-dA;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 100, 1);
         dA.forceReshape(50, 100, 1);
@@ -1517,19 +1659,19 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& + vec&, row mode
         dC = A.toDeviceMatrix()-dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, row mode
         dC = dA-B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         C = 1-A;
@@ -1537,14 +1679,14 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         //vec&& + Matrix&, row mode
         dC = B.toDeviceMatrix()-dA;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //realM - complexV
@@ -1575,20 +1717,20 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         auto dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()-dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C= D-E;
         //Matrix& + Matrix&&, col mode
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 50, 1);
         dA.forceReshape(100, 50, 1);
@@ -1599,21 +1741,21 @@ TEST_F(CudaMatrix_Test, MatrixSubmv) {
         dC = (dA-dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
 
         //Matrix&& + Matrix&, row mode
         dC = A.toDeviceMatrix()-dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C= D-E;
         //Matrix& + Matrix&&, row mode
         dC = dB-A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
     }
 }
@@ -1637,7 +1779,7 @@ TEST_F(CudaMatrix_Test, MatrixDiv) {
         auto dhC = dC.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dhC[i]);
+            EXPECT_FLOAT_EQ(C[i],dhC[i]);
         }
         printf("Test CudaMatrix operator/(float aScalar) const \r\n");
         //CudaMatrix operator+(float aScalar) const
@@ -1646,7 +1788,7 @@ TEST_F(CudaMatrix_Test, MatrixDiv) {
         auto dhD = dD.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(D[i],dhD[i]);
+            EXPECT_FLOAT_EQ(D[i],dhD[i]);
         }
         printf("Test CudaMatrix operator/(float aScalar, const CudaMatrix &aMatrix) \r\n");
         // CudaMatrix operator+(float aScalar, const CudaMatrix &aMatrix)
@@ -1655,7 +1797,7 @@ TEST_F(CudaMatrix_Test, MatrixDiv) {
         dhD = dD.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(D[i],dhD[i]);
+            EXPECT_FLOAT_EQ(D[i],dhD[i]);
         }
         printf("Test CudaMatrix &operator/(float aScalar, CudaMatrix &&aMatrix) \r\n");
         // CudaMatrix &operator+(float aScalar, CudaMatrix &&aMatrix)
@@ -1664,7 +1806,7 @@ TEST_F(CudaMatrix_Test, MatrixDiv) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         D =  C / 0.5;
@@ -1676,7 +1818,7 @@ TEST_F(CudaMatrix_Test, MatrixDiv) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         //CudaMatrix operator+(CudaMatrix &&aMatrix) const
@@ -1687,7 +1829,7 @@ TEST_F(CudaMatrix_Test, MatrixDiv) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
         //CudaMatrix operator+(CudaMatrix &&aMatrix,CudaMatrix &aOther)
@@ -1698,7 +1840,7 @@ TEST_F(CudaMatrix_Test, MatrixDiv) {
             dhD = dD2.toHostMatrix();
             for (size_t i = 0; i < 1000; i++)
             {
-                ASSERT_FLOAT_EQ(D[i],dhD[i]);
+                EXPECT_FLOAT_EQ(D[i],dhD[i]);
             }
         }
     }
@@ -1720,37 +1862,37 @@ TEST_F(CudaMatrix_Test, MatrixDiv) {
         auto dC = (dA/scalar);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = A.toDeviceMatrix()/scalar;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = B/A;
         dC = scalar/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = A/2.0;
         dC = (dA/2.0);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index"<<i;
         }
 
         dC = A.toDeviceMatrix()/2.0;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = 2.0/A;
         dC = 2.0/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -1772,19 +1914,19 @@ TEST_F(CudaMatrix_Test, MatrixDivmm) {
         auto dC = (dB/dA);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = dB / A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         dC = B.toDeviceMatrix()/dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complex
@@ -1806,20 +1948,20 @@ TEST_F(CudaMatrix_Test, MatrixDivmm) {
         auto dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& / Matrix&
         dC = A.toDeviceMatrix()/dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = B/A;
         //Matrix& / Matrix&&,
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //real & complex
@@ -1842,38 +1984,38 @@ TEST_F(CudaMatrix_Test, MatrixDivmm) {
         auto dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i<<", A:"<<A[i]<<", B:"<<B[i/2];
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i<<", A:"<<A[i]<<", B:"<<B[i/2];
         }
         //Matrix&&(real) / Matrix&(complex)
         dC = A.toDeviceMatrix()/dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&(real) / Matrix&&(complex)
         dC = dA/B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = B/A;
         //Matrix&(complex) / Matrix&&(real)
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&&(complex) / Matrix&(real)
         dC = B.toDeviceMatrix()/dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&(complex) / Matrix&(real)
         dC = dB/dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
 }
@@ -1890,32 +2032,32 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         auto dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()/dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA/B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = 2/A;
         //Matrix& + Matrix&&, col mode
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, col mode
         dC = B.toDeviceMatrix()/dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 100, 1);
         C = A/2;
@@ -1923,7 +2065,7 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& - Matrix&, row mode
         dC = A.toDeviceMatrix()/dB;
@@ -1935,20 +2077,20 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         dC = dA/B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = 2/ A;
         //Matrix& + Matrix&&, row mode
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //vec&& + Matrix&&, row mode
         dC = B.toDeviceMatrix()/dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complex
@@ -1972,32 +2114,32 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         auto dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()/dB;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, col mode
         dC = dA/B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C = D/A;
         //Matrix& + Matrix&&, col mode
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //vec&& + Matrix&&, col mode
         dC = B.toDeviceMatrix()/dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 50, 1);
         B.forceReshape(1, 50, 1);
@@ -2010,7 +2152,7 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& / Matrix&, row mode
         dC = A.toDeviceMatrix()/dB;
@@ -2022,7 +2164,7 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         dC = dA/B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         C= D / A;
@@ -2030,14 +2172,14 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         //vec&& / Matrix&&, row mode
         dC = B.toDeviceMatrix()/dA;
         for (size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //complexM - realv
@@ -2062,33 +2204,33 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         auto dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix&& / vec&, col mode
         dC = A.toDeviceMatrix()/dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& / vec&&, col mode
         dC = dA/B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C=2/A;
         //vec& + Matrix&&, col mode
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         //vec&& + Matrix&, col mode
         dC = B.toDeviceMatrix()/dA;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 100, 1);
         dA.forceReshape(50, 100, 1);
@@ -2100,19 +2242,19 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& + vec&, row mode
         dC = A.toDeviceMatrix()/dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         //Matrix& + vec&&, row mode
         dC = dA/B.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         C = 2/A;
@@ -2120,14 +2262,14 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
 
         //vec&& + Matrix&, row mode
         dC = B.toDeviceMatrix()/dA;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
     }
     //realM - complexV
@@ -2158,20 +2300,20 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         auto dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
         //Matrix&& + Matrix&, col mode
         dC = A.toDeviceMatrix()/dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C= D/E;
         //Matrix& + Matrix&&, col mode
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         dB.forceReshape(1, 50, 1);
         dA.forceReshape(100, 50, 1);
@@ -2182,21 +2324,21 @@ TEST_F(CudaMatrix_Test, MatrixDivmv) {
         dC = (dA/dB);
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
 
         //Matrix&& + Matrix&, row mode
         dC = A.toDeviceMatrix()/dB;
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i));
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i));
         }
         C= D/E;
         //Matrix& + Matrix&&, row mode
         dC = dB/A.toDeviceMatrix();
         for (size_t i = 0; i < C.getDataSize()*2; i++)
         {
-            ASSERT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
+            EXPECT_FLOAT_EQ(C[i],dC.getValue(i))<<"index:"<<i;
         }
     }
 }
@@ -2218,33 +2360,33 @@ TEST_F(CudaMatrix_Test, MatrixComplexMulAndDiv){
     auto ret2 = dA*dB;
     for (size_t i = 0; i < ret1.getDataSize()*2; i++)
     {
-        ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
+        EXPECT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
     }
     ret2 = A.toDeviceMatrix()*dB;
     for (size_t i = 0; i < ret1.getDataSize()*2; i++)
     {
-        ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
+        EXPECT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
     }
     ret2 = dA*B.toDeviceMatrix();
     for (size_t i = 0; i < ret1.getDataSize()*2; i++)
     {
-        ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
+        EXPECT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
     }
     ret1 = A/B;
     ret2 = dA/dB;
     for (size_t i = 0; i < ret1.getDataSize()*2; i++)
     {
-        ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
+        EXPECT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
     }
     ret2 = A.toDeviceMatrix()/dB;
     for (size_t i = 0; i < ret1.getDataSize()*2; i++)
     {
-        ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
+        EXPECT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
     }
     ret2 = dA/B.toDeviceMatrix();
     for (size_t i = 0; i < ret1.getDataSize()*2; i++)
     {
-        ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
+        EXPECT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<"index:"<<i;
     }
 
 }
@@ -2263,7 +2405,7 @@ TEST_F(CudaMatrix_Test, MatrixPow){
         auto dhR = dR.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2272,7 +2414,7 @@ TEST_F(CudaMatrix_Test, MatrixPow){
         auto dhR = dR.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
         {
@@ -2281,7 +2423,7 @@ TEST_F(CudaMatrix_Test, MatrixPow){
         auto dhR = dR.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2290,7 +2432,7 @@ TEST_F(CudaMatrix_Test, MatrixPow){
         auto dhR = dR.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2299,7 +2441,7 @@ TEST_F(CudaMatrix_Test, MatrixPow){
         auto dhR = dR.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
 
@@ -2307,7 +2449,6 @@ TEST_F(CudaMatrix_Test, MatrixPow){
 
 TEST_F(CudaMatrix_Test, MatrixNeg){
     auto A = Aurora::zeros(1000,1,1);
-    auto B = Aurora::zeros(1000,1,1);
     for (size_t i = 0; i < 1000; i++)
     {
         A[i] = -1+0.2*i;
@@ -2319,7 +2460,7 @@ TEST_F(CudaMatrix_Test, MatrixNeg){
         auto dhR = dR.toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
 }
@@ -2338,7 +2479,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (dA<dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2346,7 +2487,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (dA>dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }    
     {
@@ -2354,7 +2495,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (dA<=dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2362,7 +2503,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (dA>=dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2370,7 +2511,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (dA==dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2378,7 +2519,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (dA!=dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2386,7 +2527,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (9<dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2394,7 +2535,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (9>dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2402,7 +2543,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (9<=dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2410,7 +2551,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (9>=dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2418,7 +2559,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (9==dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
     {
@@ -2426,7 +2567,7 @@ TEST_F(CudaMatrix_Test, MatrixCompare){
         auto dhR = (9!=dB).toHostMatrix();
         for (size_t i = 0; i < 1000; i++)
         {
-            ASSERT_FLOAT_EQ(R[i],dhR[i]);
+            EXPECT_FLOAT_EQ(R[i],dhR[i]);
         }
     }
 }
@@ -2445,7 +2586,7 @@ TEST_F(CudaMatrix_Test, matrixfunction)
         Aurora::Matrix block2 = A.toHostMatrix().block(0, 1, 2);
         for (size_t i = 0; i < block1.getDataSize(); i++)
         {   
-            ASSERT_FLOAT_EQ(block1[i], block2[i]);
+            EXPECT_FLOAT_EQ(block1[i], block2[i]);
         }
 
         block2 = A.toHostMatrix();
@@ -2454,7 +2595,7 @@ TEST_F(CudaMatrix_Test, matrixfunction)
         block2.setBlockValue(0, 1, 2,-1);
         for (size_t i = 0; i < block1.getDataSize(); i++)
         {   
-            ASSERT_FLOAT_EQ(block1[i], block2[i]);
+            EXPECT_FLOAT_EQ(block1[i], block2[i]);
         }
 
         Aurora::CudaMatrix C = Aurora::zeros(2,3).toDeviceMatrix();
@@ -2463,7 +2604,7 @@ TEST_F(CudaMatrix_Test, matrixfunction)
         block2 = C.toHostMatrix();
         for(size_t i = 0; i < C.getDataSize(); i++)
         {
-            ASSERT_FLOAT_EQ(block1[i], block2[i]);
+            EXPECT_FLOAT_EQ(block1[i], block2[i]);
         }       
     }
 
diff --git a/test/Function2D_Cuda_Test.cpp b/test/Function2D_Cuda_Test.cpp
index e9f7b0e..b340214 100644
--- a/test/Function2D_Cuda_Test.cpp
+++ b/test/Function2D_Cuda_Test.cpp
@@ -39,7 +39,7 @@ TEST_F(Function2D_Cuda_Test, min)
         B = Aurora::Matrix::fromRawData(dataB, 4096, 41472);
         dB = B.toDeviceMatrix();
         long r,c;
-
+        // column
         auto ret1 = Aurora::min(B, Aurora::Column,r,c);
 
         auto ret2 = Aurora::min(dB, Aurora::Column,r,c);
@@ -53,7 +53,7 @@ TEST_F(Function2D_Cuda_Test, min)
             ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
         }
 
-
+        //row
         ret1 = Aurora::min(B, Aurora::FunctionDirection::Row,r,c);
 
         ret2 = Aurora::min(dB, Aurora::FunctionDirection::Row,r,c);
@@ -66,6 +66,23 @@ TEST_F(Function2D_Cuda_Test, min)
         {
             ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
         }
+        
+        //ALL
+        long r1,c1;
+        ret1 = Aurora::min(B, Aurora::All,r,c);
+
+        ret2 = Aurora::min(dB, Aurora::All,r1,c1);
+
+        ASSERT_EQ(ret1.getDimSize(0),ret2.getDimSize(0));
+        ASSERT_EQ(ret1.getDimSize(1),ret2.getDimSize(1));
+        ASSERT_EQ(ret1.getDimSize(2),ret2.getDimSize(2));
+        ASSERT_EQ(r, r1);
+        ASSERT_EQ(c, c1);
+
+        for (size_t i = 0; i < ret1.getDataSize(); i++)
+        {
+            ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
+        }
     }
     // different size speed 
     // Aurora::Matrix Aurora::min(const Aurora::Matrix &aMatrix,
@@ -100,6 +117,22 @@ TEST_F(Function2D_Cuda_Test, min)
         ASSERT_EQ(ret1.getDimSize(1),ret2.getDimSize(1));
         ASSERT_EQ(ret1.getDimSize(2),ret2.getDimSize(2));
 
+        for (size_t i = 0; i < ret1.getDataSize(); i++)
+        {
+            ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
+        }
+        //ALL
+        long r1,c1;
+        ret1 = Aurora::min(B, Aurora::All,r,c);
+
+        ret2 = Aurora::min(dB, Aurora::All,r1,c1);
+
+        ASSERT_EQ(ret1.getDimSize(0),ret2.getDimSize(0));
+        ASSERT_EQ(ret1.getDimSize(1),ret2.getDimSize(1));
+        ASSERT_EQ(ret1.getDimSize(2),ret2.getDimSize(2));
+        ASSERT_EQ(r, r1);
+        ASSERT_EQ(c, c1);
+
         for (size_t i = 0; i < ret1.getDataSize(); i++)
         {
             ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
@@ -254,6 +287,22 @@ TEST_F(Function2D_Cuda_Test, max)
         ASSERT_EQ(ret1.getDimSize(1),ret2.getDimSize(1));
         ASSERT_EQ(ret1.getDimSize(2),ret2.getDimSize(2));
 
+        for (size_t i = 0; i < ret1.getDataSize(); i++)
+        {
+            ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
+        }
+        //ALL
+        long r1,c1;
+        ret1 = Aurora::max(B, Aurora::All,r,c);
+
+        ret2 = Aurora::max(dB, Aurora::All,r1,c1);
+
+        ASSERT_EQ(ret1.getDimSize(0),ret2.getDimSize(0));
+        ASSERT_EQ(ret1.getDimSize(1),ret2.getDimSize(1));
+        ASSERT_EQ(ret1.getDimSize(2),ret2.getDimSize(2));
+        ASSERT_EQ(r, r1);
+        ASSERT_EQ(c, c1);
+
         for (size_t i = 0; i < ret1.getDataSize(); i++)
         {
             ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
@@ -296,6 +345,23 @@ TEST_F(Function2D_Cuda_Test, max)
         {
             ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
         }
+
+        //ALL
+        long r1,c1;
+        ret1 = Aurora::max(B, Aurora::All,r,c);
+
+        ret2 = Aurora::max(dB, Aurora::All,r1,c1);
+
+        ASSERT_EQ(ret1.getDimSize(0),ret2.getDimSize(0));
+        ASSERT_EQ(ret1.getDimSize(1),ret2.getDimSize(1));
+        ASSERT_EQ(ret1.getDimSize(2),ret2.getDimSize(2));
+        ASSERT_EQ(r, r1);
+        ASSERT_EQ(c, c1);
+
+        for (size_t i = 0; i < ret1.getDataSize(); i++)
+        {
+            ASSERT_FLOAT_EQ(ret1[i], ret2.getValue(i))<<", index at :"<<i;
+        }
     }
     // test
     // Aurora::Matrix Aurora::max(const Aurora::Matrix &aMatrix, float aValue)
@@ -907,4 +973,14 @@ TEST_F(Function2D_Cuda_Test, ifft_symmetric) {
     {
         EXPECT_FLOAT_AE(result1[i], result2[i]);
     }
-}
\ No newline at end of file
+}
+
+TEST_F(Function2D_Cuda_Test, std){
+    float *dataMA=  new float [9]{1, 2, 3, 2, 2, 6, 3, 3, 6};
+    auto A = Aurora::Matrix::fromRawData(dataMA,3,3);
+    auto D= Aurora::std(A.toDeviceMatrix());
+    EXPECT_FLOAT_EQ(1.0, D.getValue(0));
+    EXPECT_FLOAT_EQ(2.3094, fourDecimalRound(D.getValue(1)));
+    EXPECT_FLOAT_EQ(1.7321, fourDecimalRound(D.getValue(2)));
+
+}