Merge 'master' into 'amd-master'
Change-Id: I02f86242eebb6f3035681ed074eb024b8556420e
[ROCm/hip commit: b4c7ec1f94]
Этот коммит содержится в:
Jenkins
@@ -241,11 +241,11 @@ if ($HIP_PLATFORM eq "clang") {
|
||||
}
|
||||
|
||||
} elsif ($HIP_PLATFORM eq "nvcc") {
|
||||
$CUDA_PATH=$ENV{'CUDA_PATH'} // '/usr/local/cuda';
|
||||
$HIP_INCLUDE_PATH = "$HIP_PATH/include";
|
||||
if ($verbose & 0x2) {
|
||||
print ("CUDA_PATH=$CUDA_PATH\n");
|
||||
}
|
||||
$CUDA_PATH=$ENV{'CUDA_PATH'} // '/usr/local/cuda';
|
||||
|
||||
$HIPCC="$CUDA_PATH/bin/nvcc";
|
||||
$HIPCXXFLAGS .= " -Wno-deprecated-gpu-targets ";
|
||||
|
||||
@@ -33,7 +33,7 @@ set(CMAKE_HIP_ARCHIVE_FINISH ${CMAKE_CXX_ARCHIVE_FINISH})
|
||||
set(CMAKE_SHARED_LIBRARY_SONAME_HIP_FLAG ${CMAKE_SHARED_LIBRARY_SONAME_CXX_FLAG})
|
||||
set(CMAKE_SHARED_LIBRARY_CREATE_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_CREATE_CXX_FLAGS})
|
||||
set(CMAKE_SHARED_LIBRARY_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_CXX_FLAGS})
|
||||
set(CMAKE_SHARED_LIBRARY_LINK_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_LINK_CXX_FLAGS})
|
||||
#set(CMAKE_SHARED_LIBRARY_LINK_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_LINK_CXX_FLAGS})
|
||||
set(CMAKE_SHARED_LIBRARY_RUNTIME_HIP_FLAG ${CMAKE_SHARED_LIBRARY_RUNTIME_CXX_FLAG})
|
||||
set(CMAKE_SHARED_LIBRARY_RUNTIME_HIP_FLAG_SEP ${CMAKE_SHARED_LIBRARY_RUNTIME_CXX_FLAG_SEP})
|
||||
set(CMAKE_SHARED_LIBRARY_LINK_STATIC_HIP_FLAGS ${CMAKE_SHARED_LIBRARY_LINK_STATIC_CXX_FLAGS})
|
||||
@@ -380,7 +380,7 @@ endmacro()
|
||||
###############################################################################
|
||||
macro(HIP_PREPARE_TARGET_COMMANDS _target _format _generated_files _source_files)
|
||||
set(_hip_flags "")
|
||||
set(_hip_build_configuration "${CMAKE_BUILD_TYPE}")
|
||||
string(TOUPPER _hip_build_configuration "${CMAKE_BUILD_TYPE}")
|
||||
if(HIP_HOST_COMPILATION_CPP)
|
||||
set(HIP_C_OR_CXX CXX)
|
||||
else()
|
||||
|
||||
@@ -0,0 +1,516 @@
|
||||
# CUBLAS API supported by HIP
|
||||
|
||||
## **1. CUBLAS Data types**
|
||||
|
||||
| **type** | **CUDA** | **HIP** |**HIP value** (if differs) |
|
||||
|-------------:|---------------------------------------------------------------|------------------------------------------------------------|---------------------------|
|
||||
| enum |***`cublasStatus`*** |***`hipblasStatus_t`*** |
|
||||
| enum |***`cublasStatus_t`*** |***`hipblasStatus_t`*** |
|
||||
| 0 |*`CUBLAS_STATUS_SUCCESS`* |*`HIPBLAS_STATUS_SUCCESS`* |
|
||||
| 1 |*`CUBLAS_STATUS_NOT_INITIALIZED`* |*`HIPBLAS_STATUS_NOT_INITIALIZED`* |
|
||||
| 3 |*`CUBLAS_STATUS_ALLOC_FAILED`* |*`HIPBLAS_STATUS_ALLOC_FAILED`* | 2 |
|
||||
| 7 |*`CUBLAS_STATUS_INVALID_VALUE`* |*`HIPBLAS_STATUS_INVALID_VALUE`* | 3 |
|
||||
| 8 |*`CUBLAS_STATUS_ARCH_MISMATCH`* |*`HIPBLAS_STATUS_ARCH_MISMATCH`* | |
|
||||
| 11 |*`CUBLAS_STATUS_MAPPING_ERROR`* |*`HIPBLAS_STATUS_MAPPING_ERROR`* | 4 |
|
||||
| 13 |*`CUBLAS_STATUS_EXECUTION_FAILED`* |*`HIPBLAS_STATUS_EXECUTION_FAILED`* | 5 |
|
||||
| 14 |*`CUBLAS_STATUS_INTERNAL_ERROR`* |*`HIPBLAS_STATUS_INTERNAL_ERROR`* | 6 |
|
||||
| 15 |*`CUBLAS_STATUS_NOT_SUPPORTED`* |*`HIPBLAS_STATUS_NOT_SUPPORTED`* | 7 |
|
||||
| 16 |*`CUBLAS_STATUS_LICENSE_ERROR`* | |
|
||||
| enum |***`cublasOperation_t`*** |***`hipblasOperation_t`*** |
|
||||
| 0 |*`CUBLAS_OP_N`* |*`HIPBLAS_OP_N`* | 111 |
|
||||
| 1 |*`CUBLAS_OP_T`* |*`HIPBLAS_OP_T`* | 112 |
|
||||
| 2 |*`CUBLAS_OP_C`* |*`HIPBLAS_OP_C`* | 113 |
|
||||
| enum |***`cublasFillMode_t`*** |***`hipblasFillMode_t`*** |
|
||||
| 0 |*`CUBLAS_FILL_MODE_LOWER`* |*`HIPBLAS_FILL_MODE_LOWER`* | 121 |
|
||||
| 1 |*`CUBLAS_FILL_MODE_UPPER`* |*`HIPBLAS_FILL_MODE_UPPER`* | 122 |
|
||||
| enum |***`cublasDiagType_t`*** |***`hipblasDiagType_t`*** |
|
||||
| 0 |*`CUBLAS_DIAG_NON_UNIT`* |*`HIPBLAS_DIAG_NON_UNIT`* | 131 |
|
||||
| 1 |*`CUBLAS_DIAG_UNIT`* |*`HIPBLAS_DIAG_UNIT`* | 132 |
|
||||
| enum |***`cublasSideMode_t`*** |***`hipblasSideMode_t`*** |
|
||||
| 0 |*`CUBLAS_SIDE_LEFT`* |*`HIPBLAS_SIDE_LEFT`* | 141 |
|
||||
| 1 |*`CUBLAS_SIDE_RIGHT`* |*`HIPBLAS_SIDE_RIGHT`* | 142 |
|
||||
| enum |***`cublasPointerMode_t`*** |***`hipblasPointerMode_t`*** |
|
||||
| 0 |*`CUBLAS_POINTER_MODE_HOST`* |*`HIPBLAS_POINTER_MODE_HOST`* |
|
||||
| 1 |*`CUBLAS_POINTER_MODE_DEVICE`* |*`HIPBLAS_POINTER_MODE_DEVICE`* |
|
||||
| enum |***`cublasAtomicsMode_t`*** | |
|
||||
| 0 |*`CUBLAS_ATOMICS_NOT_ALLOWED`* | |
|
||||
| 1 |*`CUBLAS_ATOMICS_ALLOWED`* | |
|
||||
| enum |***`cublasAtomicsMode_t`*** | |
|
||||
| -1 |*`CUBLAS_GEMM_DFALT`* | |
|
||||
| -1 |*`CUBLAS_GEMM_DEFAULT`* | |
|
||||
| 0 |*`CUBLAS_GEMM_ALGO0`* | |
|
||||
| 1 |*`CUBLAS_GEMM_ALGO1`* | |
|
||||
| 2 |*`CUBLAS_GEMM_ALGO2`* | |
|
||||
| 3 |*`CUBLAS_GEMM_ALGO3`* | |
|
||||
| 4 |*`CUBLAS_GEMM_ALGO4`* | |
|
||||
| 5 |*`CUBLAS_GEMM_ALGO5`* | |
|
||||
| 6 |*`CUBLAS_GEMM_ALGO6`* | |
|
||||
| 7 |*`CUBLAS_GEMM_ALGO7`* | |
|
||||
| 8 |*`CUBLAS_GEMM_ALGO8`* | |
|
||||
| 9 |*`CUBLAS_GEMM_ALGO9`* | |
|
||||
| 10 |*`CUBLAS_GEMM_ALGO10`* | |
|
||||
| 11 |*`CUBLAS_GEMM_ALGO11`* | |
|
||||
| 12 |*`CUBLAS_GEMM_ALGO12`* | |
|
||||
| 13 |*`CUBLAS_GEMM_ALGO13`* | |
|
||||
| 14 |*`CUBLAS_GEMM_ALGO14`* | |
|
||||
| 15 |*`CUBLAS_GEMM_ALGO15`* | |
|
||||
| 16 |*`CUBLAS_GEMM_ALGO16`* | |
|
||||
| 17 |*`CUBLAS_GEMM_ALGO17`* | |
|
||||
| 18 |*`CUBLAS_GEMM_ALGO18`* | |
|
||||
| 19 |*`CUBLAS_GEMM_ALGO19`* | |
|
||||
| 20 |*`CUBLAS_GEMM_ALGO20`* | |
|
||||
| 21 |*`CUBLAS_GEMM_ALGO21`* | |
|
||||
| 22 |*`CUBLAS_GEMM_ALGO22`* | |
|
||||
| 23 |*`CUBLAS_GEMM_ALGO23`* | |
|
||||
| 99 |*`CUBLAS_GEMM_DEFAULT_TENSOR_OP`* | |
|
||||
| 99 |*`CUBLAS_GEMM_DFALT_TENSOR_OP`* | |
|
||||
| 100 |*`CUBLAS_GEMM_ALGO0_TENSOR_OP`* | |
|
||||
| 101 |*`CUBLAS_GEMM_ALGO1_TENSOR_OP`* | |
|
||||
| 102 |*`CUBLAS_GEMM_ALGO2_TENSOR_OP`* | |
|
||||
| 103 |*`CUBLAS_GEMM_ALGO3_TENSOR_OP`* | |
|
||||
| 104 |*`CUBLAS_GEMM_ALGO4_TENSOR_OP`* | |
|
||||
| 105 |*`CUBLAS_GEMM_ALGO5_TENSOR_OP`* | |
|
||||
| 106 |*`CUBLAS_GEMM_ALGO6_TENSOR_OP`* | |
|
||||
| 107 |*`CUBLAS_GEMM_ALGO7_TENSOR_OP`* | |
|
||||
| 108 |*`CUBLAS_GEMM_ALGO8_TENSOR_OP`* | |
|
||||
| 109 |*`CUBLAS_GEMM_ALGO9_TENSOR_OP`* | |
|
||||
| 110 |*`CUBLAS_GEMM_ALGO10_TENSOR_OP`* | |
|
||||
| 111 |*`CUBLAS_GEMM_ALGO11_TENSOR_OP`* | |
|
||||
| 112 |*`CUBLAS_GEMM_ALGO12_TENSOR_OP`* | |
|
||||
| 113 |*`CUBLAS_GEMM_ALGO13_TENSOR_OP`* | |
|
||||
| 114 |*`CUBLAS_GEMM_ALGO14_TENSOR_OP`* | |
|
||||
| 115 |*`CUBLAS_GEMM_ALGO15_TENSOR_OP`* | |
|
||||
| enum |***`cublasMath_t`*** | |
|
||||
| 0 |*`CUBLAS_DEFAULT_MATH`* | |
|
||||
| 1 |*`CUBLAS_TENSOR_OP_MATH`* | |
|
||||
| enum* |`cublasDataType_t` | |
|
||||
| struct |`cublasContext` | |
|
||||
| struct* |`cublasHandle_t` |`hipblasHandle_t` |
|
||||
|
||||
## **2. CUBLAS API functions**
|
||||
|
||||
| **CUDA** | **HIP** |
|
||||
|-----------------------------------------------------------|-------------------------------------------------|
|
||||
|`cublasCreate` |`hipblasCreate` |
|
||||
|`cublasCreate_v2` |`hipblasCreate` |
|
||||
|`cublasDestroy` |`hipblasDestroy` |
|
||||
|`cublasDestroy_v2` |`hipblasDestroy` |
|
||||
|`cublasGetVersion` | |
|
||||
|`cublasGetVersion_v2` | |
|
||||
|`cublasGetProperty` | |
|
||||
|`cublasGetStream` |`hipblasGetStream` |
|
||||
|`cublasGetStream_v2` |`hipblasGetStream_v2` |
|
||||
|`cublasSetStream` |`hipblasSetStream` |
|
||||
|`cublasSetStream_v2` |`hipblasSetStream_v2` |
|
||||
|`cublasGetPointerMode` |`hipblasGetPointerMode` |
|
||||
|`cublasGetPointerMode_v2` |`hipblasGetPointerMode_v2` |
|
||||
|`cublasSetPointerMode` |`hipblasSetPointerMode` |
|
||||
|`cublasSetPointerMode_v2` |`hipblasSetPointerMode_v2` |
|
||||
|`cublasGetAtomicsMode` | |
|
||||
|`cublasSetAtomicsMode` | |
|
||||
|`cublasGetMathMode` | |
|
||||
|`cublasSetMathMode` | |
|
||||
|`cublasLogCallback` | |
|
||||
|`cublasLoggerConfigure` | |
|
||||
|`cublasSetLoggerCallback` | |
|
||||
|`cublasGetLoggerCallback` | |
|
||||
|`cublasSetVector` |`hipblasSetVector` |
|
||||
|`cublasGetVector` |`hipblasGetVector` |
|
||||
|`cublasSetMatrix` |`hipblasSetMatrix` |
|
||||
|`cublasGetMatrix` |`hipblasGetMatrix` |
|
||||
|`cublasSetVectorAsync` | |
|
||||
|`cublasGetVectorAsync` | |
|
||||
|`cublasSetMatrixAsync` | |
|
||||
|`cublasGetMatrixAsync` | |
|
||||
|`cublasXerbla` | |
|
||||
|`cublasNrm2Ex` | |
|
||||
|`cublasSnrm2` |`hipblasSnrm2` |
|
||||
|`cublasSnrm2_v2` |`hipblasSnrm2` |
|
||||
|`cublasDnrm2` |`hipblasDnrm2` |
|
||||
|`cublasDnrm2_v2` |`hipblasDnrm2` |
|
||||
|`cublasScnrm2` | |
|
||||
|`cublasScnrm2_v2` | |
|
||||
|`cublasDznrm2` | |
|
||||
|`cublasDznrm2_v2` | |
|
||||
|`cublasDotEx` | |
|
||||
|`cublasDotcEx` | |
|
||||
|`cublasSdot` |`hipblasSdot` |
|
||||
|`cublasSdot_v2` |`hipblasSdot` |
|
||||
|`cublasDdot` |`hipblasDdot` |
|
||||
|`cublasDdot_v2` |`hipblasDdot` |
|
||||
|`cublasCdotu` | |
|
||||
|`cublasCdotu_v2` | |
|
||||
|`cublasCdotc` | |
|
||||
|`cublasCdotc_v2` | |
|
||||
|`cublasZdotu` | |
|
||||
|`cublasZdotu_v2` | |
|
||||
|`cublasZdotc` | |
|
||||
|`cublasZdotc_v2` | |
|
||||
|`cublasScalEx` | |
|
||||
|`cublasSscal` |`hipblasSscal` |
|
||||
|`cublasSscal_v2` |`hipblasSscal` |
|
||||
|`cublasDscal` |`hipblasDscal` |
|
||||
|`cublasDscal_v2` |`hipblasDscal` |
|
||||
|`cublasCscal` | |
|
||||
|`cublasCscal_v2` | |
|
||||
|`cublasCsscal` | |
|
||||
|`cublasCsscal_v2` | |
|
||||
|`cublasZscal` | |
|
||||
|`cublasZscal_v2` | |
|
||||
|`cublasZdscal` | |
|
||||
|`cublasZdscal_v2` | |
|
||||
|`cublasAxpyEx` | |
|
||||
|`cublasSaxpy` |`hipblasSaxpy` |
|
||||
|`cublasSaxpy_v2` |`hipblasSaxpy` |
|
||||
|`cublasDaxpy` |`hipblasDaxpy` |
|
||||
|`cublasDaxpy_v2` |`hipblasDaxpy` |
|
||||
|`cublasCaxpy` | |
|
||||
|`cublasCaxpy_v2` | |
|
||||
|`cublasZaxpy` | |
|
||||
|`cublasZaxpy_v2` | |
|
||||
|`cublasScopy` |`hipblasScopy` |
|
||||
|`cublasScopy_v2` |`hipblasScopy` |
|
||||
|`cublasDcopy` |`hipblasDcopy` |
|
||||
|`cublasDcopy_v2` |`hipblasDcopy` |
|
||||
|`cublasCcopy` | |
|
||||
|`cublasCcopy_v2` | |
|
||||
|`cublasZcopy` | |
|
||||
|`cublasZcopy_v2` | |
|
||||
|`cublasSswap` | |
|
||||
|`cublasSswap_v2` | |
|
||||
|`cublasDswap` | |
|
||||
|`cublasDswap_v2` | |
|
||||
|`cublasCswap` | |
|
||||
|`cublasCswap_v2` | |
|
||||
|`cublasZswap` | |
|
||||
|`cublasZswap_v2` | |
|
||||
|`cublasIsamax` |`hipblasIsamax` |
|
||||
|`cublasIsamax_v2` |`hipblasIsamax` |
|
||||
|`cublasIdamax` |`hipblasIdamax` |
|
||||
|`cublasIdamax_v2` |`hipblasIdamax` |
|
||||
|`cublasIcamax` | |
|
||||
|`cublasIcamax_v2` | |
|
||||
|`cublasIzamax` | |
|
||||
|`cublasIzamax_v2` | |
|
||||
|`cublasIsamin` | |
|
||||
|`cublasIsamin_v2` | |
|
||||
|`cublasIdamin` | |
|
||||
|`cublasIdamin_v2` | |
|
||||
|`cublasIcamin` | |
|
||||
|`cublasIcamin_v2` | |
|
||||
|`cublasIzamin` | |
|
||||
|`cublasIzamin_v2` | |
|
||||
|`cublasSasum` |`hipblasSasum` |
|
||||
|`cublasSasum_v2` |`hipblasSasum` |
|
||||
|`cublasDasum` |`hipblasDasum` |
|
||||
|`cublasDasum_v2` |`hipblasDasum` |
|
||||
|`cublasScasum` | |
|
||||
|`cublasScasum_v2` | |
|
||||
|`cublasDzasum` | |
|
||||
|`cublasDzasum_v2` | |
|
||||
|`cublasSrot` | |
|
||||
|`cublasSrot_v2` | |
|
||||
|`cublasDrot` | |
|
||||
|`cublasDrot_v2` | |
|
||||
|`cublasCrot` | |
|
||||
|`cublasCrot_v2` | |
|
||||
|`cublasZrot` | |
|
||||
|`cublasZrot_v2` | |
|
||||
|`cublasZdrot` | |
|
||||
|`cublasZdrot_v2` | |
|
||||
|`cublasSrotg` | |
|
||||
|`cublasSrotg_v2` | |
|
||||
|`cublasDrotg` | |
|
||||
|`cublasDrotg_v2` | |
|
||||
|`cublasCrotg` | |
|
||||
|`cublasCrotg_v2` | |
|
||||
|`cublasZrotg` | |
|
||||
|`cublasZrotg_v2` | |
|
||||
|`cublasSrotm` | |
|
||||
|`cublasSrotm_v2` | |
|
||||
|`cublasDrotm` | |
|
||||
|`cublasDrotm_v2` | |
|
||||
|`cublasSrotmg` | |
|
||||
|`cublasSrotmg_v2` | |
|
||||
|`cublasDrotmg` | |
|
||||
|`cublasDrotmg_v2` | |
|
||||
|`cublasSgemv` |`hipblasSgemv` |
|
||||
|`cublasSgemv_v2` |`hipblasSgemv` |
|
||||
|`cublasDgemv` |`hipblasDgemv` |
|
||||
|`cublasDgemv_v2` |`hipblasDgemv` |
|
||||
|`cublasCgemv` | |
|
||||
|`cublasCgemv_v2` | |
|
||||
|`cublasZgemv` | |
|
||||
|`cublasZgemv_v2` | |
|
||||
|`cublasSgbmv` | |
|
||||
|`cublasSgbmv_v2` | |
|
||||
|`cublasDgbmv` | |
|
||||
|`cublasDgbmv_v2` | |
|
||||
|`cublasCgbmv` | |
|
||||
|`cublasCgbmv_v2` | |
|
||||
|`cublasZgbmv` | |
|
||||
|`cublasZgbmv_v2` | |
|
||||
|`cublasStrmv` | |
|
||||
|`cublasStrmv_v2` | |
|
||||
|`cublasDtrmv` | |
|
||||
|`cublasDtrmv_v2` | |
|
||||
|`cublasCtrmv` | |
|
||||
|`cublasCtrmv_v2` | |
|
||||
|`cublasZtrmv` | |
|
||||
|`cublasZtrmv_v2` | |
|
||||
|`cublasStbmv` | |
|
||||
|`cublasStbmv_v2` | |
|
||||
|`cublasDtbmv` | |
|
||||
|`cublasDtbmv_v2` | |
|
||||
|`cublasCtbmv` | |
|
||||
|`cublasCtbmv_v2` | |
|
||||
|`cublasZtbmv` | |
|
||||
|`cublasZtbmv_v2` | |
|
||||
|`cublasStpmv` | |
|
||||
|`cublasStpmv_v2` | |
|
||||
|`cublasDtpmv` | |
|
||||
|`cublasDtpmv_v2` | |
|
||||
|`cublasCtpmv` | |
|
||||
|`cublasCtpmv_v2` | |
|
||||
|`cublasZtpmv` | |
|
||||
|`cublasZtpmv_v2` | |
|
||||
|`cublasStrsv` | |
|
||||
|`cublasStrsv_v2` | |
|
||||
|`cublasDtrsv` | |
|
||||
|`cublasDtrsv_v2` | |
|
||||
|`cublasCtrsv` | |
|
||||
|`cublasCtrsv_v2` | |
|
||||
|`cublasZtrsv` | |
|
||||
|`cublasZtrsv_v2` | |
|
||||
|`cublasStpsv` | |
|
||||
|`cublasStpsv_v2` | |
|
||||
|`cublasDtpsv` | |
|
||||
|`cublasDtpsv_v2` | |
|
||||
|`cublasCtpsv` | |
|
||||
|`cublasCtpsv_v2` | |
|
||||
|`cublasZtpsv` | |
|
||||
|`cublasZtpsv_v2` | |
|
||||
|`cublasStbsv` | |
|
||||
|`cublasStbsv_v2` | |
|
||||
|`cublasDtbsv` | |
|
||||
|`cublasDtbsv_v2` | |
|
||||
|`cublasCtbsv` | |
|
||||
|`cublasCtbsv_v2` | |
|
||||
|`cublasZtbsv` | |
|
||||
|`cublasZtbsv_v2` | |
|
||||
|`cublasSsymv` | |
|
||||
|`cublasSsymv_v2` | |
|
||||
|`cublasDsymv` | |
|
||||
|`cublasDsymv_v2` | |
|
||||
|`cublasCsymv` | |
|
||||
|`cublasCsymv_v2` | |
|
||||
|`cublasZsymv` | |
|
||||
|`cublasZsymv_v2` | |
|
||||
|`cublasChemv` | |
|
||||
|`cublasChemv_v2` | |
|
||||
|`cublasZhemv` | |
|
||||
|`cublasZhemv_v2` | |
|
||||
|`cublasSsbmv` | |
|
||||
|`cublasSsbmv_v2` | |
|
||||
|`cublasDsbmv` | |
|
||||
|`cublasDsbmv_v2` | |
|
||||
|`cublasChbmv` | |
|
||||
|`cublasChbmv_v2` | |
|
||||
|`cublasZhbmv` | |
|
||||
|`cublasZhbmv_v2` | |
|
||||
|`cublasSspmv` | |
|
||||
|`cublasSspmv_v2` | |
|
||||
|`cublasDspmv` | |
|
||||
|`cublasDspmv_v2` | |
|
||||
|`cublasChpmv` | |
|
||||
|`cublasChpmv_v2` | |
|
||||
|`cublasZhpmv` | |
|
||||
|`cublasZhpmv_v2` | |
|
||||
|`cublasSger` |`hipblasSger` |
|
||||
|`cublasSger_v2` |`hipblasSger` |
|
||||
|`cublasDger` |`hipblasDger` |
|
||||
|`cublasDger_v2` |`hipblasDger` |
|
||||
|`cublasCgeru` | |
|
||||
|`cublasCgeru_v2` | |
|
||||
|`cublasCgerc` | |
|
||||
|`cublasCgerc_v2` | |
|
||||
|`cublasZgeru` | |
|
||||
|`cublasZgeru_v2` | |
|
||||
|`cublasZgerc` | |
|
||||
|`cublasZgerc_v2` | |
|
||||
|`cublasSsyr` | |
|
||||
|`cublasSsyr_v2` | |
|
||||
|`cublasDsyr` | |
|
||||
|`cublasDsyr_v2` | |
|
||||
|`cublasCsyr` | |
|
||||
|`cublasCsyr_v2` | |
|
||||
|`cublasZsyr` | |
|
||||
|`cublasZsyr_v2` | |
|
||||
|`cublasCher` | |
|
||||
|`cublasCher_v2` | |
|
||||
|`cublasZher` | |
|
||||
|`cublasZher_v2` | |
|
||||
|`cublasSspr` | |
|
||||
|`cublasSspr_v2` | |
|
||||
|`cublasDspr` | |
|
||||
|`cublasDspr_v2` | |
|
||||
|`cublasChpr` | |
|
||||
|`cublasChpr_v2` | |
|
||||
|`cublasZhpr` | |
|
||||
|`cublasZhpr_v2` | |
|
||||
|`cublasSsyr2` | |
|
||||
|`cublasSsyr2_v2` | |
|
||||
|`cublasDsyr2` | |
|
||||
|`cublasDsyr2_v2` | |
|
||||
|`cublasCsyr2` | |
|
||||
|`cublasCsyr2_v2` | |
|
||||
|`cublasZsyr2` | |
|
||||
|`cublasZsyr2_v2` | |
|
||||
|`cublasCher2` | |
|
||||
|`cublasCher2_v2` | |
|
||||
|`cublasZher2` | |
|
||||
|`cublasZher2_v2` | |
|
||||
|`cublasSspr2` | |
|
||||
|`cublasSspr2_v2` | |
|
||||
|`cublasDspr2` | |
|
||||
|`cublasDspr2_v2` | |
|
||||
|`cublasChpr2` | |
|
||||
|`cublasChpr2_v2` | |
|
||||
|`cublasZhpr2` | |
|
||||
|`cublasZhpr2_v2` | |
|
||||
|`cublasSgemm` |`hipblasSgemm` |
|
||||
|`cublasSgemm_v2` |`hipblasSgemm` |
|
||||
|`cublasDgemm` |`hipblasDgemm` |
|
||||
|`cublasDgemm_v2` |`hipblasDgemm` |
|
||||
|`cublasCgemm` |`hipblasCgemm` |
|
||||
|`cublasCgemm_v2` |`hipblasCgemm` |
|
||||
|`cublasCgemm3m` | |
|
||||
|`cublasCgemm3mEx` | |
|
||||
|`cublasZgemm` | |
|
||||
|`cublasZgemm_v2` | |
|
||||
|`cublasZgemm3m` | |
|
||||
|`cublasHgemm` |`hipblasHgemm` |
|
||||
|`cublasSgemmEx` | |
|
||||
|`cublasGemmEx` | |
|
||||
|`cublasCgemmEx` | |
|
||||
|`cublasUint8gemmBias` | |
|
||||
|`cublasSsyrk` | |
|
||||
|`cublasSsyrk_v2` | |
|
||||
|`cublasDsyrk` | |
|
||||
|`cublasDsyrk_v2` | |
|
||||
|`cublasCsyrk` | |
|
||||
|`cublasCsyrk_v2` | |
|
||||
|`cublasZsyrk` | |
|
||||
|`cublasZsyrk_v2` | |
|
||||
|`cublasCsyrkEx` | |
|
||||
|`cublasCsyrk3mEx` | |
|
||||
|`cublasCherk` | |
|
||||
|`cublasCherk_v2` | |
|
||||
|`cublasZherk` | |
|
||||
|`cublasZherk_v2` | |
|
||||
|`cublasCherkEx` | |
|
||||
|`cublasCherk3mEx` | |
|
||||
|`cublasSsyr2k` | |
|
||||
|`cublasSsyr2k_v2` | |
|
||||
|`cublasDsyr2k` | |
|
||||
|`cublasDsyr2k_v2` | |
|
||||
|`cublasCsyr2k` | |
|
||||
|`cublasCsyr2k_v2` | |
|
||||
|`cublasZsyr2k` | |
|
||||
|`cublasZsyr2k_v2` | |
|
||||
|`cublasCher2k` | |
|
||||
|`cublasCher2k_v2` | |
|
||||
|`cublasZher2k` | |
|
||||
|`cublasZher2k_v2` | |
|
||||
|`cublasSsyrkx` | |
|
||||
|`cublasDsyrkx` | |
|
||||
|`cublasCsyrkx` | |
|
||||
|`cublasZsyrkx` | |
|
||||
|`cublasCherkx` | |
|
||||
|`cublasZherkx` | |
|
||||
|`cublasSsymm` | |
|
||||
|`cublasSsymm_v2` | |
|
||||
|`cublasDsymm` | |
|
||||
|`cublasDsymm_v2` | |
|
||||
|`cublasCsymm` | |
|
||||
|`cublasCsymm_v2` | |
|
||||
|`cublasZsymm` | |
|
||||
|`cublasZsymm_v2` | |
|
||||
|`cublasChemm` | |
|
||||
|`cublasChemm_v2` | |
|
||||
|`cublasZhemm` | |
|
||||
|`cublasZhemm_v2` | |
|
||||
|`cublasStrsm` |`hipblasStrsm` |
|
||||
|`cublasStrsm_v2` |`hipblasStrsm` |
|
||||
|`cublasDtrsm` |`hipblasDtrsm` |
|
||||
|`cublasDtrsm_v2` |`hipblasDtrsm` |
|
||||
|`cublasCtrsm` | |
|
||||
|`cublasCtrsm_v2` | |
|
||||
|`cublasZtrsm` | |
|
||||
|`cublasZtrsm_v2` | |
|
||||
|`cublasStrmm` | |
|
||||
|`cublasStrmm_v2` | |
|
||||
|`cublasDtrmm` | |
|
||||
|`cublasDtrmm_v2` | |
|
||||
|`cublasCtrmm` | |
|
||||
|`cublasCtrmm_v2` | |
|
||||
|`cublasZtrmm` | |
|
||||
|`cublasZtrmm_v2` | |
|
||||
|`cublasHgemmBatched` | |
|
||||
|`cublasSgemmBatched` |`hipblasSgemmBatched` |
|
||||
|`cublasDgemmBatched` |`hipblasDgemmBatched` |
|
||||
|`cublasCgemmBatched` | |
|
||||
|`cublasCgemm3mBatched` | |
|
||||
|`cublasZgemmBatched` | |
|
||||
|`cublasGemmBatchedEx` | |
|
||||
|`cublasGemmStridedBatchedEx` | |
|
||||
|`cublasSgemmStridedBatched` | |
|
||||
|`cublasDgemmStridedBatched` | |
|
||||
|`cublasCgemmStridedBatched` | |
|
||||
|`cublasCgemm3mStridedBatched` | |
|
||||
|`cublasZgemmStridedBatched` | |
|
||||
|`cublasHgemmStridedBatched` | |
|
||||
|`cublasSgeam` |`hipblasSgeam` |
|
||||
|`cublasDgeam` |`hipblasDgeam` |
|
||||
|`cublasCgeam` | |
|
||||
|`cublasZgeam` | |
|
||||
|`cublasSgetrfBatched` | |
|
||||
|`cublasDgetrfBatched` | |
|
||||
|`cublasCgetrfBatched` | |
|
||||
|`cublasZgetrfBatched` | |
|
||||
|`cublasSgetriBatched` | |
|
||||
|`cublasDgetriBatched` | |
|
||||
|`cublasCgetriBatched` | |
|
||||
|`cublasZgetriBatched` | |
|
||||
|`cublasSgetrsBatched` | |
|
||||
|`cublasDgetrsBatched` | |
|
||||
|`cublasCgetrsBatched` | |
|
||||
|`cublasZgetrsBatched` | |
|
||||
|`cublasStrsmBatched` | |
|
||||
|`cublasDtrsmBatched` | |
|
||||
|`cublasCtrsmBatched` | |
|
||||
|`cublasZtrsmBatched` | |
|
||||
|`cublasSmatinvBatched` | |
|
||||
|`cublasDmatinvBatched` | |
|
||||
|`cublasCmatinvBatched` | |
|
||||
|`cublasZmatinvBatched` | |
|
||||
|`cublasSgeqrfBatched` | |
|
||||
|`cublasDgeqrfBatched` | |
|
||||
|`cublasCgeqrfBatched` | |
|
||||
|`cublasZgeqrfBatched` | |
|
||||
|`cublasSgelsBatched` | |
|
||||
|`cublasDgelsBatched` | |
|
||||
|`cublasCgelsBatched` | |
|
||||
|`cublasZgelsBatched` | |
|
||||
|`cublasSdgmm` | |
|
||||
|`cublasDdgmm` | |
|
||||
|`cublasCdgmm` | |
|
||||
|`cublasZdgmm` | |
|
||||
|`cublasStpttr` | |
|
||||
|`cublasDtpttr` | |
|
||||
|`cublasCtpttr` | |
|
||||
|`cublasZtpttr` | |
|
||||
|`cublasStrttp` | |
|
||||
|`cublasDtrttp` | |
|
||||
|`cublasCtrttp` | |
|
||||
|`cublasZtrttp` | |
|
||||
@@ -0,0 +1,81 @@
|
||||
# CUFFT API supported by HIP
|
||||
|
||||
## **1. CUFFT Data types**
|
||||
|
||||
| **type** | **CUDA** | **HIP** |**HIP value** (if differs) |
|
||||
|-------------:|---------------------------------------------------------------|------------------------------------------------------------|---------------------------|
|
||||
| enum |***`cufftResult_t`*** |***`hipfftResult_t`*** |
|
||||
| enum |***`cufftResult`*** |***`hipfftResult`*** |
|
||||
| 0x0 |*`CUFFT_SUCCESS`* |*`HIPFFT_SUCCESS`* | 0 |
|
||||
| 0x1 |*`CUFFT_INVALID_PLAN`* |*`HIPFFT_INVALID_PLAN`* | 1 |
|
||||
| 0x2 |*`CUFFT_ALLOC_FAILED`* |*`HIPFFT_ALLOC_FAILED`* | 2 |
|
||||
| 0x3 |*`CUFFT_INVALID_TYPE`* |*`HIPFFT_INVALID_TYPE`* | 3 |
|
||||
| 0x4 |*`CUFFT_INVALID_VALUE`* |*`HIPFFT_INVALID_VALUE`* | 4 |
|
||||
| 0x5 |*`CUFFT_INTERNAL_ERROR`* |*`HIPFFT_INTERNAL_ERROR`* | 5 |
|
||||
| 0x6 |*`CUFFT_EXEC_FAILED`* |*`HIPFFT_EXEC_FAILED`* | 6 |
|
||||
| 0x7 |*`CUFFT_SETUP_FAILED`* |*`HIPFFT_SETUP_FAILED`* | 7 |
|
||||
| 0x8 |*`CUFFT_I;NVALID_SIZE`* |*`HIPFFT_INVALID_SIZE`* | 8 |
|
||||
| 0x9 |*`CUFFT_UNALIGNED_DATA`* |*`HIPFFT_UNALIGNED_DATA`* | 9 |
|
||||
| 0xA |*`CUFFT_INCOMPLETE_PARAMETER_LIST`* |*`HIPFFT_INCOMPLETE_PARAMETER_LIST`* | 10 |
|
||||
| 0xB |*`CUFFT_INVALID_DEVICE`* |*`HIPFFT_INVALID_DEVICE`* | 11 |
|
||||
| 0xC |*`CUFFT_PARSE_ERROR`* |*`HIPFFT_PARSE_ERROR`* | 12 |
|
||||
| 0xD |*`CUFFT_NO_WORKSPACE`* |*`HIPFFT_NO_WORKSPACE`* | 13 |
|
||||
| 0xE |*`CUFFT_NOT_IMPLEMENTED`* |*`HIPFFT_NOT_IMPLEMENTED`* | 14 |
|
||||
| 0xF |*`CUFFT_LICENSE_ERROR`* | |
|
||||
| 0x10 |*`CUFFT_NOT_SUPPORTED`* |*`HIPFFT_NOT_SUPPORTED`* | 16 |
|
||||
| float |***`cufftReal`*** |***`hipfftReal`*** |
|
||||
| double |***`cufftDoubleReal`*** |***`hipfftDoubleReal`*** |
|
||||
| float2 |***`cufftComplex`*** |***`hipfftComplex`*** |
|
||||
| double2 |***`cufftDoubleComplex`*** |***`hipfftDoubleComplex`*** |
|
||||
| define |`CUFFT_FORWARD` |`HIPFFT_FORWARD` |
|
||||
| define |`CUFFT_INVERSE` |`HIPFFT_BACKWARD` |
|
||||
| enum |***`cufftType_t`*** |***`hipfftType_t`*** |
|
||||
| enum |***`cufftType`*** |***`hipfftType`*** |
|
||||
| 0x2a |*`CUFFT_R2C`* |*`HIPFFT_R2C`* |
|
||||
| 0x2c |*`CUFFT_C2R`* |*`HIPFFT_C2R`* |
|
||||
| 0x29 |*`CUFFT_C2C`* |*`HIPFFT_C2C`* |
|
||||
| 0x6a |*`CUFFT_D2Z`* |*`HIPFFT_D2Z`* |
|
||||
| 0x6c |*`CUFFT_Z2D`* |*`HIPFFT_Z2D`* |
|
||||
| 0x69 |*`CUFFT_Z2Z`* |*`HIPFFT_Z2Z`* |
|
||||
| enum |***`cufftCompatibility_t`*** | |
|
||||
| enum |***`cufftCompatibility`*** | |
|
||||
| 0x01 |*`CUFFT_COMPATIBILITY_FFTW_PADDING`* | |
|
||||
| define |`CUFFT_COMPATIBILITY_DEFAULT` | |
|
||||
| int |***`cufftHandle`*** |***`hipfftHandle`*** |
|
||||
|
||||
## **2. CUFFT API functions**
|
||||
|
||||
| **CUDA** | **HIP** |
|
||||
|-----------------------------------------------------------|-------------------------------------------------|
|
||||
|`cufftPlan1d` |`hipfftPlan1d` |
|
||||
|`cufftPlan2d` |`hipfftPlan2d` |
|
||||
|`cufftPlan3d` |`hipfftPlan3d` |
|
||||
|`cufftPlanMany` |`hipfftPlanMany` |
|
||||
|`cufftMakePlan1d` |`hipfftMakePlan1d` |
|
||||
|`cufftMakePlan2d` |`hipfftMakePlan2d` |
|
||||
|`cufftMakePlan3d` |`hipfftMakePlan3d` |
|
||||
|`cufftMakePlanMany` |`hipfftMakePlanMany` |
|
||||
|`cufftMakePlanMany64` |`hipfftMakePlanMany64` |
|
||||
|`cufftGetSizeMany64` |`hipfftGetSizeMany64` |
|
||||
|`cufftEstimate1d` |`hipfftEstimate1d` |
|
||||
|`cufftEstimate2d` |`hipfftEstimate2d` |
|
||||
|`cufftEstimate3d` |`hipfftEstimate3d` |
|
||||
|`cufftEstimateMany` |`hipfftEstimateMany` |
|
||||
|`cufftCreate` |`hipfftCreate` |
|
||||
|`cufftGetSize1d` |`hipfftGetSize1d` |
|
||||
|`cufftGetSize2d` |`hipfftGetSize2d` |
|
||||
|`cufftGetSize3d` |`hipfftGetSize3d` |
|
||||
|`cufftGetSizeMany` |`hipfftGetSizeMany` |
|
||||
|`cufftGetSize` |`hipfftGetSize` |
|
||||
|`cufftSetWorkArea` |`hipfftSetWorkArea` |
|
||||
|`cufftSetAutoAllocation` |`hipfftSetAutoAllocation` |
|
||||
|`cufftExecC2C` |`hipfftExecC2C` |
|
||||
|`cufftExecR2C` |`hipfftExecR2C` |
|
||||
|`cufftExecC2R` |`hipfftExecC2R` |
|
||||
|`cufftExecZ2Z` |`hipfftExecZ2Z` |
|
||||
|`cufftExecD2Z` |`hipfftExecD2Z` |
|
||||
|`cufftExecZ2D` |`hipfftExecZ2D` |
|
||||
|`cufftSetStream` |`hipfftSetStream` |
|
||||
|`cufftDestroy` |`hipfftDestroy` |
|
||||
|`cufftGetVersion` |`hipfftGetVersion` |
|
||||
|`cufftGetProperty` | |
|
||||
@@ -6,19 +6,30 @@
|
||||
|
||||
<!-- toc -->
|
||||
|
||||
- [Using hipify-clang](#using-hipify-clang)
|
||||
* [Build and install](#build-and-install)
|
||||
* [Running and using hipify-clang](#running-and-using-hipify-clang)
|
||||
- [Supported CUDA APIs](#cuda-apis)
|
||||
- [Dependencies](#dependencies)
|
||||
- [Build and install](#build-and-install)
|
||||
* [Building](#building)
|
||||
* [Testing](#testing)
|
||||
* [Windows](#windows)
|
||||
- [Running and using hipify-clang](#running-and-using-hipify-clang)
|
||||
- [Disclaimer](#disclaimer)
|
||||
|
||||
<!-- tocstop -->
|
||||
|
||||
## Build and install
|
||||
## <a name="cuda-apis"></a> Supported CUDA APIs
|
||||
|
||||
### Dependencies
|
||||
- [Runtime API](../docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md)
|
||||
- [Driver API](../docs/markdown/CUDA_Driver_API_functions_supported_by_HIP.md)
|
||||
- [cuBLAS](../docs/markdown/CUBLAS_API_supported_by_HIP.md)
|
||||
- [cuRAND](../docs/markdown/CURAND_API_supported_by_HIP.md)
|
||||
- [cuDNN](../docs/markdown/CUDNN_API_supported_by_HIP.md)
|
||||
- [cuFFT](../docs/markdown/CUFFT_API_supported_by_HIP.md)
|
||||
|
||||
## <a name="dependencies"></a> Dependencies
|
||||
|
||||
`hipify-clang` requires:
|
||||
1. LLVM+CLANG of at least version 3.8.0, latest stable release is 6.0.0.
|
||||
1. LLVM+CLANG of at least version 3.8.0, latest stable and recommended release is 6.0.1.
|
||||
2. CUDA at least version 7.5, latest supported release is 9.0.
|
||||
|
||||
| **LLVM release version** | **CUDA latest supported version** |
|
||||
@@ -38,10 +49,11 @@
|
||||
In most cases, you can get a suitable version of LLVM+CLANG with your package manager.
|
||||
|
||||
Failing that or having multiple versions of LLVM, you can [download a release archive](http://releases.llvm.org/), build or install it, and set
|
||||
[CMAKE_PREFIX_PATH](https://cmake.org/cmake/help/v3.0/variable/CMAKE_PREFIX_PATH.html) so `cmake` can find it; for instance: `-DCMAKE_PREFIX_PATH=f:\LLVM\6.0.0\dist`
|
||||
[CMAKE_PREFIX_PATH](https://cmake.org/cmake/help/v3.10/variable/CMAKE_PREFIX_PATH.html) so `cmake` can find it; for instance: `-DCMAKE_PREFIX_PATH=f:\LLVM\6.0.1\dist`
|
||||
|
||||
### Build
|
||||
## <a name="build-and-install"></a> Build and install
|
||||
|
||||
### <a name="building"></a> Build
|
||||
|
||||
Assuming this repository is at `./HIP`:
|
||||
|
||||
@@ -64,14 +76,14 @@ Debug build type `-DCMAKE_BUILD_TYPE=Debug` is also supported and tested, `LLVM+
|
||||
|
||||
The binary can then be found at `./dist/bin/hipify-clang`.
|
||||
|
||||
### Test
|
||||
### <a name="testing"></a> Test
|
||||
|
||||
`hipify-clang` has unit tests using LLVM [`lit`](https://llvm.org/docs/CommandGuide/lit.html)/[`FileCheck`](https://llvm.org/docs/CommandGuide/FileCheck.html).
|
||||
|
||||
**LLVM+CLANG should be built from sources, Pre-Built Binaries are not exhaustive for testing.**
|
||||
|
||||
To run it:
|
||||
1. Download [`LLVM`](http://releases.llvm.org/6.0.0/llvm-6.0.0.src.tar.xz)+[`CLANG`](http://releases.llvm.org/6.0.0/cfe-6.0.0.src.tar.xz) sources.
|
||||
1. Download [`LLVM`](http://releases.llvm.org/6.0.1/llvm-6.0.1.src.tar.xz)+[`CLANG`](http://releases.llvm.org/6.0.1/cfe-6.0.1.src.tar.xz) sources.
|
||||
2. Build [`LLVM+CLANG`](http://llvm.org/docs/CMake.html).
|
||||
For instance:
|
||||
```shell
|
||||
@@ -110,21 +122,21 @@ To run it:
|
||||
6. Ensure `lit` and `FileCheck` are installed - these are distributed with LLVM.
|
||||
* installing `lit` into `python` might be required:
|
||||
|
||||
`python f:/LLVM/6.0.0/llvm/utils/lit/setup.py install`,
|
||||
`python f:/LLVM/6.0.1/llvm/utils/lit/setup.py install`,
|
||||
|
||||
where `f:/LLVM/6.0.0/llvm` is LLVM sources root directory.
|
||||
where `f:/LLVM/6.0.1/llvm` is LLVM sources root directory.
|
||||
|
||||
* Starting with LLVM 6.0.0 path to llvm-lit.py script should be specified by the `LLVM_EXTERNAL_LIT` option:
|
||||
* Starting with LLVM 6.0.1 path to llvm-lit.py script should be specified by the `LLVM_EXTERNAL_LIT` option:
|
||||
|
||||
`-DLLVM_EXTERNAL_LIT=f:/LLVM/6.0.0/build/Release/bin/llvm-lit.py`,
|
||||
`-DLLVM_EXTERNAL_LIT=f:/LLVM/6.0.1/build/Release/bin/llvm-lit.py`,
|
||||
|
||||
where `f:/LLVM/6.0.0/build/Release` is LLVM build directory.
|
||||
where `f:/LLVM/6.0.1/build/Release` is LLVM build directory.
|
||||
7. Build with the `HIPIFY_CLANG_TESTS` option turned on: -DHIPIFY_CLANG_TESTS=1.
|
||||
8. `make test-hipify`
|
||||
|
||||
On Windows after `cmake` the project `test-hipify` in the generated `hipify-clang.sln` should be built by `Visual Studio 15 2017` instead of `make test-hipify`.
|
||||
|
||||
### Windows
|
||||
### <a name="windows"></a >Windows
|
||||
|
||||
On Windows the following configurations are tested:
|
||||
|
||||
@@ -144,30 +156,30 @@ cmake
|
||||
-DHIPIFY_CLANG_TESTS=1 \
|
||||
-DCMAKE_BUILD_TYPE=Release \
|
||||
-DCMAKE_INSTALL_PREFIX=../dist \
|
||||
-DCMAKE_PREFIX_PATH=f:/LLVM/6.0.0/dist \
|
||||
-DCMAKE_PREFIX_PATH=f:/LLVM/6.0.1/dist \
|
||||
-DCUDA_TOOLKIT_ROOT_DIR="c:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v9.0" \
|
||||
-DCUDA_SDK_ROOT_DIR="c:/ProgramData/NVIDIA Corporation/CUDA Samples/v9.0" \
|
||||
-DCUDA_DNN_ROOT_DIR=f:/CUDNN/cudnn-9.0-windows10-x64-v7.1 \
|
||||
-DLLVM_EXTERNAL_LIT=f:/LLVM/6.0.0/build/Release/bin/llvm-lit.py \
|
||||
-DLLVM_EXTERNAL_LIT=f:/LLVM/6.0.1/build/Release/bin/llvm-lit.py \
|
||||
-Thost=x64
|
||||
..
|
||||
```
|
||||
A corresponding successful output:
|
||||
```shell
|
||||
-- Found LLVM 6.0.0:
|
||||
-- - CMake module path: F:/LLVM/6.0.0/dist/lib/cmake/llvm
|
||||
-- - Include path : F:/LLVM/6.0.0/dist/include
|
||||
-- - Binary path : F:/LLVM/6.0.0/dist/bin
|
||||
-- Found LLVM 6.0.1:
|
||||
-- - CMake module path: F:/LLVM/6.0.1/dist/lib/cmake/llvm
|
||||
-- - Include path : F:/LLVM/6.0.1/dist/include
|
||||
-- - Binary path : F:/LLVM/6.0.1/dist/bin
|
||||
-- Found PythonInterp: C:/Program Files/Python36/python.exe (found suitable version "3.6.4", minimum required is "2.7")
|
||||
-- Found lit: C:/Program Files/Python36/Scripts/lit.exe
|
||||
-- Found FileCheck: F:/LLVM/6.0.0/dist/bin/FileCheck.exe
|
||||
-- Found FileCheck: F:/LLVM/6.0.1/dist/bin/FileCheck.exe
|
||||
-- Found CUDA: C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v9.0 (found version "9.0")
|
||||
-- Configuring done
|
||||
-- Generating done
|
||||
-- Build files have been written to: f:/HIP/hipify-clang/build
|
||||
```
|
||||
|
||||
## Running and using hipify-clang
|
||||
## <a name="running-and-using-hipify-clang"></a> Running and using hipify-clang
|
||||
|
||||
To process a file, `hipify-clang` needs access to the same headers that would be needed to compile it with clang.
|
||||
|
||||
@@ -188,7 +200,7 @@ may be useful.
|
||||
|
||||
For a list of `hipify-clang` options, run `hipify-clang --help`.
|
||||
|
||||
## Disclaimer
|
||||
## <a name="disclaimer"></a> Disclaimer
|
||||
|
||||
The information contained herein is for informational purposes only, and is subject to change without notice. While every precaution has been taken in the preparation of this document, it may contain technical inaccuracies, omissions and typographical errors, and AMD is under no obligation to update or otherwise correct this information. Advanced Micro Devices, Inc. makes no representations or warranties with respect to the accuracy or completeness of the contents of this document, and assumes no liability of any kind, including the implied warranties of noninfringement, merchantability or fitness for particular purposes, with respect to the operation or use of AMD hardware, software or other products described herein. No license, including implied or arising by estoppel, to any intellectual property rights is granted by this document. Terms and limitations applicable to the purchase or use of AMD's products are as set forth in a signed agreement between the parties or in AMD's Standard Terms and Conditions of Sale.
|
||||
|
||||
|
||||
Разница между файлами не показана из-за своего большого размера
Загрузить разницу
@@ -164,6 +164,10 @@ bool HipifyAction::Exclude(const hipCounter & hipToken) {
|
||||
if (insertedDNNHeader) { return true; }
|
||||
insertedDNNHeader = true;
|
||||
return false;
|
||||
case API_FFT:
|
||||
if (insertedFFTHeader) { return true; }
|
||||
insertedFFTHeader = true;
|
||||
return false;
|
||||
default:
|
||||
return false;
|
||||
}
|
||||
|
||||
@@ -28,6 +28,7 @@ private:
|
||||
bool insertedRANDHeader = false;
|
||||
bool insertedRAND_kernelHeader = false;
|
||||
bool insertedDNNHeader = false;
|
||||
bool insertedFFTHeader = false;
|
||||
bool firstHeader = false;
|
||||
bool pragmaOnce = false;
|
||||
clang::SourceLocation firstHeaderLoc;
|
||||
|
||||
@@ -14,7 +14,7 @@ const char *counterNames[NUM_CONV_TYPES] = {
|
||||
};
|
||||
|
||||
const char *apiNames[NUM_API_TYPES] = {
|
||||
"CUDA Driver API", "CUDA RT API", "CUBLAS API", "CURAND API", "CUDNN API"
|
||||
"CUDA Driver API", "CUDA RT API", "CUBLAS API", "CURAND API", "CUDNN API", "CUFFT API"
|
||||
};
|
||||
|
||||
namespace {
|
||||
|
||||
@@ -57,6 +57,7 @@ enum ApiTypes {
|
||||
API_BLAS,
|
||||
API_RAND,
|
||||
API_DNN,
|
||||
API_FFT,
|
||||
API_LAST
|
||||
};
|
||||
constexpr int NUM_API_TYPES = (int) ApiTypes::API_LAST;
|
||||
|
||||
@@ -48,69 +48,35 @@ __device__ static inline unsigned int __popcll(unsigned long long int input) {
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __clz(unsigned int input) {
|
||||
#ifdef NVCC_COMPAT
|
||||
return input == 0 ? 32 : __builtin_clz(input);
|
||||
#else
|
||||
return input == 0 ? -1 : __builtin_clz(input);
|
||||
#endif
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __clzll(unsigned long long int input) {
|
||||
#ifdef NVCC_COMPAT
|
||||
return input == 0 ? 64 : ( input == 0 ? -1 : __builtin_clzl(input) );
|
||||
#else
|
||||
return input == 0 ? -1 : __builtin_clzl(input);
|
||||
#endif
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __clz(int input) {
|
||||
#ifdef NVCC_COMPAT
|
||||
return input == 0 ? 32 : ( input > 0 ? __builtin_clz(input) : __builtin_clz(~input) );
|
||||
#else
|
||||
if (input == 0) return -1;
|
||||
return input > 0 ? __builtin_clz(input) : __builtin_clz(~input);
|
||||
#endif
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __clzll(long long int input) {
|
||||
#ifdef NVCC_COMPAT
|
||||
return input == 0 ? 64 : input > 0 ? __builtin_clzl(input) : __builtin_clzl(~input);
|
||||
#else
|
||||
if (input == 0) return -1;
|
||||
return input > 0 ? __builtin_clzl(input) : __builtin_clzl(~input);
|
||||
#endif
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __ffs(unsigned int input) {
|
||||
#ifdef NVCC_COMPAT
|
||||
return ( input == 0 ? -1 : __builtin_ctz(input) ) + 1;
|
||||
#else
|
||||
return input == 0 ? -1 : __builtin_ctz(input);
|
||||
#endif
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __ffsll(unsigned long long int input) {
|
||||
#ifdef NVCC_COMPAT
|
||||
return ( input == 0 ? -1 : __builtin_ctzl(input) ) + 1;
|
||||
#else
|
||||
return input == 0 ? -1 : __builtin_ctzl(input);
|
||||
#endif
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __ffs(int input) {
|
||||
#ifdef NVCC_COMPAT
|
||||
return ( input == 0 ? -1 : __builtin_ctz(input) ) + 1;
|
||||
#else
|
||||
return input == 0 ? -1 : __builtin_ctz(input);
|
||||
#endif
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __ffsll(long long int input) {
|
||||
#ifdef NVCC_COMPAT
|
||||
return ( input == 0 ? -1 : __builtin_ctzl(input) ) + 1;
|
||||
#else
|
||||
return input == 0 ? -1 : __builtin_ctzl(input);
|
||||
#endif
|
||||
}
|
||||
|
||||
__device__ static inline unsigned int __brev(unsigned int input) { return __llvm_bitrev_b32(input); }
|
||||
@@ -673,14 +639,7 @@ int __all(int predicate) {
|
||||
__device__
|
||||
inline
|
||||
int __any(int predicate) {
|
||||
#ifdef NVCC_COMPAT
|
||||
if (__ockl_wfany_i32(predicate) != 0)
|
||||
return 1;
|
||||
else
|
||||
return 0;
|
||||
#else
|
||||
return __ockl_wfany_i32(predicate);
|
||||
#endif
|
||||
}
|
||||
|
||||
// XXX from llvm/include/llvm/IR/InstrTypes.h
|
||||
|
||||
@@ -57,8 +57,6 @@ THE SOFTWARE.
|
||||
|
||||
#if __HCC_OR_HIP_CLANG__
|
||||
|
||||
// Define NVCC_COMPAT for CUDA compatibility
|
||||
#define NVCC_COMPAT
|
||||
#define CUDA_SUCCESS hipSuccess
|
||||
|
||||
#include <hip/hip_runtime_api.h>
|
||||
@@ -110,9 +108,9 @@ extern int HIP_TRACE_API;
|
||||
#include <hip/hcc_detail/host_defines.h>
|
||||
#include <hip/hcc_detail/device_functions.h>
|
||||
#include <hip/hcc_detail/surface_functions.h>
|
||||
#include <hip/hcc_detail/texture_functions.h>
|
||||
#if __HCC__
|
||||
#include <hip/hcc_detail/math_functions.h>
|
||||
#include <hip/hcc_detail/texture_functions.h>
|
||||
#endif // __HCC__
|
||||
|
||||
// TODO-HCC remove old definitions ; ~1602 hcc supports __HCC_ACCELERATOR__ define.
|
||||
@@ -450,6 +448,14 @@ extern const __device__ __attribute__((weak)) __hip_builtin_gridDim_t gridDim;
|
||||
|
||||
#include <hip/hcc_detail/math_functions.h>
|
||||
|
||||
#endif
|
||||
hipError_t hipHccModuleLaunchKernel(hipFunction_t f, uint32_t globalWorkSizeX,
|
||||
uint32_t globalWorkSizeY, uint32_t globalWorkSizeZ,
|
||||
uint32_t localWorkSizeX, uint32_t localWorkSizeY,
|
||||
uint32_t localWorkSizeZ, size_t sharedMemBytes,
|
||||
hipStream_t hStream, void** kernelParams, void** extra,
|
||||
hipEvent_t startEvent = nullptr,
|
||||
hipEvent_t stopEvent = nullptr);
|
||||
|
||||
#endif // defined(__clang__) && defined(__HIP__)
|
||||
|
||||
#endif // HIP_HCC_DETAIL_RUNTIME_H
|
||||
|
||||
Разница между файлами не показана из-за своего большого размера
Загрузить разницу
@@ -1808,7 +1808,7 @@ bool ihipStream_t::canSeeMemory(const ihipCtx_t* copyEngineCtx, const hc::AmPoin
|
||||
// TODO - pointer-info stores a deviceID not a context,may have some unusual side-effects here:
|
||||
if (dstPtrInfo->_sizeBytes == 0) {
|
||||
return false;
|
||||
} else {
|
||||
} else if (dstPtrInfo->_appId != -1) {
|
||||
#if USE_APP_PTR_FOR_CTX
|
||||
ihipCtx_t* dstCtx = static_cast<ihipCtx_t*>(dstPtrInfo->_appPtr);
|
||||
#else
|
||||
@@ -1831,7 +1831,7 @@ bool ihipStream_t::canSeeMemory(const ihipCtx_t* copyEngineCtx, const hc::AmPoin
|
||||
// TODO - pointer-info stores a deviceID not a context,may have some unusual side-effects here:
|
||||
if (srcPtrInfo->_sizeBytes == 0) {
|
||||
return false;
|
||||
} else {
|
||||
} else if (srcPtrInfo->_appId != -1) {
|
||||
#if USE_APP_PTR_FOR_CTX
|
||||
ihipCtx_t* srcCtx = static_cast<ihipCtx_t*>(srcPtrInfo->_appPtr);
|
||||
#else
|
||||
|
||||
@@ -61,19 +61,20 @@ int sharePtr(void* ptr, ihipCtx_t* ctx, bool shareWithAll, unsigned hipFlags) {
|
||||
|
||||
auto device = ctx->getWriteableDevice();
|
||||
|
||||
#if USE_APP_PTR_FOR_CTX
|
||||
hc::am_memtracker_update(ptr, device->_deviceId, hipFlags, ctx);
|
||||
#else
|
||||
hc::am_memtracker_update(ptr, device->_deviceId, hipFlags);
|
||||
#endif
|
||||
|
||||
if (shareWithAll) {
|
||||
// shareWithAll memory is not mapped to any device
|
||||
hc::am_memtracker_update(ptr, -1, hipFlags);
|
||||
hsa_status_t s = hsa_amd_agents_allow_access(g_deviceCnt + 1, g_allAgents, NULL, ptr);
|
||||
tprintf(DB_MEM, " allow access to CPU + all %d GPUs (shareWithAll)\n", g_deviceCnt);
|
||||
if (s != HSA_STATUS_SUCCESS) {
|
||||
ret = -1;
|
||||
}
|
||||
} else {
|
||||
#if USE_APP_PTR_FOR_CTX
|
||||
hc::am_memtracker_update(ptr, device->_deviceId, hipFlags, ctx);
|
||||
#else
|
||||
hc::am_memtracker_update(ptr, device->_deviceId, hipFlags);
|
||||
#endif
|
||||
int peerCnt = 0;
|
||||
{
|
||||
LockedAccessor_CtxCrit_t crit(ctx->criticalData());
|
||||
|
||||
@@ -676,6 +676,8 @@ hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference*
|
||||
|
||||
auto ctx = ihipGetTlsDefaultCtx();
|
||||
if (ctx) {
|
||||
if(offset != nullptr)
|
||||
*offset = 0;
|
||||
}
|
||||
return ihipLogStatus(hip_status);
|
||||
}
|
||||
|
||||
@@ -0,0 +1,79 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
#include <stdio.h>
|
||||
#include <stdlib.h>
|
||||
#include <math.h>
|
||||
// CHECK: #include "hipblas.h"
|
||||
#include "cublas.h"
|
||||
#define M 6
|
||||
#define N 5
|
||||
#define IDX2C(i,j,ld) (((j)*(ld))+(i))
|
||||
static __inline__ void modify(float *m, int ldm, int n, int p, int q, float
|
||||
alpha, float beta) {
|
||||
// CHECK: hipblasSscal(n - p, alpha, &m[IDX2C(p, q, ldm)], ldm);
|
||||
// CHECK: hipblasSscal(ldm - p, beta, &m[IDX2C(p, q, ldm)], 1);
|
||||
cublasSscal(n - p, alpha, &m[IDX2C(p, q, ldm)], ldm);
|
||||
cublasSscal(ldm - p, beta, &m[IDX2C(p, q, ldm)], 1);
|
||||
}
|
||||
int main(void) {
|
||||
int i, j;
|
||||
// CHECK: hipblasStatus_t stat;
|
||||
cublasStatus stat;
|
||||
float* devPtrA;
|
||||
float* a = 0;
|
||||
a = (float *)malloc(M * N * sizeof(*a));
|
||||
if (!a) {
|
||||
printf("host memory allocation failed");
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
for (j = 0; j < N; j++) {
|
||||
for (i = 0; i < M; i++) {
|
||||
a[IDX2C(i, j, M)] = (float)(i * M + j + 1);
|
||||
}
|
||||
}
|
||||
// cublasInit is not supported yet
|
||||
cublasInit();
|
||||
stat = cublasAlloc(M*N, sizeof(*a), (void**)&devPtrA);
|
||||
// CHECK: if (stat != HIPBLAS_STATUS_SUCCESS) {
|
||||
if (stat != CUBLAS_STATUS_SUCCESS) {
|
||||
printf("device memory allocation failed");
|
||||
// cublasShutdown is not supported yet
|
||||
cublasShutdown();
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
// CHECK: stat = hipblasSetMatrix(M, N, sizeof(*a), a, M, devPtrA, M);
|
||||
stat = cublasSetMatrix(M, N, sizeof(*a), a, M, devPtrA, M);
|
||||
// CHECK: if (stat != HIPBLAS_STATUS_SUCCESS) {
|
||||
if (stat != CUBLAS_STATUS_SUCCESS) {
|
||||
printf("data download failed");
|
||||
// cublasFree is not supported yet
|
||||
cublasFree(devPtrA);
|
||||
// cublasShutdown is not supported yet
|
||||
cublasShutdown();
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
modify(devPtrA, M, N, 1, 2, 16.0f, 12.0f);
|
||||
// CHECK: stat = hipblasGetMatrix(M, N, sizeof(*a), devPtrA, M, a, M);
|
||||
stat = cublasGetMatrix(M, N, sizeof(*a), devPtrA, M, a, M);
|
||||
// CHECK: if (stat != HIPBLAS_STATUS_SUCCESS) {
|
||||
if (stat != CUBLAS_STATUS_SUCCESS) {
|
||||
printf("data upload failed");
|
||||
// cublasFree is not supported yet
|
||||
cublasFree(devPtrA);
|
||||
// cublasShutdown is not supported yet
|
||||
cublasShutdown();
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
// cublasFree is not supported yet
|
||||
cublasFree(devPtrA);
|
||||
// cublasShutdown is not supported yet
|
||||
cublasShutdown();
|
||||
for (j = 0; j < N; j++) {
|
||||
for (i = 0; i < M; i++) {
|
||||
printf("%7.0f", a[IDX2C(i, j, M)]);
|
||||
}
|
||||
printf("\n");
|
||||
}
|
||||
free(a);
|
||||
return EXIT_SUCCESS;
|
||||
}
|
||||
@@ -0,0 +1,90 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
#include <stdio.h>
|
||||
#include <stdlib.h>
|
||||
#include <math.h>
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
#include <cuda_runtime.h>
|
||||
// CHECK: #include "hipblas.h"
|
||||
#include "cublas_v2.h"
|
||||
#define M 6
|
||||
#define N 5
|
||||
#define IDX2F(i,j,ld) ((((j)-1)*(ld))+((i)-1))
|
||||
// CHECK: static __inline__ void modify(hipblasHandle_t handle, float *m, int ldm, int
|
||||
static __inline__ void modify(cublasHandle_t handle, float *m, int ldm, int
|
||||
n, int p, int q, float alpha, float beta) {
|
||||
// CHECK: hipblasSscal(handle, n - p + 1, &alpha, &m[IDX2F(p, q, ldm)], ldm);
|
||||
// CHECK: hipblasSscal(handle, ldm - p + 1, &beta, &m[IDX2F(p, q, ldm)], 1);
|
||||
cublasSscal(handle, n - p + 1, &alpha, &m[IDX2F(p, q, ldm)], ldm);
|
||||
cublasSscal(handle, ldm - p + 1, &beta, &m[IDX2F(p, q, ldm)], 1);
|
||||
}
|
||||
int main(void) {
|
||||
// CHECK: hipError_t cudaStat;
|
||||
// CHECK: hipblasStatus_t stat;
|
||||
// CHECK: hipblasHandle_t handle;
|
||||
cudaError_t cudaStat;
|
||||
cublasStatus_t stat;
|
||||
cublasHandle_t handle;
|
||||
int i, j;
|
||||
float* devPtrA;
|
||||
float* a = 0;
|
||||
a = (float *)malloc(M * N * sizeof(*a));
|
||||
if (!a) {
|
||||
printf("host memory allocation failed");
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
for (j = 1; j <= N; j++) {
|
||||
for (i = 1; i <= M; i++) {
|
||||
a[IDX2F(i, j, M)] = (float)((i - 1) * M + j);
|
||||
}
|
||||
}
|
||||
// CHECK: cudaStat = hipMalloc((void**)&devPtrA, M*N * sizeof(*a));
|
||||
cudaStat = cudaMalloc((void**)&devPtrA, M*N * sizeof(*a));
|
||||
// CHECK: if (cudaStat != hipSuccess) {
|
||||
if (cudaStat != cudaSuccess) {
|
||||
printf("device memory allocation failed");
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
// CHECK: stat = hipblasCreate(&handle);
|
||||
stat = cublasCreate(&handle);
|
||||
// CHECK: if (stat != HIPBLAS_STATUS_SUCCESS) {
|
||||
if (stat != CUBLAS_STATUS_SUCCESS) {
|
||||
printf("CUBLAS initialization failed\n");
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
// CHECK: stat = hipblasSetMatrix(M, N, sizeof(*a), a, M, devPtrA, M);
|
||||
stat = cublasSetMatrix(M, N, sizeof(*a), a, M, devPtrA, M);
|
||||
// CHECK: if (stat != HIPBLAS_STATUS_SUCCESS) {
|
||||
if (stat != CUBLAS_STATUS_SUCCESS) {
|
||||
printf("data download failed");
|
||||
// CHECK: hipFree(devPtrA);
|
||||
// CHECK: hipblasDestroy(handle);
|
||||
cudaFree(devPtrA);
|
||||
cublasDestroy(handle);
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
modify(handle, devPtrA, M, N, 2, 3, 16.0f, 12.0f);
|
||||
// CHECK: stat = hipblasGetMatrix(M, N, sizeof(*a), devPtrA, M, a, M);
|
||||
stat = cublasGetMatrix(M, N, sizeof(*a), devPtrA, M, a, M);
|
||||
// CHECK: if (stat != HIPBLAS_STATUS_SUCCESS) {
|
||||
if (stat != CUBLAS_STATUS_SUCCESS) {
|
||||
printf("data upload failed");
|
||||
// CHECK: hipFree(devPtrA);
|
||||
// CHECK: hipblasDestroy(handle);
|
||||
cudaFree(devPtrA);
|
||||
cublasDestroy(handle);
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
// CHECK: hipFree(devPtrA);
|
||||
// CHECK: hipblasDestroy(handle);
|
||||
cudaFree(devPtrA);
|
||||
cublasDestroy(handle);
|
||||
for (j = 1; j <= N; j++) {
|
||||
for (i = 1; i <= M; i++) {
|
||||
printf("%7.0f", a[IDX2F(i, j, M)]);
|
||||
}
|
||||
printf("\n");
|
||||
}
|
||||
free(a);
|
||||
return EXIT_SUCCESS;
|
||||
}
|
||||
@@ -0,0 +1,108 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
#include <stdio.h>
|
||||
#include <stdlib.h>
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
#include <cuda_runtime.h>
|
||||
// CHECK: #include "hipblas.h"
|
||||
#include "cublas_v2.h"
|
||||
#define IDX2C(i,j,ld) (((j)*(ld))+(i))
|
||||
#define m 6
|
||||
#define n 4
|
||||
#define k 5
|
||||
int main(void) {
|
||||
// CHECK: hipError_t cudaStat;
|
||||
// CHECK: hipblasStatus_t stat;
|
||||
// CHECK: hipblasHandle_t handle;
|
||||
cudaError_t cudaStat;
|
||||
cublasStatus_t stat;
|
||||
cublasHandle_t handle;
|
||||
int i, j;
|
||||
float * a;
|
||||
float * b;
|
||||
float * c;
|
||||
a = (float *)malloc(m*k * sizeof(float));
|
||||
b = (float *)malloc(k*n * sizeof(float));
|
||||
c = (float *)malloc(m*n * sizeof(float));
|
||||
int ind = 11;
|
||||
for (j = 0; j<k; j++) {
|
||||
for (i = 0; i<m; i++) {
|
||||
a[IDX2C(i, j, m)] = (float)ind++;
|
||||
}
|
||||
}
|
||||
printf("a:\n");
|
||||
for (i = 0; i<m; i++) {
|
||||
for (j = 0; j<k; j++) {
|
||||
printf(" %5.0f", a[IDX2C(i, j, m)]);
|
||||
}
|
||||
printf("\n");
|
||||
}
|
||||
ind = 11;
|
||||
for (j = 0; j<n; j++) {
|
||||
for (i = 0; i<k; i++) {
|
||||
b[IDX2C(i, j, k)] = (float)ind++;
|
||||
}
|
||||
}
|
||||
printf("b:\n");
|
||||
for (i = 0; i<k; i++) {
|
||||
for (j = 0; j<n; j++) {
|
||||
printf(" %5.0f", b[IDX2C(i, j, k)]);
|
||||
}
|
||||
printf("\n");
|
||||
}
|
||||
ind = 11;
|
||||
for (j = 0; j<n; j++) {
|
||||
for (i = 0; i<m; i++) {
|
||||
c[IDX2C(i, j, m)] = (float)ind++;
|
||||
}
|
||||
}
|
||||
printf("c:\n");
|
||||
for (i = 0; i<m; i++) {
|
||||
for (j = 0; j<n; j++) {
|
||||
printf(" %5.0f", c[IDX2C(i, j, m)]);
|
||||
}
|
||||
printf("\n");
|
||||
}
|
||||
float * d_a;
|
||||
float * d_b;
|
||||
float * d_c;
|
||||
// CHECK: cudaStat = hipMalloc((void **)& d_a, m*k * sizeof(*a));
|
||||
// CHECK: cudaStat = hipMalloc((void **)& d_b, k*n * sizeof(*b));
|
||||
// CHECK: cudaStat = hipMalloc((void **)& d_c, m*n * sizeof(*c));
|
||||
cudaStat = cudaMalloc((void **)& d_a, m*k * sizeof(*a));
|
||||
cudaStat = cudaMalloc((void **)& d_b, k*n * sizeof(*b));
|
||||
cudaStat = cudaMalloc((void **)& d_c, m*n * sizeof(*c));
|
||||
// CHECK: stat = hipblasCreate(&handle);
|
||||
stat = cublasCreate(&handle);
|
||||
// CHECK: stat = hipblasSetMatrix(m, k, sizeof(*a), a, m, d_a, m);
|
||||
// CHECK: stat = hipblasSetMatrix(k, n, sizeof(*b), b, k, d_b, k);
|
||||
// CHECK: stat = hipblasSetMatrix(m, n, sizeof(*c), c, m, d_c, m);
|
||||
stat = cublasSetMatrix(m, k, sizeof(*a), a, m, d_a, m);
|
||||
stat = cublasSetMatrix(k, n, sizeof(*b), b, k, d_b, k);
|
||||
stat = cublasSetMatrix(m, n, sizeof(*c), c, m, d_c, m);
|
||||
float al = 1.0f;
|
||||
float bet = 1.0f;
|
||||
// CHECK: stat = hipblasSgemm(handle, HIPBLAS_OP_N, HIPBLAS_OP_N, m, n, k, &al, d_a, m, d_b, k, &bet, d_c, m);
|
||||
stat = cublasSgemm(handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, &al, d_a, m, d_b, k, &bet, d_c, m);
|
||||
// CHECK: stat = hipblasGetMatrix(m, n, sizeof(*c), d_c, m, c, m);
|
||||
stat = cublasGetMatrix(m, n, sizeof(*c), d_c, m, c, m);
|
||||
printf("c after Sgemm :\n");
|
||||
for (i = 0; i<m; i++) {
|
||||
for (j = 0; j<n; j++) {
|
||||
printf(" %7.0f", c[IDX2C(i, j, m)]);
|
||||
}
|
||||
printf("\n");
|
||||
}
|
||||
// CHECK: hipFree(d_a);
|
||||
// CHECK: hipFree(d_b);
|
||||
// CHECK: hipFree(d_c);
|
||||
// CHECK: hipblasDestroy(handle);
|
||||
cudaFree(d_a);
|
||||
cudaFree(d_b);
|
||||
cudaFree(d_c);
|
||||
cublasDestroy(handle);
|
||||
free(a);
|
||||
free(b);
|
||||
free(c);
|
||||
return EXIT_SUCCESS;
|
||||
}
|
||||
@@ -0,0 +1,78 @@
|
||||
// RUN: %run_test hipify "%s" "%t" %cuda_args
|
||||
|
||||
// CHECK: #include <hip/hip_runtime.h>
|
||||
#include <cuda.h>
|
||||
// CHECK: #include <hipfft.h>
|
||||
#include <cufft.h>
|
||||
#include <stdio.h>
|
||||
#include <math.h>
|
||||
|
||||
#define DATASIZE 8
|
||||
#define BATCH 2
|
||||
|
||||
#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
|
||||
// CHECK: inline void gpuAssert(hipError_t code, const char *file, int line, bool abort = true)
|
||||
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort = true)
|
||||
{
|
||||
// CHECK: if (code != hipSuccess)
|
||||
if (code != cudaSuccess)
|
||||
{
|
||||
// CHECK: fprintf(stderr, "GPUassert: %s %s %dn", hipGetErrorString(code), file, line);
|
||||
fprintf(stderr, "GPUassert: %s %s %dn", cudaGetErrorString(code), file, line);
|
||||
if (abort) exit(code);
|
||||
}
|
||||
}
|
||||
|
||||
int main()
|
||||
{
|
||||
// --- Host side input data allocation and initialization
|
||||
// CHECK: hipfftReal *hostInputData = (hipfftReal*)malloc(DATASIZE*BATCH * sizeof(hipfftReal));
|
||||
cufftReal *hostInputData = (cufftReal*)malloc(DATASIZE*BATCH * sizeof(cufftReal));
|
||||
for (int i = 0; i<BATCH; i++)
|
||||
for (int j = 0; j<DATASIZE; j++) hostInputData[i*DATASIZE + j] = (cufftReal)(i + 1);
|
||||
|
||||
// --- Device side input data allocation and initialization
|
||||
cufftReal *deviceInputData; gpuErrchk(cudaMalloc((void**)&deviceInputData, DATASIZE * BATCH * sizeof(cufftReal)));
|
||||
// CHECK: hipMemcpy(deviceInputData, hostInputData, DATASIZE * BATCH * sizeof(hipfftReal), hipMemcpyHostToDevice);
|
||||
cudaMemcpy(deviceInputData, hostInputData, DATASIZE * BATCH * sizeof(cufftReal), cudaMemcpyHostToDevice);
|
||||
|
||||
// --- Host side output data allocation
|
||||
cufftComplex *hostOutputData = (cufftComplex*)malloc((DATASIZE / 2 + 1) * BATCH * sizeof(cufftComplex));
|
||||
|
||||
// --- Device side output data allocation
|
||||
cufftComplex *deviceOutputData; gpuErrchk(cudaMalloc((void**)&deviceOutputData, (DATASIZE / 2 + 1) * BATCH * sizeof(cufftComplex)));
|
||||
|
||||
// --- Batched 1D FFTs
|
||||
// CHECK: hipfftHandle handle;
|
||||
cufftHandle handle;
|
||||
int rank = 1; // --- 1D FFTs
|
||||
int n[] = { DATASIZE }; // --- Size of the Fourier transform
|
||||
int istride = 1, ostride = 1; // --- Distance between two successive input/output elements
|
||||
int idist = DATASIZE, odist = (DATASIZE / 2 + 1); // --- Distance between batches
|
||||
int inembed[] = { 0 }; // --- Input size with pitch (ignored for 1D transforms)
|
||||
int onembed[] = { 0 }; // --- Output size with pitch (ignored for 1D transforms)
|
||||
int batch = BATCH; // --- Number of batched executions
|
||||
// CHECK: hipfftPlanMany(&handle, rank, n,
|
||||
cufftPlanMany(&handle, rank, n,
|
||||
inembed, istride, idist,
|
||||
// CHECK: onembed, ostride, odist, HIPFFT_R2C, batch);
|
||||
onembed, ostride, odist, CUFFT_R2C, batch);
|
||||
|
||||
// CHECK: hipfftExecR2C(handle, deviceInputData, deviceOutputData);
|
||||
cufftExecR2C(handle, deviceInputData, deviceOutputData);
|
||||
|
||||
// --- Device->Host copy of the results
|
||||
// CHECK: gpuErrchk(hipMemcpy(hostOutputData, deviceOutputData, (DATASIZE / 2 + 1) * BATCH * sizeof(hipfftComplex), hipMemcpyDeviceToHost));
|
||||
gpuErrchk(cudaMemcpy(hostOutputData, deviceOutputData, (DATASIZE / 2 + 1) * BATCH * sizeof(cufftComplex), cudaMemcpyDeviceToHost));
|
||||
|
||||
for (int i = 0; i<BATCH; i++)
|
||||
for (int j = 0; j<(DATASIZE / 2 + 1); j++)
|
||||
printf("%i %i %f %fn", i, j, hostOutputData[i*(DATASIZE / 2 + 1) + j].x, hostOutputData[i*(DATASIZE / 2 + 1) + j].y);
|
||||
|
||||
// CHECK: hipfftDestroy(handle);
|
||||
cufftDestroy(handle);
|
||||
// CHECK: gpuErrchk(hipFree(deviceOutputData));
|
||||
// CHECK: gpuErrchk(hipFree(deviceInputData));
|
||||
gpuErrchk(cudaFree(deviceOutputData));
|
||||
gpuErrchk(cudaFree(deviceInputData));
|
||||
}
|
||||
@@ -47,6 +47,8 @@
|
||||
|
||||
// CHECK: #include <string>
|
||||
|
||||
// CHECK: #include "hipfft.h"
|
||||
|
||||
#include <cuda.h>
|
||||
|
||||
#include <memory>
|
||||
@@ -91,3 +93,5 @@
|
||||
#include "curand_uniform.h"
|
||||
|
||||
#include <string>
|
||||
|
||||
#include "cufft.h"
|
||||
|
||||
@@ -17,8 +17,8 @@ OUT OF OR INN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
|
||||
THE SOFTWARE.
|
||||
*/
|
||||
/* HIT_START
|
||||
* BUILD: %t %s
|
||||
* RUN: %t
|
||||
* BUILD: %t %s NVCC_OPTIONS -std=c++11
|
||||
* RUN: %t EXCLUDE_HIP_PLATFORM nvcc
|
||||
* HIT_END
|
||||
*/
|
||||
#include "test_common.h"
|
||||
|
||||
@@ -86,24 +86,15 @@ int main(int argc, char* argv[]) {
|
||||
printf("warp no. %d __all = %d \n", i, host_all[i]);
|
||||
|
||||
if (host_all[i] != 1) ++allcount;
|
||||
#if defined(__HIP_PLATFORM_HCC__) && !defined(NVCC_COMPAT)
|
||||
if (host_any[i] != 64) ++anycount;
|
||||
#else
|
||||
if (host_any[i] != 1) ++anycount;
|
||||
#endif
|
||||
}
|
||||
|
||||
#if defined(__HIP_PLATFORM_HCC__) && !defined(NVCC_COMPAT)
|
||||
if (anycount == 1 && allcount == 1)
|
||||
printf("PASSED\n");
|
||||
else
|
||||
printf("FAILED\n");
|
||||
#else
|
||||
if (anycount == 0 && allcount == 1)
|
||||
printf("PASSED\n");
|
||||
else
|
||||
else {
|
||||
printf("FAILED\n");
|
||||
#endif
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
|
||||
return EXIT_SUCCESS;
|
||||
}
|
||||
|
||||
@@ -89,7 +89,9 @@ int main(int argc, char* argv[]) {
|
||||
|
||||
if (divergent_count == 1)
|
||||
printf("PASSED\n");
|
||||
else
|
||||
else {
|
||||
printf("FAILED\n");
|
||||
return EXIT_FAILURE;
|
||||
}
|
||||
return EXIT_SUCCESS;
|
||||
}
|
||||
|
||||
@@ -45,12 +45,7 @@ THE SOFTWARE.
|
||||
|
||||
unsigned int firstbit_u32(unsigned int a) {
|
||||
if (a == 0) {
|
||||
#if defined(__HIP_PLATFORM_HCC__) && !defined(NVCC_COMPAT)
|
||||
|
||||
return -1;
|
||||
#else
|
||||
return 32;
|
||||
#endif
|
||||
}
|
||||
unsigned int pos = 0;
|
||||
while ((int)a > 0) {
|
||||
@@ -62,11 +57,7 @@ unsigned int firstbit_u32(unsigned int a) {
|
||||
|
||||
unsigned int firstbit_u64(unsigned long long int a) {
|
||||
if (a == 0) {
|
||||
#if defined(__HIP_PLATFORM_HCC__) && !defined(NVCC_COMPAT)
|
||||
return -1;
|
||||
#else
|
||||
return 64;
|
||||
#endif
|
||||
}
|
||||
unsigned int pos = 0;
|
||||
while ((long long int)a > 0) {
|
||||
|
||||
@@ -49,21 +49,13 @@ THE SOFTWARE.
|
||||
template <typename T>
|
||||
int lastbit(T a) {
|
||||
if (a == 0)
|
||||
#if defined(__HIP_PLATFORM_HCC__) && !defined(NVCC_COMPAT)
|
||||
return -1;
|
||||
#else
|
||||
return 0;
|
||||
#endif
|
||||
int pos = 1;
|
||||
while ((a & 1) != 1) {
|
||||
a >>= 1;
|
||||
pos++;
|
||||
}
|
||||
#if defined(__HIP_PLATFORM_HCC__) && !defined(NVCC_COMPAT)
|
||||
return pos - 1;
|
||||
#else
|
||||
return pos;
|
||||
#endif
|
||||
}
|
||||
|
||||
|
||||
|
||||
@@ -18,7 +18,7 @@ THE SOFTWARE.
|
||||
*/
|
||||
|
||||
/* HIT_START
|
||||
* BUILD: %t %s ../test_common.cpp
|
||||
* BUILD: %t %s ../test_common.cpp NVCC_OPTIONS -std=c++11
|
||||
* RUN: %t
|
||||
* HIT_END
|
||||
*/
|
||||
@@ -54,7 +54,7 @@ THE SOFTWARE.
|
||||
// Bit fields are broken
|
||||
#define ENABLE_BIT_FIELDS 0
|
||||
|
||||
static const int BLOCK_DIM_SIZE = 1024;
|
||||
static const int BLOCK_DIM_SIZE = 512;
|
||||
|
||||
// allocate memory on device and host for result validation
|
||||
static bool *result_d, *result_h;
|
||||
@@ -882,9 +882,10 @@ int main() {
|
||||
// Test: Passing struct which is initiazed out of order
|
||||
// accessing same elements in order from device
|
||||
ResetValidationMem();
|
||||
hipLaunchKernelStruct_t20 hipLaunchKernelStruct_h20 =
|
||||
// out of order initalization
|
||||
{.name = 'A', .rank = 2, .age = 42};
|
||||
hipLaunchKernelStruct_t20 hipLaunchKernelStruct_h20;
|
||||
hipLaunchKernelStruct_h20.name = 'A';
|
||||
hipLaunchKernelStruct_h20.rank = 2;
|
||||
hipLaunchKernelStruct_h20.age = 42;
|
||||
bool *result_d20, *result_h20;
|
||||
#if ENABLE_OUT_OF_ORDER_INITIALIZATION
|
||||
hipLaunchKernelGGL(HIP_KERNEL_NAME(hipLaunchKernelStructFunc20),
|
||||
|
||||
@@ -0,0 +1,78 @@
|
||||
/*
|
||||
Copyright (c) 2015-Present Advanced Micro Devices, Inc. All rights reserved.
|
||||
|
||||
Permission is hereby granted, free of charge, to any person obtaining a copy
|
||||
of this software and associated documentation files (the "Software"), to deal
|
||||
in the Software without restriction, including without limitation the rights
|
||||
to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
|
||||
copies of the Software, and to permit persons to whom the Software is
|
||||
furnished to do so, subject to the following conditions:
|
||||
|
||||
The above copyright notice and this permission notice shall be included in
|
||||
all copies or substantial portions of the Software.
|
||||
|
||||
THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
|
||||
IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
|
||||
FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
|
||||
AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
|
||||
LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
|
||||
OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
|
||||
THE SOFTWARE.
|
||||
*/
|
||||
|
||||
/* HIT_START
|
||||
* BUILD: %t %s ../test_common.cpp
|
||||
* RUN: %t
|
||||
* HIT_END
|
||||
*/
|
||||
#include <hip/hip_runtime.h>
|
||||
#include "test_common.h"
|
||||
|
||||
using namespace std;
|
||||
#define R 8 //rows, height
|
||||
#define C 8 //columns, width
|
||||
|
||||
texture<int, hipTextureType2D,hipReadModeElementType> tex;
|
||||
|
||||
bool runTest(void);
|
||||
|
||||
int main(int argc, char** argv) {
|
||||
bool testResult=runTest();
|
||||
|
||||
if (testResult) {
|
||||
passed();
|
||||
} else {
|
||||
exit(EXIT_FAILURE);
|
||||
}
|
||||
}
|
||||
|
||||
bool runTest()
|
||||
{
|
||||
int val[R][C],i,j;
|
||||
size_t offset;
|
||||
|
||||
for(i=0;i<R;i++)
|
||||
for(j=0;j<C;j++)
|
||||
{
|
||||
val[i][j]=(i+1)*(j+1);
|
||||
}
|
||||
hipChannelFormatDesc chan_desc=hipCreateChannelDesc(32,0,0,0,hipChannelFormatKindSigned);
|
||||
hipArray *hipArray;
|
||||
HIPCHECK(hipMallocArray(&hipArray, &chan_desc,C,R,0));
|
||||
|
||||
HIPCHECK(hipMemcpyToArray(hipArray,0,0, val, R*C*sizeof(int), hipMemcpyHostToDevice));
|
||||
|
||||
tex.addressMode[0]=hipAddressModeWrap;
|
||||
tex.addressMode[1]=hipAddressModeWrap;
|
||||
tex.filterMode=hipFilterModePoint;
|
||||
tex.normalized=0;
|
||||
|
||||
HIPCHECK(hipBindTextureToArray(&tex, hipArray, &chan_desc));
|
||||
HIPCHECK(hipGetTextureAlignmentOffset(&offset,&tex));
|
||||
HIPCHECK(hipUnbindTexture(&tex));
|
||||
HIPCHECK(hipFreeArray(hipArray));
|
||||
if(offset != 0)
|
||||
return false;
|
||||
else
|
||||
return true;
|
||||
}
|
||||
@@ -0,0 +1,63 @@
|
||||
/*
|
||||
Copyright (c) 2015-Present Advanced Micro Devices, Inc. All rights reserved.
|
||||
|
||||
Permission is hereby granted, free of charge, to any person obtaining a copy
|
||||
of this software and associated documentation files (the "Software"), to deal
|
||||
in the Software without restriction, including without limitation the rights
|
||||
to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
|
||||
copies of the Software, and to permit persons to whom the Software is
|
||||
furnished to do so, subject to the following conditions:
|
||||
|
||||
The above copyright notice and this permission notice shall be included in
|
||||
all copies or substantial portions of the Software.
|
||||
|
||||
THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
|
||||
IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
|
||||
FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
|
||||
AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
|
||||
LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
|
||||
OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
|
||||
THE SOFTWARE.
|
||||
*/
|
||||
|
||||
/* HIT_START
|
||||
* BUILD: %t %s ../test_common.cpp
|
||||
* RUN: %t
|
||||
* HIT_END
|
||||
*/
|
||||
|
||||
#include <hip/hip_runtime.h>
|
||||
#include "test_common.h"
|
||||
|
||||
using namespace std;
|
||||
#define R 8 //rows, height
|
||||
#define C 8 //columns, width
|
||||
|
||||
bool runTest(void);
|
||||
|
||||
int main(int argc, char** argv) {
|
||||
bool testResult=runTest();
|
||||
|
||||
if (testResult) {
|
||||
passed();
|
||||
} else {
|
||||
exit(EXIT_FAILURE);
|
||||
}
|
||||
}
|
||||
|
||||
bool runTest()
|
||||
{
|
||||
bool testResult=true;
|
||||
hipChannelFormatDesc chan_test,chan_desc=hipCreateChannelDesc(32,0,0,0,hipChannelFormatKindSigned);
|
||||
hipArray *hipArray;
|
||||
HIPCHECK(hipMallocArray(&hipArray, &chan_desc,C,R,0));
|
||||
HIPCHECK(hipGetChannelDesc(&chan_test,hipArray));
|
||||
|
||||
if((chan_test.x == 32)&&(chan_test.y == 0)&&(chan_test.z == 0)&&(chan_test.f == 0))
|
||||
testResult=true;
|
||||
else
|
||||
testResult=false;
|
||||
|
||||
HIPCHECK(hipFreeArray(hipArray));
|
||||
return testResult;
|
||||
}
|
||||
Ссылка в новой задаче
Block a user