From 971b75588ed82d7b0470ba6e1848f5bd37f2ea3c Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Thu, 15 Aug 2019 19:13:00 -0700 Subject: [PATCH 01/14] [dtests] refactor windows specific changes (#1313) * [dtests] refactor windows specific changes * Refactor hipMemoryAllocateCoherentDriver - PR- 1309 * Fix missing z in _putenv_s * Revert "Fix missing z in _putenv_s" This reverts commit 099a1b20a5c75c5f122d57c0ad2bca01745cdc9c. * Refactor changes from PR 1299 * Update hipEnvVarDriver.cpp [ROCm/hip commit: 44422065a1bf6b63fafaf529341dfdb35c87a247] --- projects/hip/tests/src/hipEnvVar.cpp | 9 +-------- projects/hip/tests/src/hipEnvVarDriver.cpp | 8 +------- .../hip/tests/src/runtimeApi/memory/hipMemcpy_simple.cpp | 3 --- .../memory/hipMemoryAllocateCoherentDriver.cpp | 7 +------ projects/hip/tests/src/test_common.h | 8 ++++++++ 5 files changed, 11 insertions(+), 24 deletions(-) diff --git a/projects/hip/tests/src/hipEnvVar.cpp b/projects/hip/tests/src/hipEnvVar.cpp index 7c261b1fa3..a0b9503b41 100644 --- a/projects/hip/tests/src/hipEnvVar.cpp +++ b/projects/hip/tests/src/hipEnvVar.cpp @@ -31,14 +31,7 @@ THE SOFTWARE. #include "clara/clara.hpp" #include #include "hip/hip_runtime.h" - -#ifdef _WIN64 -#include -#endif - -#ifdef _WIN64 -#define setenv(x,y,z) _putenv_s(x,y) -#endif +#include "test_common.h" using namespace std; using namespace clara; diff --git a/projects/hip/tests/src/hipEnvVarDriver.cpp b/projects/hip/tests/src/hipEnvVarDriver.cpp index 37e51faa31..4817d157a4 100644 --- a/projects/hip/tests/src/hipEnvVarDriver.cpp +++ b/projects/hip/tests/src/hipEnvVarDriver.cpp @@ -30,13 +30,7 @@ OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWA #include "hip/hip_runtime.h" #include #include - -#ifdef _WIN64 -#define popen(x,y) _popen(x,y) -#define pclose(x) _pclose(x) -#define setenv(x,y,z) _putenv_s(x,y) -#define unsetenv(x) _putenv_s(x,"") -#endif +#include "test_common.h" using namespace std; diff --git a/projects/hip/tests/src/runtimeApi/memory/hipMemcpy_simple.cpp b/projects/hip/tests/src/runtimeApi/memory/hipMemcpy_simple.cpp index dcec1fd834..8ce02d6164 100644 --- a/projects/hip/tests/src/runtimeApi/memory/hipMemcpy_simple.cpp +++ b/projects/hip/tests/src/runtimeApi/memory/hipMemcpy_simple.cpp @@ -29,9 +29,6 @@ THE SOFTWARE. #include "hip/hip_runtime.h" #include "test_common.h" -#ifdef _WIN64 -#define aligned_alloc _aligned_malloc -#endif bool p_async = false; diff --git a/projects/hip/tests/src/runtimeApi/memory/hipMemoryAllocateCoherentDriver.cpp b/projects/hip/tests/src/runtimeApi/memory/hipMemoryAllocateCoherentDriver.cpp index a064a6ad0a..b9d7602039 100644 --- a/projects/hip/tests/src/runtimeApi/memory/hipMemoryAllocateCoherentDriver.cpp +++ b/projects/hip/tests/src/runtimeApi/memory/hipMemoryAllocateCoherentDriver.cpp @@ -29,12 +29,7 @@ OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWA #include #include #include "hip/hip_runtime.h" - -#ifdef _WIN64 -#define popen(x,y) _popen(x,y) -#define pclose(x) _pclose(x) -#define setenv(x,y,z) _putenv_s(x,y) -#endif +#include "test_common.h" using namespace std; diff --git a/projects/hip/tests/src/test_common.h b/projects/hip/tests/src/test_common.h index 8820381826..78aa845076 100644 --- a/projects/hip/tests/src/test_common.h +++ b/projects/hip/tests/src/test_common.h @@ -95,6 +95,14 @@ THE SOFTWARE. } \ } +#ifdef _WIN64 +#include +#define aligned_alloc _aligned_malloc +#define popen(x,y) _popen(x,y) +#define pclose(x) _pclose(x) +#define setenv(x,y,z) _putenv_s(x,y) +#endif + // standard command-line variables: extern size_t N; extern char memsetval; From a984acf2452728fb6c2a66615c6ddcc99caa5bc2 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Thu, 15 Aug 2019 19:13:16 -0700 Subject: [PATCH 02/14] Add hipMemcpy3DAsync (#1320) * Add hipMemcpy3DAsync * Fix CI build error * Move back stream resolution to internal function * Remove stream redefinition and check [ROCm/hip commit: 2405621f62716e748bed32c97241746b0f19870c] --- .../include/hip/hcc_detail/hip_runtime_api.h | 13 ++++++++ .../include/hip/nvcc_detail/hip_runtime_api.h | 5 +++ projects/hip/src/hip_memory.cpp | 31 +++++++++++++++---- 3 files changed, 43 insertions(+), 6 deletions(-) diff --git a/projects/hip/include/hip/hcc_detail/hip_runtime_api.h b/projects/hip/include/hip/hcc_detail/hip_runtime_api.h index 8db9e23317..97ff4bd1d7 100644 --- a/projects/hip/include/hip/hcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/hcc_detail/hip_runtime_api.h @@ -2022,6 +2022,19 @@ hipError_t hipMemcpyHtoA(hipArray* dstArray, size_t dstOffset, const void* srcHo */ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p); +/** + * @brief Copies data between host and device asynchronously. + * + * @param[in] p 3D memory copy parameters + * @param[in] stream Stream to use + * @return #hipSuccess, #hipErrorInvalidValue, #hipErrorInvalidPitchValue, + * #hipErrorInvalidDevicePointer, #hipErrorInvalidMemcpyDirection + * + * @see hipMemcpy, hipMemcpy2DToArray, hipMemcpy2D, hipMemcpyFromArray, hipMemcpyToSymbol, + * hipMemcpyAsync + */ +hipError_t hipMemcpy3DAsync(const struct hipMemcpy3DParms* p, hipStream_t stream __dparm(0)); + // doxygen end Memory /** * @} diff --git a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h index e1ff25d9f2..33dbcb82f3 100644 --- a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h @@ -607,6 +607,11 @@ inline static hipError_t hipMemcpy3D(const struct hipMemcpy3DParms *p) return hipCUDAErrorTohipError(cudaMemcpy3D(p)); } +inline static hipError_t hipMemcpy3DAsync(const struct hipMemcpy3DParms *p, hipStream_t stream) +{ + return hipCUDAErrorTohipError(cudaMemcpy3DAsync(p, stream)); +} + inline static hipError_t hipMemcpy2DAsync(void* dst, size_t dpitch, const void* src, size_t spitch, size_t width, size_t height, hipMemcpyKind kind, hipStream_t stream) { diff --git a/projects/hip/src/hip_memory.cpp b/projects/hip/src/hip_memory.cpp index ccea09bb39..d7cd9f32d8 100644 --- a/projects/hip/src/hip_memory.cpp +++ b/projects/hip/src/hip_memory.cpp @@ -1387,10 +1387,9 @@ hipError_t hipMemcpyAtoH(void* dst, hipArray* srcArray, size_t srcOffset, size_t return ihipLogStatus(e); } -hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) { - HIP_INIT_SPECIAL_API(hipMemcpy3D, (TRACE_MCMD), p); +hipError_t ihipMemcpy3D(const struct hipMemcpy3DParms* p, hipStream_t stream, bool isAsync) { hipError_t e = hipSuccess; - if (p) { + if(p) { size_t byteSize; size_t depth; size_t height; @@ -1448,11 +1447,14 @@ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) { ySize = p->srcPtr.ysize; dstPitch = p->dstPtr.pitch; } - hipStream_t stream = ihipSyncAndResolveStream(hipStreamNull); + stream = ihipSyncAndResolveStream(stream); hc::completion_future marker; try { if((widthInBytes == dstPitch) && (widthInBytes == srcPitch)) { - stream->locked_copySync((void*)dstPtr, (void*)srcPtr, widthInBytes*height*depth, p->kind, false); + if(isAsync) + stream->locked_copyAsync((void*)dstPtr, (void*)srcPtr, widthInBytes*height*depth, p->kind); + else + stream->locked_copySync((void*)dstPtr, (void*)srcPtr, widthInBytes*height*depth, p->kind, false); } else { for (int i = 0; i < depth; i++) { for (int j = 0; j < height; j++) { @@ -1461,7 +1463,10 @@ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) { (unsigned char*)srcPtr + i * ySize * srcPitch + j * srcPitch; unsigned char* dst = (unsigned char*)dstPtr + i * height * dstPitch + j * dstPitch; - stream->locked_copySync(dst, src, widthInBytes, p->kind); + if(isAsync) + stream->locked_copyAsync(dst, src, widthInBytes, p->kind); + else + stream->locked_copySync(dst, src, widthInBytes, p->kind); } } } @@ -1471,6 +1476,20 @@ hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) { } else { e = hipErrorInvalidValue; } + return e; +} + +hipError_t hipMemcpy3D(const struct hipMemcpy3DParms* p) { + HIP_INIT_SPECIAL_API(hipMemcpy3D, (TRACE_MCMD), p); + hipError_t e = hipSuccess; + e = ihipMemcpy3D(p, hipStreamNull, false); + return ihipLogStatus(e); +} + +hipError_t hipMemcpy3DAsync(const struct hipMemcpy3DParms* p, hipStream_t stream) { + HIP_INIT_SPECIAL_API(hipMemcpy3DAsync, (TRACE_MCMD), p, stream); + hipError_t e = hipSuccess; + e = ihipMemcpy3D(p, stream, true); return ihipLogStatus(e); } From 4f0fb9236e892fc5474fb506f3c83ed53bec0928 Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Thu, 15 Aug 2019 22:13:33 -0400 Subject: [PATCH 03/14] Fix assert for windows. (#1329) MSVC assert.h has no guard for include once. The macro assert overrides device assert definition. Do not include it for device compilation. [ROCm/hip commit: 7aa7a4ce222a4ee401016a885c7d70054f56f60d] --- projects/hip/include/hip/hcc_detail/math_functions.h | 7 +++++++ 1 file changed, 7 insertions(+) diff --git a/projects/hip/include/hip/hcc_detail/math_functions.h b/projects/hip/include/hip/hcc_detail/math_functions.h index c384c5573d..19def9ec7e 100644 --- a/projects/hip/include/hip/hcc_detail/math_functions.h +++ b/projects/hip/include/hip/hcc_detail/math_functions.h @@ -29,7 +29,14 @@ THE SOFTWARE. #include #include + +// assert.h is only for the host version of assert. +// The device version of assert is implemented in hip/hcc_detail/hip_runtime.h. +// Users should include hip_runtime.h for the device version of assert. +#if !__HIP_DEVICE_COMPILE__ #include +#endif + #include #include #include From 74443e34abb2b8d57212eb9b776a9a5cf895b840 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Thu, 15 Aug 2019 19:13:45 -0700 Subject: [PATCH 04/14] Fix build issues seen with hip-clang path (#1331) [ROCm/hip commit: 47b832bba7a63eaa4e3c056c5071d795b86123f9] --- projects/hip/src/hip_clang.cpp | 8 +++++++- 1 file changed, 7 insertions(+), 1 deletion(-) diff --git a/projects/hip/src/hip_clang.cpp b/projects/hip/src/hip_clang.cpp index 85dfb09f62..4e36b9323a 100644 --- a/projects/hip/src/hip_clang.cpp +++ b/projects/hip/src/hip_clang.cpp @@ -29,6 +29,10 @@ THE SOFTWARE. #include "hip_fatbin.h" #include "trace_helper.h" +#ifdef __GNUC__ +#pragma GCC visibility push (default) +#endif + extern "C" std::vector* __hipRegisterFatBinary(const void* data) { @@ -249,4 +253,6 @@ hipError_t hipLaunchByPtr(const void *hostFunction) return ihipLogStatus(e); } - +#ifdef __GNUC__ +#pragma GCC visibility pop +#endif From 1ddc06c042fd8376da2a63a1e4695c22a9ee837a Mon Sep 17 00:00:00 2001 From: mhbliao <47895780+mhbliao@users.noreply.github.com> Date: Thu, 15 Aug 2019 22:13:59 -0400 Subject: [PATCH 05/14] [hip] Allow from/to half conversion on host side. (#1334) [ROCm/hip commit: e919a8246e588ca75e6fc83b1e6bbf866eb94cdf] --- projects/hip/include/hip/hcc_detail/hip_fp16.h | 5 ++--- 1 file changed, 2 insertions(+), 3 deletions(-) diff --git a/projects/hip/include/hip/hcc_detail/hip_fp16.h b/projects/hip/include/hip/hcc_detail/hip_fp16.h index 74424a9f8b..11e1b6be11 100644 --- a/projects/hip/include/hip/hcc_detail/hip_fp16.h +++ b/projects/hip/include/hip/hcc_detail/hip_fp16.h @@ -106,7 +106,7 @@ THE SOFTWARE. #if !defined(__HIP_NO_HALF_CONVERSIONS__) template< typename T, Enable_if_t{}>* = nullptr> - __device__ + __host__ __device__ __half(T x) : data{static_cast<_Float16>(x)} {} #endif @@ -233,11 +233,10 @@ THE SOFTWARE. return __half_raw{data}; } - // ACCESSORS - DEVICE ONLY #if !defined(__HIP_NO_HALF_CONVERSIONS__) template< typename T, Enable_if_t{}>* = nullptr> - __device__ + __host__ __device__ operator T() const { return data; } #endif From aae6ad3ab4f79e14e5a871d8b8de663540b3ebb8 Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Fri, 16 Aug 2019 07:57:24 -0400 Subject: [PATCH 06/14] Fix missing decl for hip-clang Add back decl for hipHccModuleLaunchKernel and hipExtModuleLaunchKernel for HIP/VDI only [ROCm/hip commit: 51f0b3f3a61a85be94565e7063a3fa6969722279] --- .../hip/include/hip/hcc_detail/hip_runtime.h | 19 +++++++++++++++++++ 1 file changed, 19 insertions(+) diff --git a/projects/hip/include/hip/hcc_detail/hip_runtime.h b/projects/hip/include/hip/hcc_detail/hip_runtime.h index 921d9951d8..2d35d189bf 100644 --- a/projects/hip/include/hip/hcc_detail/hip_runtime.h +++ b/projects/hip/include/hip/hcc_detail/hip_runtime.h @@ -462,6 +462,25 @@ hc_get_workitem_absolute_id(int dim) #undef __CUDA__ #pragma pop_macro("__CUDA__") +#if __HIP_VDI__ +hipError_t hipExtModuleLaunchKernel(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, + uint32_t flags = 0); + +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) + __attribute__((deprecated("use hipExtModuleLaunchKernel instead"))); +#endif // __HIP_VDI__ #endif // defined(__clang__) && defined(__HIP__) #include From e181fd62155075cd8c3a7b3e6e14eb010c7372eb Mon Sep 17 00:00:00 2001 From: Maneesh Gupta Date: Tue, 20 Aug 2019 12:05:30 +0000 Subject: [PATCH 07/14] [ci] Enable tests on ROCm 2.7 (#1337) Change-Id: I06fb58b48af7f52f309f659969bb42bf9e8e9faf [ROCm/hip commit: 24c3d5b5ad9d4db15e05f939f9e6904af4f449ac] --- projects/hip/Jenkinsfile | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/projects/hip/Jenkinsfile b/projects/hip/Jenkinsfile index c7ac3e5346..e07cf349c4 100644 --- a/projects/hip/Jenkinsfile +++ b/projects/hip/Jenkinsfile @@ -295,13 +295,13 @@ def docker_upload_dockerhub( String local_org, String image_name, String remote_ String build_config = 'Release' String job_name = env.JOB_NAME.toLowerCase( ) -// The following launches 3 builds in parallel: rocm-head, rocm-2.6.x and cuda-10.x -parallel rocm_2_6: +// The following launches 3 builds in parallel: rocm-head, rocm-2.7.x and cuda-10.x +parallel rocm_2_7: { node('hip-rocm') { - String hcc_ver = 'rocm-2.6.x' - String from_image = 'ci_test_nodes/rocm-2.6.x/ubuntu-16.04:latest' + String hcc_ver = 'rocm-2.7.x' + String from_image = 'ci_test_nodes/rocm-2.7.x/ubuntu-16.04:latest' String inside_args = '--device=/dev/kfd --device=/dev/dri --group-add=video' // Checkout source code, dependencies and version files From 3f704a330346f8c4bc043ff3ebde97506a4b21f7 Mon Sep 17 00:00:00 2001 From: Sarbojit2019 <52527887+SarbojitAMD@users.noreply.github.com> Date: Tue, 20 Aug 2019 17:36:22 +0530 Subject: [PATCH 08/14] [hipcc] Fix for uninitialized variables on nvcc path Fixes #1186 [ROCm/hip commit: 938502dfcbc2b1972aa018629cf7c7d9092fd03e] --- projects/hip/bin/hipcc | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/projects/hip/bin/hipcc b/projects/hip/bin/hipcc index e8bca65663..4be1c084fb 100755 --- a/projects/hip/bin/hipcc +++ b/projects/hip/bin/hipcc @@ -107,7 +107,7 @@ $HIP_COMPILER= $hipConfig{'HIP_COMPILER'}; $HIP_RUNTIME= $hipConfig{'HIP_RUNTIME'}; # If using VDI runtime, need to find HIP_VDI_HOME -if ($HIP_RUNTIME eq "VDI" and !defined $HIP_VDI_HOME) { +if (defined $HIP_RUNTIME and $HIP_RUNTIME eq "VDI" and !defined $HIP_VDI_HOME) { my $hipcc_dir = dirname($0); if (-e "$hipcc_dir/../lib/bitcode") { $HIP_VDI_HOME = abs_path($hipcc_dir . "/.."); @@ -133,7 +133,7 @@ if (defined $HIP_VDI_HOME) { } } -if ($HIP_COMPILER eq "clang") { +if (defined $HIP_COMPILER and $HIP_COMPILER eq "clang") { $HIP_PLATFORM = "clang"; if (!defined $HIP_CLANG_PATH) { $HIP_CLANG_PATH = "/opt/rocm/llvm/bin"; From 612833bcfe53329d99dc44ea9f684e511b315ab6 Mon Sep 17 00:00:00 2001 From: kpyzhov Date: Tue, 20 Aug 2019 08:06:36 -0400 Subject: [PATCH 09/14] Corrected declaration of __ockl_clz_u64() (#1340) [ROCm/hip commit: 0e3198be250a17952c233e591f58d25aeeb4f86e] --- projects/hip/include/hip/hcc_detail/device_functions.h | 2 +- projects/hip/include/hip/hcc_detail/device_library_decls.h | 3 ++- 2 files changed, 3 insertions(+), 2 deletions(-) diff --git a/projects/hip/include/hip/hcc_detail/device_functions.h b/projects/hip/include/hip/hcc_detail/device_functions.h index 465655680f..684449aa6b 100644 --- a/projects/hip/include/hip/hcc_detail/device_functions.h +++ b/projects/hip/include/hip/hcc_detail/device_functions.h @@ -50,7 +50,7 @@ __device__ static inline int __clz(int input) { } __device__ static inline int __clzll(long long int input) { - return __ockl_clz_u64((ulong)input); + return __ockl_clz_u64((ullong)input); } __device__ static inline unsigned int __ffs(unsigned int input) { diff --git a/projects/hip/include/hip/hcc_detail/device_library_decls.h b/projects/hip/include/hip/hcc_detail/device_library_decls.h index 2bf3c8cc51..4e9772c80a 100644 --- a/projects/hip/include/hip/hcc_detail/device_library_decls.h +++ b/projects/hip/include/hip/hcc_detail/device_library_decls.h @@ -34,6 +34,7 @@ typedef unsigned char uchar; typedef unsigned short ushort; typedef unsigned int uint; typedef unsigned long ulong; +typedef unsigned long long ullong; extern "C" __device__ __attribute__((const)) bool __ockl_wfany_i32(int); extern "C" __device__ __attribute__((const)) bool __ockl_wfall_i32(int); @@ -48,7 +49,7 @@ extern "C" __device__ __attribute__((const)) uint __ockl_sad_u32(uint, uint, uin extern "C" __device__ __attribute__((const)) uchar __ockl_clz_u8(uchar); extern "C" __device__ __attribute__((const)) ushort __ockl_clz_u16(ushort); extern "C" __device__ __attribute__((const)) uint __ockl_clz_u32(uint); -extern "C" __device__ __attribute__((const)) ulong __ockl_clz_u64(ulong); +extern "C" __device__ __attribute__((const)) ullong __ockl_clz_u64(ullong); extern "C" __device__ __attribute__((const)) float __ocml_floor_f32(float); extern "C" __device__ __attribute__((const)) float __ocml_rint_f32(float); From 7dca0455e9351083b45e513d1f36a360e3fbae3f Mon Sep 17 00:00:00 2001 From: Jatin Chaudhary <51944368+cjatin@users.noreply.github.com> Date: Wed, 21 Aug 2019 15:30:10 +0530 Subject: [PATCH 10/14] Using HSA API for hipMemsetAsync (#1346) [ROCm/hip commit: ac62d7a5c031994279bd8b01d1b6911ac0e472da] --- projects/hip/src/hip_memory.cpp | 7 +++++++ 1 file changed, 7 insertions(+) diff --git a/projects/hip/src/hip_memory.cpp b/projects/hip/src/hip_memory.cpp index d7cd9f32d8..66f776e966 100644 --- a/projects/hip/src/hip_memory.cpp +++ b/projects/hip/src/hip_memory.cpp @@ -1557,6 +1557,13 @@ hipError_t ihipMemPtrGetInfo(void* ptr, size_t* size) { template void ihipMemsetKernel(hipStream_t stream, T* ptr, T val, size_t count) { + // Just Use count, instead of dividing by 4, the calling API already does it + if (sizeof(T) == sizeof(uint32_t) && (count % sizeof(uint32_t) == 0) && + !hsa_amd_memory_fill(ptr, reinterpret_cast(val), count)) { + // Only return if the execution completes without error + // if error occured, try the normal version + return; + } static constexpr uint32_t block_dim = 256; const uint32_t grid_dim = clamp_integer(count / block_dim, 1, UINT32_MAX); From edb699bd327b1defce4ef4f535f823f617c031cb Mon Sep 17 00:00:00 2001 From: Sarbojit2019 <52527887+SarbojitAMD@users.noreply.github.com> Date: Wed, 21 Aug 2019 15:31:03 +0530 Subject: [PATCH 11/14] Compilation failure on nvcc path when using hipChannelFormatKind (#1345) Fix for github #1183 issue reported [ROCm/hip commit: 84de192c9b3b811f101e134e2338fb02ebaca1b6] --- .../hip/include/hip/nvcc_detail/hip_runtime_api.h | 12 ++++++------ 1 file changed, 6 insertions(+), 6 deletions(-) diff --git a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h index 33dbcb82f3..532e1d827e 100644 --- a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h @@ -67,12 +67,12 @@ typedef enum hipMemcpyKind { #define hipReadModeElementType cudaReadModeElementType #define hipReadModeNormalizedFloat cudaReadModeNormalizedFloat -typedef enum hipChannelFormatKind { - hipChannelFormatKindSigned = 0, - hipChannelFormatKindUnsigned = 1, - hipChannelFormatKindFloat = 2, - hipChannelFormatKindNone = 3 -} hipChannelFormatKind; +// hipChannelFormatKind +#define hipChannelFormatKind cudaChannelFormatKind +#define hipChannelFormatKindSigned cudaChannelFormatKindSigned +#define hipChannelFormatKindUnsigned cudaChannelFormatKindUnsigned +#define hipChannelFormatKindFloat cudaChannelFormatKindFloat +#define hipChannelFormatKindNone cudaChannelFormatKindNone #define hipSurfaceBoundaryMode cudaSurfaceBoundaryMode #define hipBoundaryModeZero cudaBoundaryModeZero From f1f202a4e9e5abe3876deffbc7ee4f14b89ede3d Mon Sep 17 00:00:00 2001 From: Maneesh Gupta Date: Wed, 21 Aug 2019 10:01:23 +0000 Subject: [PATCH 12/14] Packaging and versioning changes (#1348) - Fixes SWDEV-200435 - Brings versioning and packaging naming inline with rest of the ROCm components. - Minor bug fixes. Change-Id: I543a24bf18baffe74fa3ccac8c93bddba7231c5e [ROCm/hip commit: c42075715b6e720ea513b715c7acf21d070fa9ce] --- projects/hip/CMakeLists.txt | 23 ++++++++++++++++++++--- projects/hip/packaging/hip_doc.txt | 3 ++- projects/hip/packaging/hip_hcc.txt | 5 +++-- projects/hip/packaging/hip_nvcc.txt | 3 ++- projects/hip/packaging/hip_samples.txt | 3 ++- 5 files changed, 29 insertions(+), 8 deletions(-) diff --git a/projects/hip/CMakeLists.txt b/projects/hip/CMakeLists.txt index 85715e5791..2c994e77e9 100644 --- a/projects/hip/CMakeLists.txt +++ b/projects/hip/CMakeLists.txt @@ -44,12 +44,25 @@ execute_process(COMMAND git rev-parse --short HEAD OUTPUT_VARIABLE HIP_VERSION_GITHASH OUTPUT_STRIP_TRAILING_WHITESPACE) +# get commit count +execute_process(COMMAND git rev-list --count HEAD + WORKING_DIRECTORY ${PROJECT_SOURCE_DIR} + OUTPUT_VARIABLE HIP_VERSION_GITCOUNT + OUTPUT_STRIP_TRAILING_WHITESPACE) + set(HIP_VERSION_PATCH ${HIP_VERSION_GITDATE}-${HIP_VERSION_GITHASH}) -set(HIP_VERSION ${HIP_VERSION_MAJOR}.${HIP_VERSION_MINOR}.${HIP_VERSION_PATCH}) add_to_config(_versionInfo HIP_VERSION_MAJOR) add_to_config(_versionInfo HIP_VERSION_MINOR) add_to_config(_versionInfo HIP_VERSION_PATCH) +# overwrite HIP_VERSION_PATCH for packaging +if(DEFINED ENV{ROCM_BUILD_ID}) + set(HIP_VERSION_PATCH ${HIP_VERSION_GITDATE}.${HIP_VERSION_GITCOUNT}-$ENV{ROCM_BUILD_ID}-${HIP_VERSION_GITHASH}) +else() + set(HIP_VERSION_PATCH ${HIP_VERSION_GITDATE}.${HIP_VERSION_GITCOUNT}-${HIP_VERSION_GITHASH}) +endif() +set(HIP_VERSION ${HIP_VERSION_MAJOR}.${HIP_VERSION_MINOR}.${HIP_VERSION_PATCH}) + ############################# # Configure variables ############################# @@ -424,7 +437,7 @@ add_custom_target(pkg_hip_hcc COMMAND ${CMAKE_COMMAND} . COMMAND cp *.rpm ${PROJECT_BINARY_DIR} COMMAND cp *.tar.gz ${PROJECT_BINARY_DIR} WORKING_DIRECTORY ${BUILD_DIR} - DEPENDS hip_hcc hip_hcc_static) + DEPENDS hip_hcc hip_hcc_static hiprtc) # Package: hip_nvcc set(BUILD_DIR ${CMAKE_CURRENT_BINARY_DIR}/packages/hip_nvcc) @@ -464,7 +477,11 @@ if(POLICY CMP0037) cmake_policy(PUSH) cmake_policy(SET CMP0037 OLD) endif() -add_custom_target(package DEPENDS pkg_hip_base pkg_hip_hcc pkg_hip_nvcc pkg_hip_doc pkg_hip_samples) +add_custom_target(package + COMMAND rename -v "'s/([a-z0-9_.\-]).deb/$1-amd64.deb/'" *.deb + COMMAND rename -v "'s/([a-z0-9_.\-]).rpm/$1.x86_64.rpm/'" *.rpm + WORKING_DIRECTORY ${PROJECT_BINARY_DIR} + DEPENDS pkg_hip_base pkg_hip_hcc pkg_hip_nvcc pkg_hip_doc pkg_hip_samples) if(POLICY CMP0037) cmake_policy(POP) endif() diff --git a/projects/hip/packaging/hip_doc.txt b/projects/hip/packaging/hip_doc.txt index 5987c2c128..2c1619f221 100644 --- a/projects/hip/packaging/hip_doc.txt +++ b/projects/hip/packaging/hip_doc.txt @@ -35,7 +35,8 @@ set(CPACK_DEBIAN_PACKAGE_DEPENDS "hip_base (= ${CPACK_PACKAGE_VERSION})") set(CPACK_BINARY_RPM "ON") set(CPACK_RPM_PACKAGE_ARCHITECTURE "${CMAKE_SYSTEM_PROCESSOR}") set(CPACK_RPM_PACKAGE_AUTOREQPROV " no") -set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${CPACK_PACKAGE_VERSION}") +string(REPLACE "-" "_" HIP_BASE_VERSION ${CPACK_PACKAGE_VERSION}) +set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${HIP_BASE_VERSION}") set(CPACK_RPM_EXCLUDE_FROM_AUTO_FILELIST_ADDITION "/opt") set(CPACK_SOURCE_GENERATOR "TGZ") include(CPack) diff --git a/projects/hip/packaging/hip_hcc.txt b/projects/hip/packaging/hip_hcc.txt index e2f912c7ca..c688433cc0 100644 --- a/projects/hip/packaging/hip_hcc.txt +++ b/projects/hip/packaging/hip_hcc.txt @@ -42,10 +42,11 @@ set(CPACK_RPM_PACKAGE_ARCHITECTURE "${CMAKE_SYSTEM_PROCESSOR}") set(CPACK_RPM_POST_INSTALL_SCRIPT_FILE "${PROJECT_BINARY_DIR}/postinst") set(CPACK_RPM_PRE_UNINSTALL_SCRIPT_FILE "${PROJECT_BINARY_DIR}/prerm") set(CPACK_RPM_PACKAGE_AUTOREQPROV " no") +string(REPLACE "-" "_" HIP_BASE_VERSION ${CPACK_PACKAGE_VERSION}) if(@COMPILE_HIP_ATP_MARKER@) - set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${CPACK_PACKAGE_VERSION}, ${HCC_PACKAGE_NAME} = @HCC_PACKAGE_VERSION@, rocm-profiler, comgr >= 1.1") + set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${HIP_BASE_VERSION}, ${HCC_PACKAGE_NAME} = @HCC_PACKAGE_VERSION@, rocm-profiler, comgr >= 1.1") else() - set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${CPACK_PACKAGE_VERSION}, ${HCC_PACKAGE_NAME} = @HCC_PACKAGE_VERSION@, comgr >= 1.1") + set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${HIP_BASE_VERSION}, ${HCC_PACKAGE_NAME} = @HCC_PACKAGE_VERSION@, comgr >= 1.1") endif() set(CPACK_RPM_EXCLUDE_FROM_AUTO_FILELIST_ADDITION "/opt") set(CPACK_SOURCE_GENERATOR "TGZ") diff --git a/projects/hip/packaging/hip_nvcc.txt b/projects/hip/packaging/hip_nvcc.txt index 254b7a956a..975176f934 100644 --- a/projects/hip/packaging/hip_nvcc.txt +++ b/projects/hip/packaging/hip_nvcc.txt @@ -24,7 +24,8 @@ set(CPACK_RPM_PACKAGE_ARCHITECTURE "${CMAKE_SYSTEM_PROCESSOR}") #set(CPACK_RPM_POST_INSTALL_SCRIPT_FILE "${PROJECT_BINARY_DIR}/postinst") #set(CPACK_RPM_PRE_UNINSTALL_SCRIPT_FILE "${PROJECT_BINARY_DIR}/prerm") set(CPACK_RPM_PACKAGE_AUTOREQPROV " no") -set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${CPACK_PACKAGE_VERSION}, cuda >= 7.5") +string(REPLACE "-" "_" HIP_BASE_VERSION ${CPACK_PACKAGE_VERSION}) +set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${HIP_BASE_VERSION}, cuda >= 7.5") set(CPACK_RPM_EXCLUDE_FROM_AUTO_FILELIST_ADDITION "/opt") set(CPACK_SOURCE_GENERATOR "TGZ") include(CPack) diff --git a/projects/hip/packaging/hip_samples.txt b/projects/hip/packaging/hip_samples.txt index c1707e42f1..4c1e04cb23 100644 --- a/projects/hip/packaging/hip_samples.txt +++ b/projects/hip/packaging/hip_samples.txt @@ -23,7 +23,8 @@ set(CPACK_DEBIAN_PACKAGE_DEPENDS "hip_base (= ${CPACK_PACKAGE_VERSION})") set(CPACK_BINARY_RPM "ON") set(CPACK_RPM_PACKAGE_ARCHITECTURE "${CMAKE_SYSTEM_PROCESSOR}") set(CPACK_RPM_PACKAGE_AUTOREQPROV " no") -set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${CPACK_PACKAGE_VERSION}") +string(REPLACE "-" "_" HIP_BASE_VERSION ${CPACK_PACKAGE_VERSION}) +set(CPACK_RPM_PACKAGE_REQUIRES "hip_base = ${HIP_BASE_VERSION}") set(CPACK_RPM_EXCLUDE_FROM_AUTO_FILELIST_ADDITION "/opt") set(CPACK_SOURCE_GENERATOR "TGZ") include(CPack) From 7509ffa68d34a31020c855b0a36191a3c0b81430 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Wed, 21 Aug 2019 03:01:31 -0700 Subject: [PATCH 13/14] Remove -g from hipBusBandwidth makefile (#1351) [ROCm/hip commit: b3a0d4266bee51f869ac99f51c656e800bdf3612] --- projects/hip/samples/1_Utils/hipBusBandwidth/Makefile | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/projects/hip/samples/1_Utils/hipBusBandwidth/Makefile b/projects/hip/samples/1_Utils/hipBusBandwidth/Makefile index 418f25f8ed..45665adc45 100644 --- a/projects/hip/samples/1_Utils/hipBusBandwidth/Makefile +++ b/projects/hip/samples/1_Utils/hipBusBandwidth/Makefile @@ -5,7 +5,7 @@ endif HIPCC=$(HIP_PATH)/bin/hipcc EXE=hipBusBandwidth -CXXFLAGS = -O3 -g +CXXFLAGS = -O3 all: install From 83ad27b7efd20040064704347a40635bc74034f2 Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Wed, 21 Aug 2019 21:06:20 +0300 Subject: [PATCH 14/14] [HIPIFY] CUDA 10.1 Update 2 support [ROCm/hip commit: 54fee7abdc9085036641d8adb5576e6497e3f5ce] --- .../markdown/CUBLAS_API_supported_by_HIP.md | 5 + .../markdown/CURAND_API_supported_by_HIP.md | 223 +++++++++--------- .../markdown/CUSPARSE_API_supported_by_HIP.md | 5 + .../src/CUDA2HIP_BLAS_API_types.cpp | 7 + .../src/CUDA2HIP_RAND_API_types.cpp | 11 +- .../src/CUDA2HIP_SPARSE_API_types.cpp | 6 + 6 files changed, 147 insertions(+), 110 deletions(-) diff --git a/projects/hip/docs/markdown/CUBLAS_API_supported_by_HIP.md b/projects/hip/docs/markdown/CUBLAS_API_supported_by_HIP.md index 62e014c944..56fc2a89ac 100644 --- a/projects/hip/docs/markdown/CUBLAS_API_supported_by_HIP.md +++ b/projects/hip/docs/markdown/CUBLAS_API_supported_by_HIP.md @@ -4,6 +4,11 @@ | **type** | **CUDA** |**CUDA version\***| **HIP** |**HIP value** (if differs) | |-------------:|---------------------------------------------------------------|:----------------:|------------------------------------------------------------|---------------------------| +| define |`CUBLAS_VER_MAJOR` | 10.1 Update 2 | | +| define |`CUBLAS_VER_MINOR` | 10.1 Update 2 | | +| define |`CUBLAS_VER_PATCH` | 10.1 Update 2 | | +| define |`CUBLAS_VER_BUILD` | 10.1 Update 2 | | +| define |`CUBLAS_VERSION` | 10.1 Update 2 | | | enum |***`cublasStatus`*** | |***`hipblasStatus_t`*** | | enum |***`cublasStatus_t`*** | |***`hipblasStatus_t`*** | | 0 |*`CUBLAS_STATUS_SUCCESS`* | |*`HIPBLAS_STATUS_SUCCESS`* | diff --git a/projects/hip/docs/markdown/CURAND_API_supported_by_HIP.md b/projects/hip/docs/markdown/CURAND_API_supported_by_HIP.md index 83c41e59ff..a46a2c1afd 100644 --- a/projects/hip/docs/markdown/CURAND_API_supported_by_HIP.md +++ b/projects/hip/docs/markdown/CURAND_API_supported_by_HIP.md @@ -2,115 +2,120 @@ ## **1. CURAND Data types** -| **type** | **CUDA** | **HIP** | **HIP value** (if differs) | -|-------------:|---------------------------------------------------------------|------------------------------------------------------------|----------------------------| -| enum |***`curandStatus`*** |***`hiprandStatus`*** | -| enum |***`curandStatus_t`*** |***`hiprandStatus_t`*** | -| 0 |*`CURAND_STATUS_SUCCESS`* |*`HIPRAND_STATUS_SUCCESS`* | -| 100 |*`CURAND_STATUS_VERSION_MISMATCH`* |*`HIPRAND_STATUS_VERSION_MISMATCH`* | -| 101 |*`CURAND_STATUS_NOT_INITIALIZED`* |*`HIPRAND_STATUS_NOT_INITIALIZED`* | -| 102 |*`CURAND_STATUS_ALLOCATION_FAILED`* |*`HIPRAND_STATUS_ALLOCATION_FAILED`* | -| 103 |*`CURAND_STATUS_TYPE_ERROR`* |*`HIPRAND_STATUS_TYPE_ERROR`* | -| 104 |*`CURAND_STATUS_OUT_OF_RANGE`* |*`HIPRAND_STATUS_OUT_OF_RANGE`* | -| 105 |*`CURAND_STATUS_LENGTH_NOT_MULTIPLE`* |*`HIPRAND_STATUS_LENGTH_NOT_MULTIPLE`* | -| 106 |*`CURAND_STATUS_DOUBLE_PRECISION_REQUIRED`* |*`HIPRAND_STATUS_DOUBLE_PRECISION_REQUIRED`* | -| 201 |*`CURAND_STATUS_LAUNCH_FAILURE`* |*`HIPRAND_STATUS_LAUNCH_FAILURE`* | -| 202 |*`CURAND_STATUS_PREEXISTING_FAILURE`* |*`HIPRAND_STATUS_PREEXISTING_FAILURE`* | -| 203 |*`CURAND_STATUS_INITIALIZATION_FAILED`* |*`HIPRAND_STATUS_INITIALIZATION_FAILED`* | -| 204 |*`CURAND_STATUS_ARCH_MISMATCH`* |*`HIPRAND_STATUS_ARCH_MISMATCH`* | -| 999 |*`CURAND_STATUS_INTERNAL_ERROR`* |*`HIPRAND_STATUS_INTERNAL_ERROR`* | -| enum |***`curandRngType`*** |***`hiprandRngType`*** | -| enum |***`curandRngType_t`*** |***`hiprandRngType_t`*** | -| 0 |*`CURAND_RNG_TEST`* |*`HIPRAND_RNG_TEST`* | -| 100 |*`CURAND_RNG_PSEUDO_DEFAULT`* |*`HIPRAND_RNG_PSEUDO_DEFAULT`* | 400 | -| 101 |*`CURAND_RNG_PSEUDO_XORWOW`* |*`HIPRAND_RNG_PSEUDO_XORWOW`* | 401 | -| 121 |*`CURAND_RNG_PSEUDO_MRG32K3A`* |*`HIPRAND_RNG_PSEUDO_MRG32K3A`* | 402 | -| 141 |*`CURAND_RNG_PSEUDO_MTGP32`* |*`HIPRAND_RNG_PSEUDO_MTGP32`* | 403 | -| 142 |*`CURAND_RNG_PSEUDO_MT19937`* |*`HIPRAND_RNG_PSEUDO_MT19937`* | 404 | -| 161 |*`CURAND_RNG_PSEUDO_PHILOX4_32_10`* |*`HIPRAND_RNG_PSEUDO_PHILOX4_32_10`* | 405 | -| 200 |*`CURAND_RNG_QUASI_DEFAULT`* |*`HIPRAND_RNG_QUASI_DEFAULT`* | 500 | -| 201 |*`CURAND_RNG_QUASI_SOBOL32`* |*`HIPRAND_RNG_QUASI_SOBOL32`* | 501 | -| 202 |*`CURAND_RNG_QUASI_SCRAMBLED_SOBOL32`* |*`HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL32`* | 502 | -| 203 |*`CURAND_RNG_QUASI_SOBOL64`* |*`HIPRAND_RNG_QUASI_SOBOL64`* | 503 | -| 204 |*`CURAND_RNG_QUASI_SCRAMBLED_SOBOL64`* |*`HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL64`* | 504 | -| enum |***`curandOrdering`*** | | -| enum |***`curandOrdering_t`*** | | -| 100 |*`CURAND_ORDERING_PSEUDO_BEST`* | | -| 101 |*`CURAND_ORDERING_PSEUDO_DEFAULT`* | | -| 102 |*`CURAND_ORDERING_PSEUDO_SEEDED`* | | -| 201 |*`CURAND_ORDERING_QUASI_DEFAULT`* | | -| enum |***`curandDirectionVectorSet`*** | | -| enum |***`curandDirectionVectorSet_t`*** | | -| 101 |*`CURAND_DIRECTION_VECTORS_32_JOEKUO6`* | | -| 102 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`* | | -| 103 |*`CURAND_DIRECTION_VECTORS_64_JOEKUO6`* | | -| 104 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`* | | -| uint | `curandDirectionVectors32_t` | `hiprandDirectionVectors32_t` | -| uint | `curandDirectionVectors64_t` | | -| struct | `curandGenerator_st` | `hiprandGenerator_st` | -| struct* | `curandGenerator_t` | `hiprandGenerator_t` | -| enum |***`curandOrdering`*** | | -| enum |***`curandOrdering_t`*** | | -| 100 |*`CURAND_ORDERING_PSEUDO_BEST`* | | -| 101 |*`CURAND_ORDERING_PSEUDO_DEFAULT`* | | -| 102 |*`CURAND_ORDERING_PSEUDO_SEEDED`* | | -| 201 |*`CURAND_ORDERING_QUASI_DEFAULT`* | | -| enum |***`curandDirectionVectorSet`*** | | -| enum |***`curandDirectionVectorSet_t`*** | | -| 101 |*`CURAND_DIRECTION_VECTORS_32_JOEKUO6`* | | -| 102 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`* | | -| 103 |*`CURAND_DIRECTION_VECTORS_64_JOEKUO6`* | | -| 104 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`* | | -| uint | `curandDirectionVectors32_t` | `hiprandDirectionVectors32_t` | -| uint | `curandDirectionVectors64_t` | | -| double | `curandDistribution_st` | | -| double | `curandHistogramM2V_st` | | -| double* | `curandDistribution_t` | | -| double* | `curandHistogramM2V_t` | | -| struct | `curandDistributionShift_st` | | -| struct* | `curandDistributionShift_t` | | -| struct | `curandDistributionM2Shift_st` | | -| struct* | `curandDistributionM2Shift_t` | | -| struct | `curandHistogramM2_st` | | -| struct* | `curandHistogramM2_t` | | -| uint | `curandHistogramM2K_st` | | -| uint* | `curandHistogramM2K_t` | | -| struct | `curandDiscreteDistribution_st` | `hiprandDiscreteDistribution_st` | -| struct* | `curandDiscreteDistribution_t` | `hiprandDiscreteDistribution_t` | -| enum |***`curandMethod`*** | | -| enum |***`curandMethod_t`*** | | -| 0 |*`CURAND_CHOOSE_BEST`* | | -| 1 |*`CURAND_ITR`* | | -| 2 |*`CURAND_KNUTH`* | | -| 3 |*`CURAND_HITR`* | | -| 4 |*`CURAND_M1`* | | -| 5 |*`CURAND_M2`* | | -| 6 |*`CURAND_BINARY_SEARCH`* | | -| 7 |*`CURAND_DISCRETE_GAUSS`* | | -| 8 |*`CURAND_REJECTION`* | | -| 9 |*`CURAND_DEVICE_API`* | | -| 10 |*`CURAND_FAST_REJECTION`* | | -| 11 |*`CURAND_3RD`* | | -| 12 |*`CURAND_DEFINITION`* | | -| 13 |*`CURAND_POISSON`* | | -| struct | `curandStateMtgp32` | `hiprandStateMtgp32` | -| typedef | `curandStateMtgp32_t` | `hiprandStateMtgp32_t` | -| struct | `curandStateScrambledSobol64` | | -| typedef | `curandStateScrambledSobol64_t` | | -| struct | `curandStateSobol64` | | -| typedef | `curandStateSobol64_t` | | -| struct | `curandStateScrambledSobol32` | | -| typedef | `curandStateScrambledSobol32_t` | | -| struct | `curandStateSobol32` | `hiprandStateSobol32` | -| typedef | `curandStateSobol32_t` | `hiprandStateSobol32_t` | -| struct | `curandStateMRG32k3a` | `hiprandStateMRG32k3a` | -| typedef | `curandStateMRG32k3a_t` | `hiprandStateMRG32k3a_t` | -| struct | `curandStatePhilox4_32_10` | `hiprandStatePhilox4_32_10` | -| typedef | `curandStatePhilox4_32_10_t` | `hiprandStatePhilox4_32_10_t` | -| struct | `curandStateXORWOW` | `hiprandStateXORWOW` | -| typedef | `curandStateXORWOW_t` | `hiprandStateXORWOW_t` | -| struct | `curandState` | `hiprandState` | -| typedef | `curandState_t` | `hiprandState_t` | +| **type** | **CUDA** |**CUDA version\***| **HIP** | **HIP value** (if differs) | +|-------------:|---------------------------------------------------------------|:----------------:|------------------------------------------------------------|----------------------------| +| define |`CURAND_VER_MAJOR` | 10.1 Update 2 | | +| define |`CURAND_VER_MINOR` | 10.1 Update 2 | | +| define |`CURAND_VER_PATCH` | 10.1 Update 2 | | +| define |`CURAND_VER_BUILD` | 10.1 Update 2 | | +| define |`CURAND_VERSION` | 10.1 Update 2 | | +| enum |***`curandStatus`*** | |***`hiprandStatus`*** | +| enum |***`curandStatus_t`*** | |***`hiprandStatus_t`*** | +| 0 |*`CURAND_STATUS_SUCCESS`* | |*`HIPRAND_STATUS_SUCCESS`* | +| 100 |*`CURAND_STATUS_VERSION_MISMATCH`* | |*`HIPRAND_STATUS_VERSION_MISMATCH`* | +| 101 |*`CURAND_STATUS_NOT_INITIALIZED`* | |*`HIPRAND_STATUS_NOT_INITIALIZED`* | +| 102 |*`CURAND_STATUS_ALLOCATION_FAILED`* | |*`HIPRAND_STATUS_ALLOCATION_FAILED`* | +| 103 |*`CURAND_STATUS_TYPE_ERROR`* | |*`HIPRAND_STATUS_TYPE_ERROR`* | +| 104 |*`CURAND_STATUS_OUT_OF_RANGE`* | |*`HIPRAND_STATUS_OUT_OF_RANGE`* | +| 105 |*`CURAND_STATUS_LENGTH_NOT_MULTIPLE`* | |*`HIPRAND_STATUS_LENGTH_NOT_MULTIPLE`* | +| 106 |*`CURAND_STATUS_DOUBLE_PRECISION_REQUIRED`* | |*`HIPRAND_STATUS_DOUBLE_PRECISION_REQUIRED`* | +| 201 |*`CURAND_STATUS_LAUNCH_FAILURE`* | |*`HIPRAND_STATUS_LAUNCH_FAILURE`* | +| 202 |*`CURAND_STATUS_PREEXISTING_FAILURE`* | |*`HIPRAND_STATUS_PREEXISTING_FAILURE`* | +| 203 |*`CURAND_STATUS_INITIALIZATION_FAILED`* | |*`HIPRAND_STATUS_INITIALIZATION_FAILED`* | +| 204 |*`CURAND_STATUS_ARCH_MISMATCH`* | |*`HIPRAND_STATUS_ARCH_MISMATCH`* | +| 999 |*`CURAND_STATUS_INTERNAL_ERROR`* | |*`HIPRAND_STATUS_INTERNAL_ERROR`* | +| enum |***`curandRngType`*** | |***`hiprandRngType`*** | +| enum |***`curandRngType_t`*** | |***`hiprandRngType_t`*** | +| 0 |*`CURAND_RNG_TEST`* | |*`HIPRAND_RNG_TEST`* | +| 100 |*`CURAND_RNG_PSEUDO_DEFAULT`* | |*`HIPRAND_RNG_PSEUDO_DEFAULT`* | 400 | +| 101 |*`CURAND_RNG_PSEUDO_XORWOW`* | |*`HIPRAND_RNG_PSEUDO_XORWOW`* | 401 | +| 121 |*`CURAND_RNG_PSEUDO_MRG32K3A`* | |*`HIPRAND_RNG_PSEUDO_MRG32K3A`* | 402 | +| 141 |*`CURAND_RNG_PSEUDO_MTGP32`* | |*`HIPRAND_RNG_PSEUDO_MTGP32`* | 403 | +| 142 |*`CURAND_RNG_PSEUDO_MT19937`* | |*`HIPRAND_RNG_PSEUDO_MT19937`* | 404 | +| 161 |*`CURAND_RNG_PSEUDO_PHILOX4_32_10`* | |*`HIPRAND_RNG_PSEUDO_PHILOX4_32_10`* | 405 | +| 200 |*`CURAND_RNG_QUASI_DEFAULT`* | |*`HIPRAND_RNG_QUASI_DEFAULT`* | 500 | +| 201 |*`CURAND_RNG_QUASI_SOBOL32`* | |*`HIPRAND_RNG_QUASI_SOBOL32`* | 501 | +| 202 |*`CURAND_RNG_QUASI_SCRAMBLED_SOBOL32`* | |*`HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL32`* | 502 | +| 203 |*`CURAND_RNG_QUASI_SOBOL64`* | |*`HIPRAND_RNG_QUASI_SOBOL64`* | 503 | +| 204 |*`CURAND_RNG_QUASI_SCRAMBLED_SOBOL64`* | |*`HIPRAND_RNG_QUASI_SCRAMBLED_SOBOL64`* | 504 | +| enum |***`curandOrdering`*** | | | +| enum |***`curandOrdering_t`*** | | | +| 100 |*`CURAND_ORDERING_PSEUDO_BEST`* | | | +| 101 |*`CURAND_ORDERING_PSEUDO_DEFAULT`* | | | +| 102 |*`CURAND_ORDERING_PSEUDO_SEEDED`* | | | +| 201 |*`CURAND_ORDERING_QUASI_DEFAULT`* | | | +| enum |***`curandDirectionVectorSet`*** | | | +| enum |***`curandDirectionVectorSet_t`*** | | | +| 101 |*`CURAND_DIRECTION_VECTORS_32_JOEKUO6`* | | | +| 102 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`* | | | +| 103 |*`CURAND_DIRECTION_VECTORS_64_JOEKUO6`* | | | +| 104 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`* | | | +| uint | `curandDirectionVectors32_t` | | `hiprandDirectionVectors32_t` | +| uint | `curandDirectionVectors64_t` | | | +| struct | `curandGenerator_st` | | `hiprandGenerator_st` | +| struct* | `curandGenerator_t` | | `hiprandGenerator_t` | +| enum |***`curandOrdering`*** | | | +| enum |***`curandOrdering_t`*** | | | +| 100 |*`CURAND_ORDERING_PSEUDO_BEST`* | | | +| 101 |*`CURAND_ORDERING_PSEUDO_DEFAULT`* | | | +| 102 |*`CURAND_ORDERING_PSEUDO_SEEDED`* | | | +| 201 |*`CURAND_ORDERING_QUASI_DEFAULT`* | | | +| enum |***`curandDirectionVectorSet`*** | | | +| enum |***`curandDirectionVectorSet_t`*** | | | +| 101 |*`CURAND_DIRECTION_VECTORS_32_JOEKUO6`* | | | +| 102 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_32_JOEKUO6`* | | | +| 103 |*`CURAND_DIRECTION_VECTORS_64_JOEKUO6`* | | | +| 104 |*`CURAND_SCRAMBLED_DIRECTION_VECTORS_64_JOEKUO6`* | | | +| uint | `curandDirectionVectors32_t` | | `hiprandDirectionVectors32_t` | +| uint | `curandDirectionVectors64_t` | | | +| double | `curandDistribution_st` | | | +| double | `curandHistogramM2V_st` | | | +| double* | `curandDistribution_t` | | | +| double* | `curandHistogramM2V_t` | | | +| struct | `curandDistributionShift_st` | | | +| struct* | `curandDistributionShift_t` | | | +| struct | `curandDistributionM2Shift_st` | | | +| struct* | `curandDistributionM2Shift_t` | | | +| struct | `curandHistogramM2_st` | | | +| struct* | `curandHistogramM2_t` | | | +| uint | `curandHistogramM2K_st` | | | +| uint* | `curandHistogramM2K_t` | | | +| struct | `curandDiscreteDistribution_st` | | `hiprandDiscreteDistribution_st` | +| struct* | `curandDiscreteDistribution_t` | | `hiprandDiscreteDistribution_t` | +| enum |***`curandMethod`*** | | | +| enum |***`curandMethod_t`*** | | | +| 0 |*`CURAND_CHOOSE_BEST`* | | | +| 1 |*`CURAND_ITR`* | | | +| 2 |*`CURAND_KNUTH`* | | | +| 3 |*`CURAND_HITR`* | | | +| 4 |*`CURAND_M1`* | | | +| 5 |*`CURAND_M2`* | | | +| 6 |*`CURAND_BINARY_SEARCH`* | | | +| 7 |*`CURAND_DISCRETE_GAUSS`* | | | +| 8 |*`CURAND_REJECTION`* | | | +| 9 |*`CURAND_DEVICE_API`* | | | +| 10 |*`CURAND_FAST_REJECTION`* | | | +| 11 |*`CURAND_3RD`* | | | +| 12 |*`CURAND_DEFINITION`* | | | +| 13 |*`CURAND_POISSON`* | | | +| struct | `curandStateMtgp32` | | `hiprandStateMtgp32` | +| typedef | `curandStateMtgp32_t` | | `hiprandStateMtgp32_t` | +| struct | `curandStateScrambledSobol64` | | | +| typedef | `curandStateScrambledSobol64_t` | | | +| struct | `curandStateSobol64` | | | +| typedef | `curandStateSobol64_t` | | | +| struct | `curandStateScrambledSobol32` | | | +| typedef | `curandStateScrambledSobol32_t` | | | +| struct | `curandStateSobol32` | | `hiprandStateSobol32` | +| typedef | `curandStateSobol32_t` | | `hiprandStateSobol32_t` | +| struct | `curandStateMRG32k3a` | | `hiprandStateMRG32k3a` | +| typedef | `curandStateMRG32k3a_t` | | `hiprandStateMRG32k3a_t` | +| struct | `curandStatePhilox4_32_10` | | `hiprandStatePhilox4_32_10` | +| typedef | `curandStatePhilox4_32_10_t` | | `hiprandStatePhilox4_32_10_t` | +| struct | `curandStateXORWOW` | | `hiprandStateXORWOW` | +| typedef | `curandStateXORWOW_t` | | `hiprandStateXORWOW_t` | +| struct | `curandState` | | `hiprandState` | +| typedef | `curandState_t` | | `hiprandState_t` | ## **2. Host API Functions** diff --git a/projects/hip/docs/markdown/CUSPARSE_API_supported_by_HIP.md b/projects/hip/docs/markdown/CUSPARSE_API_supported_by_HIP.md index f8a9d27b07..3a25acc20c 100644 --- a/projects/hip/docs/markdown/CUSPARSE_API_supported_by_HIP.md +++ b/projects/hip/docs/markdown/CUSPARSE_API_supported_by_HIP.md @@ -4,6 +4,11 @@ | **type** | **CUDA** |**CUDA version\***| **HIP** | |-------------:|---------------------------------------------------------------|:-----------------|------------------------------------------------------------| +| define |`CUSPARSE_VER_MAJOR` | 10.1 Update 2 | | +| define |`CUSPARSE_VER_MINOR` | 10.1 Update 2 | | +| define |`CUSPARSE_VER_PATCH` | 10.1 Update 2 | | +| define |`CUSPARSE_VER_BUILD` | 10.1 Update 2 | | +| define |`CUSPARSE_VERSION` | 10.1 Update 2 | | | enum |***`cusparseAction_t`*** | |***`hipsparseAction_t`*** | | 0 |*`CUSPARSE_ACTION_SYMBOLIC`* | |*`HIPSPARSE_ACTION_SYMBOLIC`* | | 1 |*`CUSPARSE_ACTION_NUMERIC`* | |*`HIPSPARSE_ACTION_NUMERIC`* | diff --git a/projects/hip/hipify-clang/src/CUDA2HIP_BLAS_API_types.cpp b/projects/hip/hipify-clang/src/CUDA2HIP_BLAS_API_types.cpp index 2e2168efa0..6785b1d41b 100644 --- a/projects/hip/hipify-clang/src/CUDA2HIP_BLAS_API_types.cpp +++ b/projects/hip/hipify-clang/src/CUDA2HIP_BLAS_API_types.cpp @@ -24,6 +24,13 @@ THE SOFTWARE. // Map of all functions const std::map CUDA_BLAS_TYPE_NAME_MAP{ + // Blas defines + {"CUBLAS_VER_MAJOR", {"HIPBLAS_VER_MAJOR", "", CONV_DEFINE, API_BLAS, HIP_UNSUPPORTED}}, + {"CUBLAS_VER_MINOR", {"HIPBLAS_VER_MINOR", "", CONV_DEFINE, API_BLAS, HIP_UNSUPPORTED}}, + {"CUBLAS_VER_PATCH", {"HIPBLAS_VER_PATCH", "", CONV_DEFINE, API_BLAS, HIP_UNSUPPORTED}}, + {"CUBLAS_VER_BUILD", {"HIPBLAS_VER_BUILD", "", CONV_DEFINE, API_BLAS, HIP_UNSUPPORTED}}, + {"CUBLAS_VERSION", {"HIPBLAS_VERSION", "", CONV_DEFINE, API_BLAS, HIP_UNSUPPORTED}}, + // Blas operations {"cublasOperation_t", {"hipblasOperation_t", "rocblas_operation", CONV_TYPE, API_BLAS}}, {"CUBLAS_OP_N", {"HIPBLAS_OP_N", "rocblas_operation_none", CONV_NUMERIC_LITERAL, API_BLAS}}, diff --git a/projects/hip/hipify-clang/src/CUDA2HIP_RAND_API_types.cpp b/projects/hip/hipify-clang/src/CUDA2HIP_RAND_API_types.cpp index e6eae1db69..67ecc93504 100644 --- a/projects/hip/hipify-clang/src/CUDA2HIP_RAND_API_types.cpp +++ b/projects/hip/hipify-clang/src/CUDA2HIP_RAND_API_types.cpp @@ -24,6 +24,14 @@ THE SOFTWARE. // Map of all functions const std::map CUDA_RAND_TYPE_NAME_MAP{ + // RAND defines + {"CURAND_VER_MAJOR", {"HIPRAND_VER_MAJOR", "", CONV_DEFINE, API_RAND, HIP_UNSUPPORTED}}, + {"CURAND_VER_MINOR", {"HIPRAND_VER_MINOR", "", CONV_DEFINE, API_RAND, HIP_UNSUPPORTED}}, + {"CURAND_VER_PATCH", {"HIPRAND_VER_PATCH", "", CONV_DEFINE, API_RAND, HIP_UNSUPPORTED}}, + {"CURAND_VER_BUILD", {"HIPRAND_VER_BUILD", "", CONV_DEFINE, API_RAND, HIP_UNSUPPORTED}}, + {"CURAND_VERSION", {"HIPRAND_VERSION", "", CONV_DEFINE, API_RAND, HIP_UNSUPPORTED}}, + + // RAND Host types {"curandStatus", {"hiprandStatus_t", "", CONV_TYPE, API_RAND}}, {"curandStatus_t", {"hiprandStatus_t", "", CONV_TYPE, API_RAND}}, {"curandRngType", {"hiprandRngType_t", "", CONV_TYPE, API_RAND}}, @@ -52,7 +60,8 @@ const std::map CUDA_RAND_TYPE_NAME_MAP{ {"curandMethod_t", {"hiprandMethod_t", "", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}}, {"curandDirectionVectors32_t", {"hiprandDirectionVectors32_t", "", CONV_TYPE, API_RAND}}, {"curandDirectionVectors64_t", {"hiprandDirectionVectors64_t", "", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}}, - // cuRAND types for Device functions + + // RAND types for Device functions {"curandStateMtgp32", {"hiprandStateMtgp32", "", CONV_TYPE, API_RAND}}, {"curandStateMtgp32_t", {"hiprandStateMtgp32_t", "", CONV_TYPE, API_RAND}}, {"curandStateScrambledSobol64", {"hiprandStateScrambledSobol64", "", CONV_TYPE, API_RAND, HIP_UNSUPPORTED}}, diff --git a/projects/hip/hipify-clang/src/CUDA2HIP_SPARSE_API_types.cpp b/projects/hip/hipify-clang/src/CUDA2HIP_SPARSE_API_types.cpp index 4a6e91288f..f9e3920af4 100644 --- a/projects/hip/hipify-clang/src/CUDA2HIP_SPARSE_API_types.cpp +++ b/projects/hip/hipify-clang/src/CUDA2HIP_SPARSE_API_types.cpp @@ -175,4 +175,10 @@ const std::map CUDA_SPARSE_TYPE_NAME_MAP{ {"CUSPARSE_INDEX_32I", {"HIPSPARSE_INDEX_32I", "", CONV_NUMERIC_LITERAL, API_SPARSE, HIP_UNSUPPORTED}}, {"CUSPARSE_INDEX_64I", {"HIPSPARSE_INDEX_64I", "", CONV_NUMERIC_LITERAL, API_SPARSE, HIP_UNSUPPORTED}}, + // 3. Defines + {"CUSPARSE_VER_MAJOR", {"HIPSPARSE_VER_MAJOR", "", CONV_DEFINE, API_SPARSE, HIP_UNSUPPORTED}}, + {"CUSPARSE_VER_MINOR", {"HIPSPARSE_VER_MINOR", "", CONV_DEFINE, API_SPARSE, HIP_UNSUPPORTED}}, + {"CUSPARSE_VER_PATCH", {"HIPSPARSE_VER_PATCH", "", CONV_DEFINE, API_SPARSE, HIP_UNSUPPORTED}}, + {"CUSPARSE_VER_BUILD", {"HIPSPARSE_VER_BUILD", "", CONV_DEFINE, API_SPARSE, HIP_UNSUPPORTED}}, + {"CUSPARSE_VERSION", {"HIPSPARSE_VERSION", "", CONV_DEFINE, API_SPARSE, HIP_UNSUPPORTED}}, };