@@ -138,6 +138,14 @@ namespace
138138 std::complex <double >,
139139 std::conditional_t <std::is_same_v<T, int8_t >, int32_t , T>>>;
140140
141+ /* **********************************************************************
142+ * Variable template to set scale type to F32 if the data type of BF16 *
143+ **********************************************************************/
144+ template <typename T>
145+ constexpr auto hipblaslt_scaletype = hipblaslt_datatype<T>;
146+
147+ template <>
148+ constexpr auto hipblaslt_scaletype<rocblas_bfloat16> = HIP_R_32F ;
141149 /* ***************************************************************
142150 * Construct a HipBlasLT GEMM from a RocblasContractionProblem *
143151 ****************************************************************/
@@ -543,7 +551,7 @@ rocblas_status runContractionProblemHipBlasLT(const RocblasContractionProblem<Ti
543551 sizeof (int64_t ))));
544552 hipblasLtMatmulDesc_t matmulDesc;
545553 RETURN_IF_ROCBLAS_ERROR (static_cast <rocblas_status>(
546- hipblasLtMatmulDescCreate (&matmulDesc, hipblaslt_compute_type<Tc>, hipblaslt_datatype <Ti>),
554+ hipblasLtMatmulDescCreate (&matmulDesc, hipblaslt_compute_type<Tc>, hipblaslt_scaletype <Ti>),
547555 HIPBLAS_STATUS_SUCCESS ));
548556 RETURN_IF_ROCBLAS_ERROR (static_cast <rocblas_status>(hipblasLtMatmulDescSetAttribute (
549557 matmulDesc, HIPBLASLT_MATMUL_DESC_TRANSA , &transA, sizeof (int32_t ))));
0 commit comments