From 0573ee79de3781d9a08bd3a6fd24724d9f8fbe4b Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Wed, 13 Jun 2018 10:01:14 -0400 Subject: [PATCH 01/22] Let hipcc handle static library for hip-clang only if it contains bundles [ROCm/hip commit: f0935e6d4f424c82f065c0c22ca5b48e6ed2099a] --- projects/hip/bin/hipcc | 21 +++++++++++++++++++-- 1 file changed, 19 insertions(+), 2 deletions(-) diff --git a/projects/hip/bin/hipcc b/projects/hip/bin/hipcc index 03f35b27fc..ca504fae4e 100755 --- a/projects/hip/bin/hipcc +++ b/projects/hip/bin/hipcc @@ -377,14 +377,31 @@ foreach $arg (@ARGV) while (my $line = <$in>) { chomp $line; if ($line =~ m/\.a$/) { + my $libFile = $line; my $path = abs_path($line); my @objs = split ('\n', `cd $tmpdir; ar xv $path`); + ## Check if all files in .a are object files. + my $allIsObj = 1; + my $realObjs = ""; foreach my $obj (@objs) { chomp $obj; $obj =~ s/^x - //; $obj = "$tmpdir/$obj"; - push (@inputs, $obj); - $new_arg = "$new_arg $obj"; + my $fileType = `file $obj`; + my $isObj = ($fileType =~ m/ELF/ or $fileType =~ m/COFF/); + $allIsObj = ($allIsObj and $isObj); + if ($isObj) { + $realObjs = $realObjs . " " . $obj; + } else { + push (@inputs, $obj); + $new_arg = "$new_arg $obj"; + } + } + if ($allIsObj) { + print $out "$line\n"; + } else { + system("cd $tmpdir; ar c $libFile $realObjs"); + print $out "$tmpdir/$libFile\n"; } } else { print $out "$line\n"; From acaf783f5269bb435150abfd8bfbf6e448aa1614 Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Sun, 17 Jun 2018 12:18:37 -0400 Subject: [PATCH 02/22] Fix handling of static library in hipcc for hip-clang [ROCm/hip commit: 2b32dbd414226adcac8c8b31211031128fffd6a8] --- projects/hip/bin/hipcc | 11 +++++++---- 1 file changed, 7 insertions(+), 4 deletions(-) diff --git a/projects/hip/bin/hipcc b/projects/hip/bin/hipcc index ca504fae4e..4c526bd2c7 100755 --- a/projects/hip/bin/hipcc +++ b/projects/hip/bin/hipcc @@ -391,17 +391,20 @@ foreach $arg (@ARGV) my $isObj = ($fileType =~ m/ELF/ or $fileType =~ m/COFF/); $allIsObj = ($allIsObj and $isObj); if ($isObj) { - $realObjs = $realObjs . " " . $obj; + $realObjs = ($realObjs . " " . $obj); } else { push (@inputs, $obj); $new_arg = "$new_arg $obj"; } } + chomp $realObjs; if ($allIsObj) { print $out "$line\n"; - } else { - system("cd $tmpdir; ar c $libFile $realObjs"); - print $out "$tmpdir/$libFile\n"; + } elsif ($realObjs) { + my $libBaseName = basename($libFile, ".a"); + $libBaseName = mktemp($libBaseName . "XXXX") . ".a"; + system("cd $tmpdir; ar c $libBaseName $realObjs"); + print $out "$tmpdir/$libBaseName\n"; } } else { print $out "$line\n"; From 8696038043a5d32a143dd884d0b8ed38f54add13 Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Mon, 18 Jun 2018 21:43:24 -0400 Subject: [PATCH 03/22] Let hipcc handle library with extension lo for hip-clang [ROCm/hip commit: 46d3c1d51e26f6e464e32872504d4219c3c5dec8] --- projects/hip/bin/hipcc | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/projects/hip/bin/hipcc b/projects/hip/bin/hipcc index 4c526bd2c7..f5fe4249ca 100755 --- a/projects/hip/bin/hipcc +++ b/projects/hip/bin/hipcc @@ -376,7 +376,7 @@ foreach $arg (@ARGV) open my $out, ">", $new_file or die "$new_file: $!"; while (my $line = <$in>) { chomp $line; - if ($line =~ m/\.a$/) { + if ($line =~ m/\.a$/ || $line =~ m/\.lo$/) { my $libFile = $line; my $path = abs_path($line); my @objs = split ('\n', `cd $tmpdir; ar xv $path`); @@ -401,8 +401,8 @@ foreach $arg (@ARGV) if ($allIsObj) { print $out "$line\n"; } elsif ($realObjs) { - my $libBaseName = basename($libFile, ".a"); - $libBaseName = mktemp($libBaseName . "XXXX") . ".a"; + my($libBaseName, $libDir, $libExt) = fileparse($libFile); + $libBaseName = mktemp($libBaseName . "XXXX") . $libExt; system("cd $tmpdir; ar c $libBaseName $realObjs"); print $out "$tmpdir/$libBaseName\n"; } From 260e50769bc1d68260c74d51c513b5681fc4d9ae Mon Sep 17 00:00:00 2001 From: Aaron Enye Shi Date: Fri, 22 Jun 2018 14:37:19 +0000 Subject: [PATCH 04/22] Add HIP Compute Mode [ROCm/hip commit: ff924ecb3d2a0135b03f653e266f63b01e9f22a4] --- projects/hip/include/hip/hip_runtime_api.h | 6 ++++++ 1 file changed, 6 insertions(+) diff --git a/projects/hip/include/hip/hip_runtime_api.h b/projects/hip/include/hip/hip_runtime_api.h index e18c17e07b..2ff562cc01 100644 --- a/projects/hip/include/hip/hip_runtime_api.h +++ b/projects/hip/include/hip/hip_runtime_api.h @@ -293,6 +293,12 @@ typedef enum hipDeviceAttribute_t { hipDeviceAttributeIntegrated, ///< iGPU } hipDeviceAttribute_t; +enum hipComputeMode { + hipComputeModeDefault = 0, + hipComputeModeExclusive = 1, + hipComputeModeProhibited = 2, + hipComputeModeExcusiveProcess = 3 +}; /** * @} From af0e9b54de8373fc0c0a499bf3c300850f48b4ed Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Mon, 25 Jun 2018 18:12:36 -0400 Subject: [PATCH 05/22] Add workaround to hipcc for build failure in tensorflow due to missing symbol __cpu_model https://github.com/tensorflow/tensorflow/issues/9593 [ROCm/hip commit: 021728cd863f212fa5ebd18cba31af1af9a3f789] --- projects/hip/bin/hipcc | 3 +++ 1 file changed, 3 insertions(+) diff --git a/projects/hip/bin/hipcc b/projects/hip/bin/hipcc index 03f35b27fc..731c9bcc96 100755 --- a/projects/hip/bin/hipcc +++ b/projects/hip/bin/hipcc @@ -601,6 +601,9 @@ if ($needLDFLAGS and not $compileOnly) { $CMD .= " $HIPLDFLAGS"; } $CMD .= " $toolArgs"; +if ($needLDFLAGS and not $compileOnly and $HIP_PLATFORM eq "clang") { + $CMD .= " -lgcc_s -lgcc"; +} if ($verbose & 0x1) { print "hipcc-cmd: ", $CMD, "\n"; From 1cccfbc6133bda9130d17fe10b3f6f6e7e903b8c Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Sat, 30 Jun 2018 11:40:32 +0530 Subject: [PATCH 06/22] Updated indentation [ROCm/hip commit: 0593b84d7cfcedd5beb15723df3a144ac134f348] --- .../docs/markdown/hip_deprecated_api_list.md | 32 +++++++++---------- 1 file changed, 16 insertions(+), 16 deletions(-) diff --git a/projects/hip/docs/markdown/hip_deprecated_api_list.md b/projects/hip/docs/markdown/hip_deprecated_api_list.md index a96f7f4d3f..6a9ed48839 100644 --- a/projects/hip/docs/markdown/hip_deprecated_api_list.md +++ b/projects/hip/docs/markdown/hip_deprecated_api_list.md @@ -4,19 +4,19 @@ CUDA supports cuCtx API, the Driver API that defines "Context" and "Devices" as separate entities. Contexts contain a single device, and a device can theoretically have multiple contexts. HIP initially added limited support for these API to facilitate easy porting from existing driver codes. These API are marked as deprecated now since there are better alternate interface (such as hipSetDevice or the stream API) to achieve the required functions. -###hipCtxCreate -###hipCtxDestroy -###hipCtxPopCurrent -###hipCtxPushCurrent -###hipCtxSetCurrent -###hipCtxGetCurrent -###hipCtxGetDevice -###hipCtxGetApiVersion -###hipCtxGetCacheConfig -###hipCtxSetCacheConfig -###hipCtxSetSharedMemConfig -###hipCtxGetSharedMemConfig -###hipCtxSynchronize -###hipCtxGetFlags -###hipCtxEnablePeerAccess -###hipCtxDisablePeerAccess +### hipCtxCreate +### hipCtxDestroy +### hipCtxPopCurrent +### hipCtxPushCurrent +### hipCtxSetCurrent +### hipCtxGetCurrent +### hipCtxGetDevice +### hipCtxGetApiVersion +### hipCtxGetCacheConfig +### hipCtxSetCacheConfig +### hipCtxSetSharedMemConfig +### hipCtxGetSharedMemConfig +### hipCtxSynchronize +### hipCtxGetFlags +### hipCtxEnablePeerAccess +### hipCtxDisablePeerAccess From e504d5472b33fdd0bc1b05616ebef76eeb3bb005 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Sat, 30 Jun 2018 11:42:17 +0530 Subject: [PATCH 07/22] Updated heading [ROCm/hip commit: 5a061f78aa71d5bce52c71abcdaf52a7e51a37ce] --- projects/hip/docs/markdown/hip_deprecated_api_list.md | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/projects/hip/docs/markdown/hip_deprecated_api_list.md b/projects/hip/docs/markdown/hip_deprecated_api_list.md index 6a9ed48839..dfb202c8ee 100644 --- a/projects/hip/docs/markdown/hip_deprecated_api_list.md +++ b/projects/hip/docs/markdown/hip_deprecated_api_list.md @@ -1,4 +1,4 @@ -# HIP Deprecated API List +# HIP Deprecated APIs ## HIP Context API From b7367558ef58563868c58ed51504e8cebb492136 Mon Sep 17 00:00:00 2001 From: Maneesh Gupta Date: Mon, 2 Jul 2018 10:37:20 +0530 Subject: [PATCH 08/22] Update hip-targets.cmake for hip::host & hip::device [ROCm/hip commit: 4b660c8382a99cddfc21b52e084cbe82919305da] --- projects/hip/packaging/hip-targets.cmake | 20 +++++++++++++++++++- 1 file changed, 19 insertions(+), 1 deletion(-) diff --git a/projects/hip/packaging/hip-targets.cmake b/projects/hip/packaging/hip-targets.cmake index 65370eec9e..5aff2ca1ed 100644 --- a/projects/hip/packaging/hip-targets.cmake +++ b/projects/hip/packaging/hip-targets.cmake @@ -16,7 +16,7 @@ set(CMAKE_IMPORT_FILE_VERSION 1) set(_targetsDefined) set(_targetsNotDefined) set(_expectedTargets) -foreach(_expectedTarget hip::hip_hcc_static hip::hip_hcc hip::hip_device) +foreach(_expectedTarget hip::hip_hcc_static hip::hip_hcc hip::hip_device hip::host hip::device) list(APPEND _expectedTargets ${_expectedTarget}) if(NOT TARGET ${_expectedTarget}) list(APPEND _targetsNotDefined ${_expectedTarget}) @@ -65,6 +65,24 @@ set_target_properties(hip::hip_device PROPERTIES INTERFACE_SYSTEM_INCLUDE_DIRECTORIES "${_IMPORT_PREFIX}/include;/opt/rocm/hsa/include" ) +# Create imported target hip::host +add_library(hip::host INTERFACE IMPORTED) + +set_target_properties(hip::host PROPERTIES + INTERFACE_LINK_LIBRARIES "hip::hip_hcc" +) + +# Create imported target hip::device +add_library(hip::device INTERFACE IMPORTED) + +set_target_properties(hip::device PROPERTIES + INTERFACE_LINK_LIBRARIES "hip::host;hip::hip_device;hcc::hccrt;hcc::hc_am" +) + +if(CMAKE_VERSION VERSION_LESS 3.0.0) + message(FATAL_ERROR "This file relies on consumers using CMake 3.0.0 or greater.") +endif() + # Load information for each installed configuration. get_filename_component(_DIR "${CMAKE_CURRENT_LIST_FILE}" PATH) file(GLOB CONFIG_FILES "${_DIR}/hip-targets-*.cmake") From c957c42c20db4b4af5a25e0fec69c04f88fcbb15 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Mon, 2 Jul 2018 14:32:11 +0530 Subject: [PATCH 09/22] Revert "Use memcpy kernel for all pinned memory cases in hipMemcpy2DAsync" [ROCm/hip commit: 7cd1d5e644b61a0fb820985382dee84d5ef5a56e] --- projects/hip/src/hip_memory.cpp | 5 +---- 1 file changed, 1 insertion(+), 4 deletions(-) diff --git a/projects/hip/src/hip_memory.cpp b/projects/hip/src/hip_memory.cpp index 787e49683b..d6c04ae98c 100644 --- a/projects/hip/src/hip_memory.cpp +++ b/projects/hip/src/hip_memory.cpp @@ -1680,12 +1680,9 @@ hipError_t hipMemcpy2DAsync(void* dst, size_t dpitch, const void* src, size_t sp actualDest = pinnedPtr; } } -#if 0 if((width == dpitch) && (width == spitch)) { hip_internal::memcpyAsync(dst, src, width*height, kind, stream); - } else -#endif - { + } else { try { if(!isLocked){ for (int i = 0; i < height; ++i) From ff253a279c96a33e336f8739f3a370cbc6a2ada6 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Tue, 3 Jul 2018 08:54:17 +0530 Subject: [PATCH 10/22] Fixed offset null check in bind texture functions [ROCm/hip commit: feff0aeea4622a3a03bd2aeb739604928db1ec70] --- projects/hip/src/hip_texture.cpp | 6 ++++-- 1 file changed, 4 insertions(+), 2 deletions(-) diff --git a/projects/hip/src/hip_texture.cpp b/projects/hip/src/hip_texture.cpp index 24c6eef3af..d6caf853de 100644 --- a/projects/hip/src/hip_texture.cpp +++ b/projects/hip/src/hip_texture.cpp @@ -389,7 +389,8 @@ hipError_t ihipBindTextureImpl(int dim, enum hipTextureReadMode readMode, size_t enum hipTextureFilterMode filterMode = tex->filterMode; int normalizedCoords = tex->normalized; hipTextureObject_t& textureObject = tex->textureObject; - *offset = 0; + if(offset != nullptr) + *offset = 0; auto ctx = ihipGetTlsDefaultCtx(); if (ctx) { hc::accelerator acc = ctx->getDevice()->_acc; @@ -459,7 +460,8 @@ hipError_t ihipBindTexture2DImpl(int dim, enum hipTextureReadMode readMode, size enum hipTextureFilterMode filterMode = tex->filterMode; int normalizedCoords = tex->normalized; hipTextureObject_t& textureObject = tex->textureObject; - *offset = 0; + if(offset != nullptr) + *offset = 0; auto ctx = ihipGetTlsDefaultCtx(); if (ctx) { hc::accelerator acc = ctx->getDevice()->_acc; From 19cb14d54cd10975c8e9ec4a6b678bf67dc1fd97 Mon Sep 17 00:00:00 2001 From: Maneesh Gupta Date: Wed, 4 Jul 2018 09:33:51 +0530 Subject: [PATCH 11/22] Update hip_hcc_internal.h Adding missing include for hip_hcc_internal in order to build with HCC [ROCm/hip commit: 0c2f985553d6223d1219a3868ecb8d474cd6de20] --- projects/hip/src/hip_hcc_internal.h | 1 + 1 file changed, 1 insertion(+) diff --git a/projects/hip/src/hip_hcc_internal.h b/projects/hip/src/hip_hcc_internal.h index 4008df1574..be257aff4f 100644 --- a/projects/hip/src/hip_hcc_internal.h +++ b/projects/hip/src/hip_hcc_internal.h @@ -26,6 +26,7 @@ THE SOFTWARE. #include #include #include +#include #include "hsa/hsa_ext_amd.h" #include "hip/hip_runtime.h" From 0d40be6866678847b2f8d059d445cd75342e28c2 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Thu, 5 Jul 2018 23:00:41 +0530 Subject: [PATCH 12/22] Added tex2dlayered mapping for HIP/NVCC [ROCm/hip commit: 19bae58e487062299151628810cf66ec02c1d7ec] --- projects/hip/include/hip/nvcc_detail/hip_runtime_api.h | 1 + 1 file changed, 1 insertion(+) 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 3b8a3661f7..afe7f11f52 100644 --- a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h @@ -163,6 +163,7 @@ typedef cudaSurfaceObject_t hipSurfaceObject_t; #define hipTextureType1D cudaTextureType1D #define hipTextureType1DLayered cudaTextureType1DLayered #define hipTextureType2D cudaTextureType2D +#define hipTextureType2DLayered cudaTextureType2DLayered #define hipTextureType3D cudaTextureType3D #define hipDeviceMapHost cudaDeviceMapHost From 399ec56015316895f4c4a1caa648ecd86d04fc57 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Thu, 5 Jul 2018 23:11:39 +0530 Subject: [PATCH 13/22] Corrected enum type [ROCm/hip commit: 4acf489bd57fc0f739f83f0869994c7b16508bcf] --- projects/hip/include/hip/nvcc_detail/hip_runtime_api.h | 8 ++++---- 1 file changed, 4 insertions(+), 4 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 3b8a3661f7..557b15e9a2 100644 --- a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h @@ -1168,20 +1168,20 @@ inline static hipError_t hipOccupancyMaxPotentialBlockSize(int* minGridSize, int return hipCUDAErrorTohipError(cerror); } -template +template inline static hipError_t hipBindTexture(size_t* offset, const struct texture& tex, const void* devPtr, size_t size = UINT_MAX) { return hipCUDAErrorTohipError(cudaBindTexture(offset, tex, devPtr, size)); } -template +template inline static hipError_t hipBindTexture(size_t* offset, struct texture& tex, const void* devPtr, const struct hipChannelFormatDesc& desc, size_t size = UINT_MAX) { return hipCUDAErrorTohipError(cudaBindTexture(offset, tex, devPtr, desc, size)); } -template +template inline static hipError_t hipUnbindTexture(struct texture* tex) { return hipCUDAErrorTohipError(cudaUnbindTexture(tex)); } @@ -1198,7 +1198,7 @@ inline static hipError_t hipBindTextureToArray(struct texture& return hipCUDAErrorTohipError(cudaBindTextureToArray(tex, array, desc)); } -template +template inline static hipError_t hipBindTextureToArray(struct texture& tex, hipArray_const_t array) { return hipCUDAErrorTohipError(cudaBindTextureToArray(tex, array)); From 9735598cd1838ce970412221cfab5a1f2cb3dc45 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Thu, 5 Jul 2018 23:41:31 +0530 Subject: [PATCH 14/22] Added another variant of bindtextoarray for direct porting [ROCm/hip commit: bd4816fc85590d6094aa6e708125aa1bc56b4c7d] --- projects/hip/include/hip/hcc_detail/hip_runtime_api.h | 7 +++++++ projects/hip/include/hip/nvcc_detail/hip_runtime_api.h | 7 +++++++ 2 files changed, 14 insertions(+) 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 0a80a583c7..573ae39af9 100644 --- a/projects/hip/include/hip/hcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/hcc_detail/hip_runtime_api.h @@ -2621,6 +2621,13 @@ hipError_t hipBindTextureToArray(struct texture& tex, hipArray return ihipBindTextureToArrayImpl(dim, readMode, array, desc, &tex); } +template +inline static hipError_t hipBindTextureToArray(struct texture *tex, + hipArray_const_t array, + const struct hipChannelFormatDesc* desc) { + return ihipBindTextureToArrayImpl(dim, readMode, array, *desc, tex); +} + // C API hipError_t hipBindTextureToMipmappedArray(const textureReference* tex, hipMipmappedArray_const_t mipmappedArray, 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 3b8a3661f7..64404574f0 100644 --- a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h @@ -1198,6 +1198,13 @@ inline static hipError_t hipBindTextureToArray(struct texture& return hipCUDAErrorTohipError(cudaBindTextureToArray(tex, array, desc)); } +template +inline static hipError_t hipBindTextureToArray(struct texture *tex, + hipArray_const_t array, + const struct hipChannelFormatDesc* desc) { + return hipCUDAErrorTohipError(cudaBindTextureToArray(tex, array, desc)); +} + template inline static hipError_t hipBindTextureToArray(struct texture& tex, hipArray_const_t array) { From 9998b882bbb7b5ee0636e65474b78e72f25a0ff7 Mon Sep 17 00:00:00 2001 From: Aaron Enye Shi Date: Fri, 6 Jul 2018 11:26:48 -0400 Subject: [PATCH 15/22] Fix hip_mbcnt test typo Ctest did not catch this test failure. When running this test manually before typo, I got __mbcnt_hi() FAILED! . This fix will fix this test for HCC and HIP clang path. [ROCm/hip commit: aa05779c64b42ca11c1401a89ff175ad92124633] --- projects/hip/tests/src/deviceLib/hip_mbcnt.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/projects/hip/tests/src/deviceLib/hip_mbcnt.cpp b/projects/hip/tests/src/deviceLib/hip_mbcnt.cpp index 0dd7169f51..9fdf36a1d3 100644 --- a/projects/hip/tests/src/deviceLib/hip_mbcnt.cpp +++ b/projects/hip/tests/src/deviceLib/hip_mbcnt.cpp @@ -88,7 +88,7 @@ int main() { for (unsigned int i = 0; i < num_threads; i++) { unsigned int this_lane_id = i % wave_size; unsigned int this_mbcnt_lo = this_lane_id >= 32 ? 32 : this_lane_id; - unsigned int this_mbcnt_hi = this_lane_id < 32 ? 0 : (this_lane_id - 22); + unsigned int this_mbcnt_hi = this_lane_id < 32 ? 0 : (this_lane_id - 32); if (host_mbcnt_lo[i] != this_mbcnt_lo) mbcnt_lo_errors++; From 5a6fe18e0d540bd53aba70270e0221dea5cd7f0d Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Tue, 10 Jul 2018 20:37:54 +0300 Subject: [PATCH 16/22] [HIPIFY][docs] Update README.md after testing with new LLVM releases 5.0.2 and 6.0.1 [ROCm/hip commit: 75164a5bf50b292ccf0eff395aa8bc7ef4a13962] --- projects/hip/hipify-clang/README.md | 23 ++++++++++++++++------- 1 file changed, 16 insertions(+), 7 deletions(-) diff --git a/projects/hip/hipify-clang/README.md b/projects/hip/hipify-clang/README.md index c6f59a0251..e099085ca6 100644 --- a/projects/hip/hipify-clang/README.md +++ b/projects/hip/hipify-clang/README.md @@ -31,7 +31,9 @@ | 4.0.1 | 8.0 | | 5.0.0 | 8.0 | | 5.0.1 | 8.0 | +| 5.0.2 | 8.0 | | 6.0.0 | 9.0 | +| 6.0.1 | 9.0 | In most cases, you can get a suitable version of LLVM+CLANG with your package manager. @@ -80,7 +82,7 @@ To run it: cmake \ -DCMAKE_INSTALL_PREFIX=../dist \ -DLLVM_SOURCE_DIR=../llvm \ - -DCMAKE_BUILD_TYPE=Debug \ + -DCMAKE_BUILD_TYPE=Release \ -Thost=x64 \ ../llvm @@ -114,9 +116,9 @@ To run it: * Starting with LLVM 6.0.0 path to llvm-lit.py script should be specified by the `LLVM_EXTERNAL_LIT` option: - `-DLLVM_EXTERNAL_LIT=f:/LLVM/6.0.0/build/Debug/bin/llvm-lit.py`, + `-DLLVM_EXTERNAL_LIT=f:/LLVM/6.0.0/build/Release/bin/llvm-lit.py`, - where `f:/LLVM/6.0.0/build/Debug` is LLVM build directory. + where `f:/LLVM/6.0.0/build/Release` is LLVM build directory. 7. Build with the `HIPIFY_CLANG_TESTS` option turned on: -DHIPIFY_CLANG_TESTS=1. 8. `make test-hipify` @@ -124,9 +126,15 @@ To run it: ### Windows -On Windows the following tested configuration is recommended: +On Windows the following configurations are tested: -LLVM 6.0.0 (exact), CUDA 9.0 (exact), cudnn-9.0 (exact), Python 3.6 (min), cmake 3.10 (min), Visual Studio 15.5 2017 (min). +LLVM 6.0.0 - 6.0.1, CUDA 9.0, cudnn-9.0 + +LLVM 5.0.0 - 5.0.2, CUDA 8.0, cudnn-8.0 + +Build system for the above configurations: + +Python 3.6 (min), cmake 3.10 (min), Visual Studio 15.5 2017 (min). Here is an example of building `hipify-clang` with testing support on `Windows 10` by `Visual Studio 15 2017`: @@ -134,13 +142,14 @@ Here is an example of building `hipify-clang` with testing support on `Windows 1 cmake -G "Visual Studio 15 2017 Win64" \ -DHIPIFY_CLANG_TESTS=1 \ - -DCMAKE_BUILD_TYPE=Debug \ + -DCMAKE_BUILD_TYPE=Release \ -DCMAKE_INSTALL_PREFIX=../dist \ -DCMAKE_PREFIX_PATH=f:/LLVM/6.0.0/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/Debug/bin/llvm-lit.py \ + -DLLVM_EXTERNAL_LIT=f:/LLVM/6.0.0/build/Release/bin/llvm-lit.py \ + -Thost=x64 .. ``` A corresponding successful output: From 0b0041e82c790efbc4ac934430965cbbd42ad246 Mon Sep 17 00:00:00 2001 From: "Yaxun (Sam) Liu" Date: Tue, 10 Jul 2018 16:49:59 -0400 Subject: [PATCH 17/22] Fix build failure in code_object_bundle.cpp [ROCm/hip commit: 7db7cce9e4a5e9bbe46fd0b903322130a36a4d7d] --- projects/hip/src/code_object_bundle.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/projects/hip/src/code_object_bundle.cpp b/projects/hip/src/code_object_bundle.cpp index ede7090a52..91258f0c75 100644 --- a/projects/hip/src/code_object_bundle.cpp +++ b/projects/hip/src/code_object_bundle.cpp @@ -38,7 +38,7 @@ std::string isa_name(std::string triple) hsa_isa_from_name(triple.c_str(), &tmp) != HSA_STATUS_SUCCESS}; if (is_old_rocr) { - auto tmp{triple.substr(triple.rfind('x') + 1)}; + std::string tmp{triple.substr(triple.rfind('x') + 1)}; triple.replace(0, std::string::npos, "AMD:AMDGPU"); for (auto&& x : tmp) { @@ -51,7 +51,7 @@ std::string isa_name(std::string triple) } hsa_isa_t hip_impl::triple_to_hsa_isa(const std::string& triple) { - const auto isa{isa_name(std::move(triple))}; + const std::string isa{isa_name(std::move(triple))}; if (isa.empty()) return hsa_isa_t({}); From 54881b7a12a1ca9e277aee67665fb84e33ac3b2e Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Wed, 11 Jul 2018 12:17:33 +0530 Subject: [PATCH 18/22] Add hipGetTextureAlignmentOffset on NVCC path [ROCm/hip commit: c93f216fe8b0af6295628b28f021fbf1fccd9fe1] --- projects/hip/include/hip/nvcc_detail/hip_runtime_api.h | 5 +++++ 1 file changed, 5 insertions(+) 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 3b8a3661f7..efa0c3b7ea 100644 --- a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h @@ -1239,6 +1239,11 @@ inline static hipError_t hipGetTextureObjectResourceDesc(hipResourceDesc* pResDe hipTextureObject_t textureObject) { return hipCUDAErrorTohipError(cudaGetTextureObjectResourceDesc( pResDesc, textureObject)); } + +inline static hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference* texref) +{ + return hipCUDAErrorTohipError(cudaGetTextureAlignmentOffset(offset,texref)); +} #endif //__CUDACC__ #endif // HIP_INCLUDE_HIP_NVCC_DETAIL_HIP_RUNTIME_API_H From 00f7f9488b0002431169b4a2abb73ea87a08186b Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Wed, 11 Jul 2018 12:37:07 +0530 Subject: [PATCH 19/22] Added hipGetChanDesc for NVCC path [ROCm/hip commit: 6bf979d9e6f3ec513e1bac6204714102a4195898] --- projects/hip/include/hip/nvcc_detail/hip_runtime_api.h | 5 +++++ 1 file changed, 5 insertions(+) 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 efa0c3b7ea..5706e1d097 100644 --- a/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h +++ b/projects/hip/include/hip/nvcc_detail/hip_runtime_api.h @@ -1244,6 +1244,11 @@ inline static hipError_t hipGetTextureAlignmentOffset(size_t* offset, const text { return hipCUDAErrorTohipError(cudaGetTextureAlignmentOffset(offset,texref)); } + +inline static hipError_t hipGetChannelDesc(hipChannelFormatDesc* desc, hipArray_const_t array) +{ + return hipCUDAErrorTohipError(cudaGetChannelDesc(desc,array)); +} #endif //__CUDACC__ #endif // HIP_INCLUDE_HIP_NVCC_DETAIL_HIP_RUNTIME_API_H From 2c8fbf1a6760eea5e8c7d1d6a4d115ca8e9677ab Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Wed, 11 Jul 2018 18:21:00 +0300 Subject: [PATCH 20/22] [HIPIFY] Add DEBUG(X) macro compatibility In LLVM 7.0 DEBUG(X) was deleted, LLVM_DEBUG(X) should be used instead. [ROCm/hip commit: 9fcb48e5c4a80641b4531c5f7895f29da4d169f6] --- projects/hip/hipify-clang/src/LLVMCompat.h | 4 ++++ projects/hip/hipify-clang/src/main.cpp | 2 +- 2 files changed, 5 insertions(+), 1 deletion(-) diff --git a/projects/hip/hipify-clang/src/LLVMCompat.h b/projects/hip/hipify-clang/src/LLVMCompat.h index 3e2fe1aebb..72b6832012 100644 --- a/projects/hip/hipify-clang/src/LLVMCompat.h +++ b/projects/hip/hipify-clang/src/LLVMCompat.h @@ -23,6 +23,10 @@ namespace llcompat { #define GET_NUM_ARGS() getNumArgs() #endif +#if LLVM_VERSION_MAJOR < 7 + #define LLVM_DEBUG(X) DEBUG(X) +#endif + void PrintStackTraceOnErrorSignal(); /** diff --git a/projects/hip/hipify-clang/src/main.cpp b/projects/hip/hipify-clang/src/main.cpp index ccf627b147..e420ab0681 100644 --- a/projects/hip/hipify-clang/src/main.cpp +++ b/projects/hip/hipify-clang/src/main.cpp @@ -132,7 +132,7 @@ int main(int argc, const char **argv) { // Hipify _all_ the things! if (Tool.runAndSave(&actionFactory)) { - DEBUG(llvm::dbgs() << "Skipped some replacements.\n"); + LLVM_DEBUG(llvm::dbgs() << "Skipped some replacements.\n"); } // Either move the tmpfile to the output, or remove it. From 51fff4049a2a5bab4d586969a7059e539819b474 Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Wed, 11 Jul 2018 20:15:49 +0300 Subject: [PATCH 21/22] [HIPIFY] Current trunk LLVM 7.0 initial support Tested with CUDA 8.0, 9.0, 9.1 and 9.2. Only 8.0 works with LLVM 7.0, due to the changes in LLVM trunc since released 6.0, which works fine with CUDA 8.0 and 9.0. So, nothing to do in hipify-clang, hope that all the CUDA 9.x related issues will be fixed in 7.0 release. [ROCm/hip commit: 433e1727765e5f0769708c652ed25cbd0cb145e3] --- projects/hip/hipify-clang/CMakeLists.txt | 6 +++++- projects/hip/hipify-clang/src/HipifyAction.cpp | 6 +++++- 2 files changed, 10 insertions(+), 2 deletions(-) diff --git a/projects/hip/hipify-clang/CMakeLists.txt b/projects/hip/hipify-clang/CMakeLists.txt index 8b3fa7e591..5d9070be28 100644 --- a/projects/hip/hipify-clang/CMakeLists.txt +++ b/projects/hip/hipify-clang/CMakeLists.txt @@ -51,6 +51,10 @@ if(WIN32) target_link_libraries(hipify-clang version) endif() +if ((LLVM_PACKAGE_VERSION VERSION_EQUAL "7") OR (LLVM_PACKAGE_VERSION VERSION_GREATER "7")) + target_link_libraries(hipify-clang clangToolingInclusions) +endif() + set(CMAKE_C_FLAGS "${CMAKE_C_FLAGS} ${EXTRA_CFLAGS}") set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} ${EXTRA_CFLAGS}") if(MSVC) @@ -95,7 +99,7 @@ if (HIPIFY_CLANG_TESTS) message(STATUS "Please install clang 4.0 or higher.") elseif (CUDA_VERSION VERSION_EQUAL "9.0") message(STATUS "Please install clang 6.0 or higher.") - elseif (CUDA_VERSION VERSION_EQUAL "9.1") + elseif ((CUDA_VERSION VERSION_EQUAL "9.1") OR (CUDA_VERSION VERSION_EQUAL "9.2")) message(STATUS "Please install clang 7.0 or higher.") endif() endif() diff --git a/projects/hip/hipify-clang/src/HipifyAction.cpp b/projects/hip/hipify-clang/src/HipifyAction.cpp index 7e5ff4357d..21b16e0699 100644 --- a/projects/hip/hipify-clang/src/HipifyAction.cpp +++ b/projects/hip/hipify-clang/src/HipifyAction.cpp @@ -428,7 +428,11 @@ public: void InclusionDirective(clang::SourceLocation hash_loc, const clang::Token& include_token, StringRef file_name, bool is_angled, clang::CharSourceRange filename_range, const clang::FileEntry* file, StringRef search_path, StringRef relative_path, - const clang::Module* imported) override { + const clang::Module* imported +#if LLVM_VERSION_MAJOR > 6 + , clang::SrcMgr::CharacteristicKind FileType +#endif + ) override { hipifyAction.InclusionDirective(hash_loc, include_token, file_name, is_angled, filename_range, file, search_path, relative_path, imported); } From b063b3eb2662982540bf1c0c7da65df354e1016f Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Sat, 14 Jul 2018 16:08:19 +0300 Subject: [PATCH 22/22] [HIPIFY] Support of cudaComputeMode / CUcomputemode + update docs + fix typo in hip_runtime_api.h [ROCm/hip commit: 55e21055e3526f4f97571479cb49db67091fdf78] --- ...A_Driver_API_functions_supported_by_HIP.md | 7 ++++++- ..._Runtime_API_functions_supported_by_HIP.md | 10 +++++----- projects/hip/hipify-clang/src/CUDA2HipMap.cpp | 20 +++++++++---------- projects/hip/include/hip/hip_runtime_api.h | 2 +- 4 files changed, 22 insertions(+), 17 deletions(-) diff --git a/projects/hip/docs/markdown/CUDA_Driver_API_functions_supported_by_HIP.md b/projects/hip/docs/markdown/CUDA_Driver_API_functions_supported_by_HIP.md index c1c30ae019..7e806886c6 100644 --- a/projects/hip/docs/markdown/CUDA_Driver_API_functions_supported_by_HIP.md +++ b/projects/hip/docs/markdown/CUDA_Driver_API_functions_supported_by_HIP.md @@ -231,6 +231,11 @@ | 0x02 |*`CU_MEMORYTYPE_DEVICE`* | | | 0x03 |*`CU_MEMORYTYPE_ARRAY`* | | | 0x04 |*`CU_MEMORYTYPE_UNIFIED`* | | +| enum |***`CUcomputemode`*** |***`hipComputeMode`*** | +| 0 |*`CU_COMPUTEMODE_DEFAULT`* |*`hipComputeModeDefault`* | +| 1 |*`CU_COMPUTEMODE_EXCLUSIVE`* |*`hipComputeModeExclusive`* | +| 2 |*`CU_COMPUTEMODE_PROHIBITED`* |*`hipComputeModeProhibited`* | +| 3 |*`CU_COMPUTEMODE_EXCLUSIVE_PROCESS`* |*`hipComputeModeExclusiveProcess`* | | enum |***`CUoccupancy_flags`*** | | | 0x00 |*`CU_OCCUPANCY_DEFAULT`* | | | 0x01 |*`CU_OCCUPANCY_DISABLE_CACHING_OVERRIDE`* | | @@ -243,7 +248,7 @@ | 6 |*`CU_POINTER_ATTRIBUTE_SYNC_MEMOPS`* | | | 7 |*`CU_POINTER_ATTRIBUTE_BUFFER_ID`* | | | 8 |*`CU_POINTER_ATTRIBUTE_IS_MANAGED`* | | -| enum |***`CUmemorytype`*** | | +| enum |***`CUresourcetype`*** | | | 0x00 |*`CU_RESOURCE_TYPE_ARRAY`* | | | 0x01 |*`CU_RESOURCE_TYPE_MIPMAPPED_ARRAY`* | | | 0x02 |*`CU_RESOURCE_TYPE_LINEAR`* | | diff --git a/projects/hip/docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md b/projects/hip/docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md index 65528da7fb..dca2683b12 100644 --- a/projects/hip/docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md +++ b/projects/hip/docs/markdown/CUDA_Runtime_API_functions_supported_by_HIP.md @@ -418,11 +418,11 @@ | 1 |*`cudaChannelFormatKindUnsigned`* |*`hipChannelFormatKindUnsigned`* | | 2 |*`cudaChannelFormatKindFloat`* |*`hipChannelFormatKindFloat`* | | 3 |*`cudaChannelFormatKindNone`* |*`hipChannelFormatKindNone`* | -| enum |***`cudaComputeMode`*** | | -| 0 |*`cudaComputeModeDefault`* | | -| 1 |*`cudaComputeModeExclusive`* | | -| 2 |*`cudaComputeModeProhibited`* | | -| 3 |*`cudaComputeModeExclusiveProcess`* | | +| enum |***`cudaComputeMode`*** |***`hipComputeMode`*** | +| 0 |*`cudaComputeModeDefault`* |*`hipComputeModeDefault`* | +| 1 |*`cudaComputeModeExclusive`* |*`hipComputeModeExclusive`* | +| 2 |*`cudaComputeModeProhibited`* |*`hipComputeModeProhibited`* | +| 3 |*`cudaComputeModeExclusiveProcess`* |*`hipComputeModeExclusiveProcess`* | | enum |***`cudaDeviceAttr`*** |***`hipDeviceAttribute_t`*** | | 1 |*`cudaDevAttrMaxThreadsPerBlock`* |*`hipDeviceAttributeMaxThreadsPerBlock`* | | 2 |*`cudaDevAttrMaxBlockDimX`* |*`hipDeviceAttributeMaxBlockDimX`* | diff --git a/projects/hip/hipify-clang/src/CUDA2HipMap.cpp b/projects/hip/hipify-clang/src/CUDA2HipMap.cpp index 7d10b35e48..588642ccb5 100644 --- a/projects/hip/hipify-clang/src/CUDA2HipMap.cpp +++ b/projects/hip/hipify-clang/src/CUDA2HipMap.cpp @@ -24,7 +24,7 @@ const std::map CUDA_TYPE_NAME_MAP{ {"CUaddress_mode", {"hipAddress_mode", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, {"CUarray_cubemap_face", {"hipArray_cubemap_face", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, {"CUarray_format", {"hipArray_format", CONV_TYPE, API_DRIVER}}, - {"CUcomputemode", {"hipComputemode", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // API_RUNTIME ANALOGUE (cudaComputeMode) + {"CUcomputemode", {"hipComputeMode", CONV_TYPE, API_DRIVER}}, // API_RUNTIME ANALOGUE (cudaComputeMode) {"CUmem_advise", {"hipMemAdvise", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // API_RUNTIME ANALOGUE (cudaComputeMode) {"CUmem_range_attribute", {"hipMemRangeAttribute", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // API_RUNTIME ANALOGUE (cudaMemRangeAttribute) {"CUctx_flags", {"hipCctx_flags", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, @@ -236,7 +236,7 @@ const std::map CUDA_TYPE_NAME_MAP{ {"cudaDeviceAttr", {"hipDeviceAttribute_t", CONV_TYPE, API_RUNTIME}}, // API_DRIVER ANALOGUE (CUdevice_attribute) {"cudaDeviceProp", {"hipDeviceProp_t", CONV_TYPE, API_RUNTIME}}, {"cudaDeviceP2PAttr", {"hipDeviceP2PAttribute", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, // API_DRIVER ANALOGUE (CUdevice_P2PAttribute) - {"cudaComputeMode", {"hipComputeMode", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, // API_DRIVER ANALOGUE (CUcomputemode) + {"cudaComputeMode", {"hipComputeMode", CONV_TYPE, API_RUNTIME}}, // API_DRIVER ANALOGUE (CUcomputemode) {"cudaFuncCache", {"hipFuncCache_t", CONV_CACHE, API_RUNTIME}}, // API_Driver ANALOGUE (CUfunc_cache) {"cudaFuncAttributes", {"hipFuncAttributes", CONV_EXEC, API_RUNTIME, HIP_UNSUPPORTED}}, {"cudaSharedMemConfig", {"hipSharedMemConfig", CONV_TYPE, API_RUNTIME}}, @@ -628,10 +628,10 @@ const std::map CUDA_IDENTIFIER_MAP{ {"CU_AD_FORMAT_FLOAT", {"HIP_AD_FORMAT_FLOAT", CONV_TYPE, API_DRIVER}}, // 0x20 // CUcomputemode enum - {"CU_COMPUTEMODE_DEFAULT", {"hipComputeModeDefault", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 0 // API_RUNTIME ANALOGUE (cudaComputeModeDefault = 0) - {"CU_COMPUTEMODE_EXCLUSIVE", {"hipComputeModeExclusive", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 1 // API_RUNTIME ANALOGUE (cudaComputeModeExclusive = 1) - {"CU_COMPUTEMODE_PROHIBITED", {"hipComputeModeProhibited", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 2 // API_RUNTIME ANALOGUE (cudaComputeModeProhibited = 2) - {"CU_COMPUTEMODE_EXCLUSIVE_PROCESS", {"hipComputeModeExclusiveProcess", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, // 3 // API_RUNTIME ANALOGUE (cudaComputeModeExclusiveProcess = 3) + {"CU_COMPUTEMODE_DEFAULT", {"hipComputeModeDefault", CONV_TYPE, API_DRIVER}}, // 0 // API_RUNTIME ANALOGUE (cudaComputeModeDefault = 0) + {"CU_COMPUTEMODE_EXCLUSIVE", {"hipComputeModeExclusive", CONV_TYPE, API_DRIVER}}, // 1 // API_RUNTIME ANALOGUE (cudaComputeModeExclusive = 1) + {"CU_COMPUTEMODE_PROHIBITED", {"hipComputeModeProhibited", CONV_TYPE, API_DRIVER}}, // 2 // API_RUNTIME ANALOGUE (cudaComputeModeProhibited = 2) + {"CU_COMPUTEMODE_EXCLUSIVE_PROCESS", {"hipComputeModeExclusiveProcess", CONV_TYPE, API_DRIVER}}, // 3 // API_RUNTIME ANALOGUE (cudaComputeModeExclusiveProcess = 3) // Memory advise values // {"CUmem_advise_enum", {"hipMemAdvise", CONV_TYPE, API_DRIVER, HIP_UNSUPPORTED}}, @@ -1698,10 +1698,10 @@ const std::map CUDA_IDENTIFIER_MAP{ {"cudaDeviceGetP2PAttribute", {"hipDeviceGetP2PAttribute", CONV_DEVICE, API_RUNTIME, HIP_UNSUPPORTED}}, // API_DRIVER ANALOGUE (cuDeviceGetP2PAttribute) // enum cudaComputeMode - {"cudaComputeModeDefault", {"hipComputeModeDefault", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, // 0 // API_DRIVER ANALOGUE (CU_COMPUTEMODE_DEFAULT = 0) - {"cudaComputeModeExclusive", {"hipComputeModeExclusive", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, // 1 // API_DRIVER ANALOGUE (CU_COMPUTEMODE_EXCLUSIVE = 1) - {"cudaComputeModeProhibited", {"hipComputeModeProhibited", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, // 2 // API_DRIVER ANALOGUE (CU_COMPUTEMODE_PROHIBITED = 2) - {"cudaComputeModeExclusiveProcess", {"hipComputeModeExclusiveProcess", CONV_TYPE, API_RUNTIME, HIP_UNSUPPORTED}}, // 3 // API_DRIVER ANALOGUE (CU_COMPUTEMODE_EXCLUSIVE_PROCESS = 3) + {"cudaComputeModeDefault", {"hipComputeModeDefault", CONV_TYPE, API_RUNTIME}}, // 0 // API_DRIVER ANALOGUE (CU_COMPUTEMODE_DEFAULT = 0) + {"cudaComputeModeExclusive", {"hipComputeModeExclusive", CONV_TYPE, API_RUNTIME}}, // 1 // API_DRIVER ANALOGUE (CU_COMPUTEMODE_EXCLUSIVE = 1) + {"cudaComputeModeProhibited", {"hipComputeModeProhibited", CONV_TYPE, API_RUNTIME}}, // 2 // API_DRIVER ANALOGUE (CU_COMPUTEMODE_PROHIBITED = 2) + {"cudaComputeModeExclusiveProcess", {"hipComputeModeExclusiveProcess", CONV_TYPE, API_RUNTIME}}, // 3 // API_DRIVER ANALOGUE (CU_COMPUTEMODE_EXCLUSIVE_PROCESS = 3) // Device Flags {"cudaGetDeviceFlags", {"hipGetDeviceFlags", CONV_DEVICE, API_RUNTIME, HIP_UNSUPPORTED}}, diff --git a/projects/hip/include/hip/hip_runtime_api.h b/projects/hip/include/hip/hip_runtime_api.h index 2ff562cc01..cd7af65265 100644 --- a/projects/hip/include/hip/hip_runtime_api.h +++ b/projects/hip/include/hip/hip_runtime_api.h @@ -297,7 +297,7 @@ enum hipComputeMode { hipComputeModeDefault = 0, hipComputeModeExclusive = 1, hipComputeModeProhibited = 2, - hipComputeModeExcusiveProcess = 3 + hipComputeModeExclusiveProcess = 3 }; /**