title: CublasLt cublasLtMatmulAlgoGetHeuristic returns CUBLAS_STATUS_INVALID_VALUE for rows major matrix
body: <p>I've just finished to refactor my program to use <strong>cublasLt</strong> lib for GEMM and I fell into a <strong>CUBLAS_STATUS_INVALID_VALUE</strong> when executing <code>cublasLtMatmulAlgoGetHeuristic</code> in the function below.</p>
  
  <p><strong>CudaMatrix.cu:product</strong></p>
  
  <pre class="lang-cpp prettyprint-override"><code>/**
   * Performs the matrix-matrix multiplication C = A x B
   *
   * @see https://docs.nvidia.com/cuda/cublas/index.html#cublasLtMatmul
   *
   * @param A - The left matrix A
   * @param B - The right matrix B
   * @param C - The result matrix C
   * @param opA - Operation to perform on matrix A before multiplication (none, transpose or hermitian)
   * @param opB - Operation to perform on matrix B before multiplication (none, transpose or hermitian)
   * @param lightHandle - cublasLt handle
   */
  template&lt;typename precision&gt;
  void CudaMatrix&lt;precision&gt;::product(const CudaMatrix           &amp;A,
                                      const CudaMatrix           &amp;B,
                                            CudaMatrix           &amp;C,
                                            cublasOperation_t    opA,
                                            cublasOperation_t    opB,
                                            cublasLtHandle_t     lightHandle
  ) {
      const precision                 zero               = 0,
                                      one                = 1;
      const int                       requestedAlgoCount = 1;
      cudaStream_t                    stream             = nullptr;
      cublasLtMatmulHeuristicResult_t heuristicResult;
      cublasLtMatmulPreference_t      preference;
      cublasLtMatmulDesc_t            computeDesc;
      int                             returnedAlgoCount;
  
      // Set matrix pre-operation such as transpose if any
      cublasLtCk(cublasLtMatmulDescCreate(&amp;computeDesc, A.cublasLtDataType));
      cublasLtCk(cublasLtMatmulDescSetAttribute(computeDesc, CUBLASLT_MATMUL_DESC_TRANSA, &amp;opA, sizeof(opA)));
      cublasLtCk(cublasLtMatmulDescSetAttribute(computeDesc, CUBLASLT_MATMUL_DESC_TRANSB, &amp;opB, sizeof(opB)));
  
      // Get the best algorithm to use
      cublasLtCk(cublasLtMatmulPreferenceCreate(&amp;preference));
      cublasLtCk(cublasLtMatmulPreferenceSetAttribute(preference, CUBLASLT_MATMUL_PREF_MAX_WORKSPACE_BYTES,
                 &amp;CudaMatrix::matMulWorkspaceSize, sizeof(CudaMatrix::matMulWorkspaceSize)));
      cublasLtCk(cublasLtMatmulAlgoGetHeuristic(lightHandle, computeDesc, A.matrixLayout, B.matrixLayout,
                 C.matrixLayout, C.matrixLayout, preference, requestedAlgoCount, &amp;heuristicResult, &amp;returnedAlgoCount));
  
      std::cout &lt;&lt; "returnedAlgoCount = " &lt;&lt; returnedAlgoCount &lt;&lt; std::endl;
  
      // Do the multiplication
      cublasLtCk(cublasLtMatmul(lightHandle, computeDesc, &amp;one, A.data, A.matrixLayout, B.data, B.matrixLayout, &amp;zero,
                 C.data, C.matrixLayout, C.data, C.matrixLayout, &amp;heuristicResult.algo,
                 &amp;CudaMatrix::matMulWorkspace, CudaMatrix::matMulWorkspaceSize, stream));
  
      // clean up
      cublasLtCk(cublasLtMatmulPreferenceDestroy(preference));
      cublasLtCk(cublasLtMatmulDescDestroy(computeDesc));
  }
  </code></pre>
  
  <p>I concatenated a minimal reproducible example below with the same source code as I have in my program (with trims).</p>
  
  <p>This error may be related to a bug I found in <a href="https://devtalk.nvidia.com/default/topic/1049211/gpu-accelerated-libraries/where-can-i-find-working-examples-for-the-new-cublaslt-library-/3" rel="nofollow noreferrer">NVIDIA forum</a> but I am not sure.</p>
  
  <p>I'm running on Ubuntu 18.04 with RTX 5000 GPU.</p>
  
  <p><strong>cublaslt_mat_mul.cu</strong></p>
  
  <pre class="lang-cpp prettyprint-override"><code>#include &lt;iostream&gt;
  #include &lt;iomanip&gt;
  #include &lt;limits&gt;
  #include &lt;vector&gt;
  #include &lt;cxxabi.h&gt;
  #include &lt;cuda_runtime.h&gt;
  #include &lt;cuda_runtime_api.h&gt;
  #include &lt;cublasLt.h&gt;
  
  // ****************************************************************************************************************** //
  //                                                    ErrorsCheck.cuh                                                 //
  // ****************************************************************************************************************** //
  
  static const char* cublasGetErrorEnum(cublasStatus_t error)
  {
      switch (error)
      {
          case CUBLAS_STATUS_SUCCESS:
              return "CUBLAS_STATUS_SUCCESS";
  
          case CUBLAS_STATUS_NOT_INITIALIZED:
              return "CUBLAS_STATUS_NOT_INITIALIZED";
  
          case CUBLAS_STATUS_ALLOC_FAILED:
              return "CUBLAS_STATUS_ALLOC_FAILED";
  
          case CUBLAS_STATUS_INVALID_VALUE:
              return "CUBLAS_STATUS_INVALID_VALUE";
  
          case CUBLAS_STATUS_ARCH_MISMATCH:
              return "CUBLAS_STATUS_ARCH_MISMATCH";
  
          case CUBLAS_STATUS_MAPPING_ERROR:
              return "CUBLAS_STATUS_MAPPING_ERROR";
  
          case CUBLAS_STATUS_EXECUTION_FAILED:
              return "CUBLAS_STATUS_EXECUTION_FAILED";
  
          case CUBLAS_STATUS_INTERNAL_ERROR:
              return "CUBLAS_STATUS_INTERNAL_ERROR";
  
          case CUBLAS_STATUS_NOT_SUPPORTED:
              return "CUBLAS_STATUS_NOT_SUPPORTED";
  
          case CUBLAS_STATUS_LICENSE_ERROR:
              return "CUBLAS_STATUS_LICENSE_ERROR";
  
          default:
              return "&lt;unknown&gt;";
      }
  }
  
  inline void cublasLtCheck(cublasStatus_t status, int iLine, const char *szFile) {
      if (status != CUBLAS_STATUS_SUCCESS) {
          std::cerr &lt;&lt; "CublasLt error " &lt;&lt; cublasGetErrorEnum(status) &lt;&lt; " at line " &lt;&lt; iLine &lt;&lt; " in file "
                    &lt;&lt; szFile &lt;&lt; std::endl;
      }
  }
  
  inline void cudaCheck(cudaError_t status, int iLine, const char *szFile) {
      if (status != cudaSuccess) {
          std::cerr &lt;&lt; "CublasLt error " &lt;&lt; cudaGetErrorString(status) &lt;&lt; " at line " &lt;&lt; iLine &lt;&lt; " in file "
                    &lt;&lt; szFile &lt;&lt; std::endl;
      }
  }
  
  #define cublasLtCk(call) cublasLtCheck(call, __LINE__, __FILE__)
  #define cudaCk(call) cudaCheck(call, __LINE__, __FILE__)
  
  // ****************************************************************************************************************** //
  //                                                    CudaMatrix.cuh                                                  //
  // ****************************************************************************************************************** //
  
  #define MB 1048576 // 2^19 byte
  
  typedef unsigned int uint;
  
  template &lt;typename precision&gt;
  struct CudaMatrix {
      // Matrix multiplication GPU workspace that can be used to improve matrix multiplication computation time
      const static void   *matMulWorkspace;
      const static size_t matMulWorkspaceSize;
  
      CudaMatrix() : width(0), height(0), data(nullptr), cublasHandle(nullptr), cublasLtHandle(nullptr), matrixLayout(nullptr) { };
      CudaMatrix(uint width, uint height, cublasHandle_t cublasHandle = nullptr, cublasLtHandle_t cublasLtHandle = nullptr,
                 cublasLtMatrixLayout_t matrixLayout = nullptr) : width(width), height(height), cublasHandle(cublasHandle),
                 cublasLtHandle(cublasLtHandle), matrixLayout(matrixLayout)
      {
          cudaCk(cudaMalloc(&amp;data, bytesSize()));
  
          if (typeid(precision).hash_code() == typeid(uint).hash_code()) {
              cublasLtDataType = CUDA_R_8U;
          } else if (typeid(precision).hash_code() == typeid(int).hash_code()) {
              cublasLtDataType = CUDA_R_8I;
          } else if (typeid(precision).hash_code() == typeid(float).hash_code()) {
              cublasLtDataType = CUDA_R_32F;
          } else if (typeid(precision).hash_code() == typeid(double).hash_code()) {
              cublasLtDataType = CUDA_R_64F;
          } else {
              throw std::runtime_error("The datatype " + std::string(typeid(precision).name()) + " is not handled in CudaMatrix");
          }
  
          cublasLtCk(cublasLtMatrixLayoutCreate(&amp;matrixLayout, cublasLtDataType, height, width, width));
  
          if  (matMulWorkspace == nullptr) {
              cudaCk(cudaMalloc(&amp;matMulWorkspace, matMulWorkspaceSize));
          }
      }
  
      __device__ __host__ uint size() const { return width * height; }
  
      static void product(const CudaMatrix &amp;A, const CudaMatrix &amp;B, CudaMatrix &amp;C, cublasOperation_t opA, cublasOperation_t opB, cublasLtHandle_t lightHandle);
  
      void freeResources() { cudaCk(cudaFree(data)); cublasLtCk(cublasLtMatrixLayoutDestroy(matrixLayout)); }
      uint bytesSize() const { return size() * sizeof(precision); }
      void setValuesFromVector(const std::vector&lt;precision&gt; &amp;vector);
      void setValuesFromVector(const std::vector&lt;std::vector&lt;precision&gt;&gt; &amp;vectors);
      void display(const std::string &amp;name = "", uint x = 0, uint y = 0, uint roiWidth = 0, uint roiHeight = 0) const;
      void product(const CudaMatrix &amp;A) { product(*this, A, *this, CUBLAS_OP_N, CUBLAS_OP_N, cublasLtHandle); }
  
      precision              *data;
      uint                   width,
                             height;
      cublasHandle_t         cublasHandle;
      cublasLtHandle_t       cublasLtHandle;
      cublasLtMatrixLayout_t matrixLayout;
      cudaDataType_t         cublasLtDataType;
  };
  
  template &lt;typename precision&gt; const size_t CudaMatrix&lt;precision&gt;::matMulWorkspaceSize = 500 * MB;
  template &lt;typename precision&gt; const void*  CudaMatrix&lt;precision&gt;::matMulWorkspace     = nullptr;
  
  // ****************************************************************************************************************** //
  //                                                     CudaMatrix.cu                                                  //
  // ****************************************************************************************************************** //
  
  /**
   * Display the matrix
   *
   * @tparam precision - The matrix precision
   *
   * @param name - The matrix name
   */
  template &lt;typename precision&gt;
  void CudaMatrix&lt;precision&gt;::display(const std::string &amp;name, uint x, uint y, uint roiWidth, uint roiHeight) const
  {
      precision *hostValues;
  
      roiWidth == 0 ? roiWidth = width : roiWidth = roiWidth;
      roiHeight == 0 ? roiHeight = height : roiHeight = roiHeight;
  
      cudaCk(cudaMallocHost(&amp;hostValues, bytesSize()));
      cudaCk(cudaMemcpy(hostValues, data, bytesSize(), cudaMemcpyDeviceToHost));
  
      std::cout &lt;&lt; std::setprecision(std::numeric_limits&lt;precision&gt;::digits10 + 1);
  
      std::cout &lt;&lt; "Matrix " &lt;&lt; name &lt;&lt; " " &lt;&lt; width &lt;&lt; " x " &lt;&lt; height &lt;&lt; " pixels of "
                &lt;&lt; abi::__cxa_demangle(typeid(precision).name(), nullptr, nullptr, nullptr)
                &lt;&lt; "\n\n";
  
      for (int i = y; i &lt; y + roiHeight; ++i) {
          std::cout &lt;&lt; "{ ";
  
          for (int j = x; j &lt; x + roiWidth - 1; ++j) {
              std::cout &lt;&lt; *(hostValues + i * width + j) &lt;&lt; ", ";
          }
  
          std::cout &lt;&lt; *(hostValues + (i + 1) * width - 1) &lt;&lt; " }\n";
      }
  
      std::cout &lt;&lt; std::endl;
  
      cudaCk(cudaFreeHost(hostValues));
  }
  
  /**
   * Set the matrix values in device CUDA memory from a host standard 1D vector
   *
   * @tparam precision - The matrix precision
   *
   * @param vector - The values to set the device CUDA memory from
   */
  template &lt;typename precision&gt;
  void CudaMatrix&lt;precision&gt;::setValuesFromVector(const std::vector&lt;precision&gt; &amp;vector)
  {
      cudaCk(cudaMemcpy(data, vector.data(), vector.size() * sizeof(precision), cudaMemcpyHostToDevice));
  }
  
  /**
   * Set the matrix values in device CUDA memory from a host standard 2D vector
   *
   * @tparam precision - The matrix precision
   *
   * @param vectors - The values to set the device CUDA memory from
   */
  template &lt;typename precision&gt;
  void CudaMatrix&lt;precision&gt;::setValuesFromVector(const std::vector&lt;std::vector&lt;precision&gt;&gt; &amp;vectors)
  {
      std::vector&lt;precision&gt; buffer;
  
      buffer.reserve(vectors.size() * vectors[0].size());
  
      for (const auto &amp;vector : vectors) {
          buffer.insert(buffer.end(), vector.begin(), vector.end());
      }
  
      setValuesFromVector(buffer);
  }
  
  /**
   * Performs the matrix-matrix multiplication C = A x B
   *
   * @see https://docs.nvidia.com/cuda/cublas/index.html#cublasLtMatmul
   *
   * @param A - The left matrix A
   * @param B - The right matrix B
   * @param C - The result matrix C
   * @param opA - Operation to perform on matrix A before multiplication (none, transpose or hermitian)
   * @param opB - Operation to perform on matrix B before multiplication (none, transpose or hermitian)
   * @param lightHandle - cublasLt handle
   */
  template&lt;typename precision&gt;
  void CudaMatrix&lt;precision&gt;::product(const CudaMatrix           &amp;A,
                                      const CudaMatrix           &amp;B,
                                            CudaMatrix           &amp;C,
                                            cublasOperation_t    opA,
                                            cublasOperation_t    opB,
                                            cublasLtHandle_t     lightHandle
  ) {
      const precision                 zero               = 0,
                                      one                = 1;
      const int                       requestedAlgoCount = 1;
      cudaStream_t                    stream             = nullptr;
      cublasLtMatmulHeuristicResult_t heuristicResult;
      cublasLtMatmulPreference_t      preference;
      cublasLtMatmulDesc_t            computeDesc;
      int                             returnedAlgoCount;
  
      // Set matrix pre-operation such as transpose if any
      cublasLtCk(cublasLtMatmulDescCreate(&amp;computeDesc, A.cublasLtDataType));
      cublasLtCk(cublasLtMatmulDescSetAttribute(computeDesc, CUBLASLT_MATMUL_DESC_TRANSA, &amp;opA, sizeof(opA)));
      cublasLtCk(cublasLtMatmulDescSetAttribute(computeDesc, CUBLASLT_MATMUL_DESC_TRANSB, &amp;opB, sizeof(opB)));
  
      // Get the best algorithm to use
      cublasLtCk(cublasLtMatmulPreferenceCreate(&amp;preference));
      cublasLtCk(cublasLtMatmulPreferenceSetAttribute(preference, CUBLASLT_MATMUL_PREF_MAX_WORKSPACE_BYTES,
                 &amp;CudaMatrix::matMulWorkspaceSize, sizeof(CudaMatrix::matMulWorkspaceSize)));
      cublasLtCk(cublasLtMatmulAlgoGetHeuristic(lightHandle, computeDesc, A.matrixLayout, B.matrixLayout,
                 C.matrixLayout, C.matrixLayout, preference, requestedAlgoCount, &amp;heuristicResult, &amp;returnedAlgoCount));
  
      std::cout &lt;&lt; "returnedAlgoCount = " &lt;&lt; returnedAlgoCount &lt;&lt; std::endl;
  
      // Do the multiplication
      cublasLtCk(cublasLtMatmul(lightHandle, computeDesc, &amp;one, A.data, A.matrixLayout, B.data, B.matrixLayout, &amp;zero,
                 C.data, C.matrixLayout, C.data, C.matrixLayout, &amp;heuristicResult.algo,
                 &amp;CudaMatrix::matMulWorkspace, CudaMatrix::matMulWorkspaceSize, stream));
  
      // clean up
      cublasLtCk(cublasLtMatmulPreferenceDestroy(preference));
      cublasLtCk(cublasLtMatmulDescDestroy(computeDesc));
  }
  
  // Forward template declarations
  template struct CudaMatrix&lt;double&gt;;
  template struct CudaMatrix&lt;float&gt;;
  template struct CudaMatrix&lt;int&gt;;
  template struct CudaMatrix&lt;uint&gt;;
  
  // ****************************************************************************************************************** //
  //                                                        main.cu                                                     //
  // ****************************************************************************************************************** //
  
  int main(int argc, char const *argv[])
  {
      cublasLtHandle_t   cublasLtHandle = nullptr;
      std::vector&lt;float&gt; r1Expect       = { 6, 6, 6, 15, 15, 15, 24, 24, 24 };
      std::vector&lt;float&gt; r2Expect       = { 1, 2, 3, 4, 5, 6, 7, 8, 9 };
  
      cublasLtCk(cublasLtCreate(&amp;cublasLtHandle));
  
      // Declare matrices
      CudaMatrix&lt;float&gt; m1(3, 3);
      CudaMatrix&lt;float&gt; m2(3, 3);
      CudaMatrix&lt;float&gt; m3(3, 3);
      CudaMatrix&lt;float&gt; deviceResult(3, 3);
  
      // Set device memory values
      m1.setValuesFromVector({ {1, 1, 1}, {1, 1, 1}, {1, 1, 1} });
      m2.setValuesFromVector({ {1, 2, 3}, {4, 5, 6}, {7, 8, 9} });
      m3.setValuesFromVector({ {1, 0, 0}, {0, 1, 0}, {0, 0, 1} });
  
      // Test results (just showing it here)
      CudaMatrix&lt;float&gt;::product(m1, m2, deviceResult, CUBLAS_OP_N, CUBLAS_OP_N, cublasLtHandle);
  
      m1.display("m1");
      m2.display("m2");
      deviceResult.display("m1 X m2");
  
      CudaMatrix&lt;float&gt;::product(m2, m3, deviceResult, CUBLAS_OP_N, CUBLAS_OP_N, cublasLtHandle);
  
      m1.display("m2");
      m2.display("m3");
      deviceResult.display("m2 X m3");
  
      // Clean up
      cublasLtCk(cublasLtDestroy(cublasLtHandle));
  
      m1.freeResources();
      m2.freeResources();
      m3.freeResources();
      deviceResult.freeResources();
  
      return 0;
  }
  </code></pre>
  
  <p><strong>CMakeLists.txt</strong></p>
  
  <pre><code>cmake_minimum_required(VERSION 3.10)
  project(test-cuda)
  
  # ------------------------------------------------ Compilation options ----------------------------------------------- #
  
  # CUDA 10 does not support C++ 17
  set(CMAKE_CXX_STANDARD 14)
  set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -std=c++14")
  set(CMAKE_BUILD_TYPE Debug) # Release or Debug
  
  # Include CUDA
  find_package(CUDA REQUIRED)
  set(CUDA_NVCC_FLAGS "${CUDA_NVCC_FLAGS} -arch=sm_75 -std=c++14 --expt-relaxed-constexpr --expt-extended-lambda")
  
  # ----------------------------------------------------- Constants ---------------------------------------------------- #
  
  if (NOT ${CMAKE_BUILD_TYPE} STREQUAL "Release")
      MESSAGE(STATUS "Debug build")
      add_definitions(-DDEBUG_CUDA)
  else ()
      MESSAGE(STATUS "Release build")
      set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -O3")
      set(CUDA_NVCC_FLAGS "${CUDA_NVCC_FLAGS} -O3")
  endif ()
  
  # ------------------------------------------------- Source code files ------------------------------------------------ #
  
  # All in one
  file(GLOB matmul "cublaslt_mat_mul.cu")
  
  # ---------------------------------------------------- Executables --------------------------------------------------- #
  
  cuda_add_executable(matmulTest ${matmul})
  
  # ---------------------------------------------------- Libraries ----------------------------------------------------- #
  
  # Path to local libraries
  file(GLOB CUDAlibs "/usr/lib/x86_64-linux-gnu/libcuda.so" "/usr/lib/x86_64-linux-gnu/libcublas.so" "/usr/lib/x86_64-linux-gnu/libcublasLt.so" "/usr/local/cuda/lib64/libcudart.so")
  # Link libraries
  target_link_libraries(matmulTest ${CUDAlibs})
  </code></pre>
  
  <p><strong>Output</strong></p>
  
  <pre><code>CublasLt error CUBLAS_STATUS_INVALID_VALUE at line 249 in file /home/rom1/Desktop/test_cuda/cublaslt_mat_mul.cu
  returnedAlgoCount = -768202864
  CublasLt error CUBLAS_STATUS_INVALID_VALUE at line 256 in file /home/rom1/Desktop/test_cuda/cublaslt_mat_mul.cu
  Matrix m1 3 x 3 pixels of float
  
  { 1, 1, 1 }
  { 1, 1, 1 }
  { 1, 1, 1 }
  
  Matrix m2 3 x 3 pixels of float
  
  { 1, 2, 3 }
  { 4, 5, 6 }
  { 7, 8, 9 }
  
  Matrix m1 X m2 3 x 3 pixels of float
  
  { 0, 0, 0 }
  { 0, 0, 0 }
  { 0, 0, 0 }
  
  CublasLt error CUBLAS_STATUS_INVALID_VALUE at line 249 in file /home/rom1/Desktop/test_cuda/cublaslt_mat_mul.cu
  returnedAlgoCount = -870514560
  CublasLt error CUBLAS_STATUS_INVALID_VALUE at line 256 in file /home/rom1/Desktop/test_cuda/cublaslt_mat_mul.cu
  Matrix m2 3 x 3 pixels of float
  
  { 1, 1, 1 }
  { 1, 1, 1 }
  { 1, 1, 1 }
  
  Matrix m3 3 x 3 pixels of float
  
  { 1, 0, 0 }
  { 0, 1, 0 }
  { 0, 0, 1 }
  
  Matrix m2 X m3 3 x 3 pixels of float
  
  { 0, 0, 0 }
  { 0, 0, 0 }
  { 0, 0, 0 }
  </code></pre>