From 1c05b06bcda254265828cd1965100c6d48e95ae2 Mon Sep 17 00:00:00 2001 From: Aaron Enye Shi Date: Fri, 19 Oct 2018 20:07:08 +0000 Subject: [PATCH] Remove libhip_device.a static library Move remaining function definitions from device_util.cpp to hip_runtime.h header. We can now remove the static library completely as device_util.cpp was the last file part of target hip::hip_device . [ROCm/clr commit: ba43d1ce1b983a3ca8ebd34e7e962bd27e9d4f40] --- projects/clr/hipamd/CMakeLists.txt | 18 ++++------ projects/clr/hipamd/bin/hipcc | 4 +-- .../include/hip/hcc_detail/device_functions.h | 9 +++++ .../include/hip/hcc_detail/hip_runtime.h | 3 +- .../packaging/hip-targets-release.cmake | 10 ------ .../clr/hipamd/packaging/hip-targets.cmake | 16 +++------ projects/clr/hipamd/packaging/hip_hcc.txt | 1 - projects/clr/hipamd/src/device_util.cpp | 33 ------------------- 8 files changed, 23 insertions(+), 71 deletions(-) delete mode 100644 projects/clr/hipamd/src/device_util.cpp diff --git a/projects/clr/hipamd/CMakeLists.txt b/projects/clr/hipamd/CMakeLists.txt index 3c62ea4365..e018a5e4fe 100644 --- a/projects/clr/hipamd/CMakeLists.txt +++ b/projects/clr/hipamd/CMakeLists.txt @@ -207,9 +207,6 @@ if(HIP_PLATFORM STREQUAL "hcc") src/env.cpp src/program_state.cpp) - set(SOURCE_FILES_DEVICE - src/device_util.cpp) - execute_process(COMMAND ${HCC_HOME}/bin/hcc-config --ldflags OUTPUT_VARIABLE HCC_LD_FLAGS) set(CMAKE_SHARED_LINKER_FLAGS "${CMAKE_SHARED_LINKER_FLAGS} ${HCC_LD_FLAGS} -Wl,-Bsymbolic") set(CMAKE_SHARED_LINKER_FLAGS "${CMAKE_SHARED_LINKER_FLAGS} --amdgpu-target=gfx701 --amdgpu-target=gfx803 --amdgpu-target=gfx900 --amdgpu-target=gfx906") @@ -222,19 +219,18 @@ if(HIP_PLATFORM STREQUAL "hcc") target_link_libraries(hip_hcc PRIVATE hc_am) target_link_libraries(hip_hcc_static PRIVATE hc_am) endif() - add_library(hip_device STATIC ${SOURCE_FILES_DEVICE}) string(REPLACE " " ";" HCC_CXX_FLAGS_LIST ${HCC_CXX_FLAGS}) - foreach(TARGET hip_hcc hip_hcc_static hip_device) + foreach(TARGET hip_hcc hip_hcc_static) target_include_directories(${TARGET} SYSTEM INTERFACE $/include>;${HSA_PATH}/include) endforeach() add_library(host INTERFACE) target_link_libraries(host INTERFACE hip_hcc) add_library(device INTERFACE) if(HIP_COMPILER STREQUAL "hcc") - target_link_libraries(device INTERFACE host hip_device hcc::hccrt hcc::hc_am) - elseif(HIP_COMPILER STREQUAL "clang") - target_link_libraries(device INTERFACE host hip_device) + target_link_libraries(device INTERFACE host hcc::hccrt hcc::hc_am) + else() + target_link_libraries(device INTERFACE host) endif() # Generate .hipInfo @@ -256,7 +252,7 @@ endif() ############################# # Install hip_hcc if platform is hcc if(HIP_PLATFORM STREQUAL "hcc") - install(TARGETS hip_hcc_static hip_hcc hip_device DESTINATION lib) + install(TARGETS hip_hcc_static hip_hcc DESTINATION lib) # Install .hipInfo install(FILES ${PROJECT_BINARY_DIR}/.hipInfo DESTINATION lib) @@ -284,7 +280,7 @@ set(BIN_INSTALL_DIR ${CMAKE_INSTALL_PREFIX}/bin) set(CONFIG_PACKAGE_INSTALL_DIR ${LIB_INSTALL_DIR}/cmake/hip) if(HIP_PLATFORM STREQUAL "hcc") - install(TARGETS hip_hcc_static hip_hcc hip_device host device EXPORT hip-targets DESTINATION ${LIB_INSTALL_DIR}) + install(TARGETS hip_hcc_static hip_hcc host device EXPORT hip-targets DESTINATION ${LIB_INSTALL_DIR}) install(EXPORT hip-targets DESTINATION ${CONFIG_PACKAGE_INSTALL_DIR} NAMESPACE hip::) include(CMakePackageConfigHelpers) @@ -353,7 +349,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_device hip_hcc_static) + DEPENDS hip_hcc hip_hcc_static) # Package: hip_nvcc set(BUILD_DIR ${CMAKE_CURRENT_BINARY_DIR}/packages/hip_nvcc) diff --git a/projects/clr/hipamd/bin/hipcc b/projects/clr/hipamd/bin/hipcc index f22884ea1b..80234f8213 100755 --- a/projects/clr/hipamd/bin/hipcc +++ b/projects/clr/hipamd/bin/hipcc @@ -719,9 +719,9 @@ if ($setStdLib eq 0 and $HIP_PLATFORM eq 'hcc') if ($needHipHcc) { if ($linkType eq 0) { - substr($HIPLDFLAGS,0,0) = " $HIP_PATH/lib/libhip_hcc_static.a $HIP_PATH/lib/libhip_device.a " ; + substr($HIPLDFLAGS,0,0) = " $HIP_PATH/lib/libhip_hcc_static.a " ; } else { - substr($HIPLDFLAGS,0,0) = " -Wl,--rpath=$HIP_PATH/lib $HIP_PATH/lib/libhip_hcc.so $HIP_PATH/lib/libhip_device.a "; + substr($HIPLDFLAGS,0,0) = " -Wl,--rpath=$HIP_PATH/lib $HIP_PATH/lib/libhip_hcc.so "; } } diff --git a/projects/clr/hipamd/include/hip/hcc_detail/device_functions.h b/projects/clr/hipamd/include/hip/hcc_detail/device_functions.h index 42927e3246..974c9b4618 100644 --- a/projects/clr/hipamd/include/hip/hcc_detail/device_functions.h +++ b/projects/clr/hipamd/include/hip/hcc_detail/device_functions.h @@ -774,6 +774,15 @@ static void __threadfence_system() __atomic_work_item_fence(0, __memory_order_seq_cst, __memory_scope_all_svm_devices); } +// abort +__device__ +inline +__attribute__((weak)) +void abort() { + return __builtin_trap(); +} + + #endif // __HCC_OR_HIP_CLANG__ #ifdef __HCC__ diff --git a/projects/clr/hipamd/include/hip/hcc_detail/hip_runtime.h b/projects/clr/hipamd/include/hip/hcc_detail/hip_runtime.h index 3db06bb15e..60d145c884 100644 --- a/projects/clr/hipamd/include/hip/hcc_detail/hip_runtime.h +++ b/projects/clr/hipamd/include/hip/hcc_detail/hip_runtime.h @@ -181,8 +181,7 @@ extern int HIP_TRACE_API; #define __HCC_C__ #endif -// abort -__device__ void abort(); +__host__ inline void* __get_dynamicgroupbaseptr() { return nullptr; } #if __HIP_ARCH_GFX701__ == 0 diff --git a/projects/clr/hipamd/packaging/hip-targets-release.cmake b/projects/clr/hipamd/packaging/hip-targets-release.cmake index 0ae7405cde..83c207810c 100644 --- a/projects/clr/hipamd/packaging/hip-targets-release.cmake +++ b/projects/clr/hipamd/packaging/hip-targets-release.cmake @@ -41,15 +41,5 @@ endif() list(APPEND _IMPORT_CHECK_TARGETS hip::hip_hcc ) list(APPEND _IMPORT_CHECK_FILES_FOR_hip::hip_hcc "/opt/rocm/hip/lib/libhip_hcc.so" ) -# Import target "hip::hip_device" for configuration "Release" -set_property(TARGET hip::hip_device APPEND PROPERTY IMPORTED_CONFIGURATIONS RELEASE) -set_target_properties(hip::hip_device PROPERTIES - IMPORTED_LINK_INTERFACE_LANGUAGES_RELEASE "CXX" - IMPORTED_LOCATION_RELEASE "/opt/rocm/hip/lib/libhip_device.a" - ) - -list(APPEND _IMPORT_CHECK_TARGETS hip::hip_device ) -list(APPEND _IMPORT_CHECK_FILES_FOR_hip::hip_device "/opt/rocm/hip/lib/libhip_device.a" ) - # Commands beyond this point should not need to know the version. set(CMAKE_IMPORT_FILE_VERSION) diff --git a/projects/clr/hipamd/packaging/hip-targets.cmake b/projects/clr/hipamd/packaging/hip-targets.cmake index ec2fa716a6..d7a6b4d588 100644 --- a/projects/clr/hipamd/packaging/hip-targets.cmake +++ b/projects/clr/hipamd/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 hip::host hip::device) +foreach(_expectedTarget hip::hip_hcc_static hip::hip_hcc hip::host hip::device) list(APPEND _expectedTargets ${_expectedTarget}) if(NOT TARGET ${_expectedTarget}) list(APPEND _targetsNotDefined ${_expectedTarget}) @@ -57,14 +57,6 @@ set_target_properties(hip::hip_hcc PROPERTIES INTERFACE_SYSTEM_INCLUDE_DIRECTORIES "${_IMPORT_PREFIX}/include;/opt/rocm/hsa/include" ) -# Create imported target hip::hip_device -add_library(hip::hip_device STATIC IMPORTED) - -set_target_properties(hip::hip_device PROPERTIES - INTERFACE_INCLUDE_DIRECTORIES "${_IMPORT_PREFIX}/include;/opt/rocm/hsa/include" - INTERFACE_SYSTEM_INCLUDE_DIRECTORIES "${_IMPORT_PREFIX}/include;/opt/rocm/hsa/include" -) - # Create imported target hip::host add_library(hip::host INTERFACE IMPORTED) @@ -75,13 +67,13 @@ set_target_properties(hip::host PROPERTIES # Create imported target hip::device add_library(hip::device INTERFACE IMPORTED) -if(HIP_COMPILER STREQUAL "clang") +if(HIP_COMPILER STREQUAL "hcc") set_target_properties(hip::device PROPERTIES - INTERFACE_LINK_LIBRARIES "hip::host;hip::hip_device" + INTERFACE_LINK_LIBRARIES "hip::host;hcc::hccrt;hcc::hc_am" ) else() set_target_properties(hip::device PROPERTIES - INTERFACE_LINK_LIBRARIES "hip::host;hip::hip_device;hcc::hccrt;hcc::hc_am" + INTERFACE_LINK_LIBRARIES "hip::host" ) endif() diff --git a/projects/clr/hipamd/packaging/hip_hcc.txt b/projects/clr/hipamd/packaging/hip_hcc.txt index 9d4b96761d..fe866e47f9 100644 --- a/projects/clr/hipamd/packaging/hip_hcc.txt +++ b/projects/clr/hipamd/packaging/hip_hcc.txt @@ -3,7 +3,6 @@ project(hip_hcc) install(FILES @PROJECT_BINARY_DIR@/libhip_hcc.so DESTINATION lib) install(FILES @PROJECT_BINARY_DIR@/libhip_hcc_static.a DESTINATION lib) -install(FILES @PROJECT_BINARY_DIR@/libhip_device.a DESTINATION lib) install(FILES @PROJECT_BINARY_DIR@/.hipInfo DESTINATION lib) install(FILES @PROJECT_BINARY_DIR@/hip-config.cmake @PROJECT_BINARY_DIR@/hip-config-version.cmake DESTINATION lib/cmake/hip) install(FILES @hip_SOURCE_DIR@/packaging/hip-targets.cmake @hip_SOURCE_DIR@/packaging/hip-targets-release.cmake DESTINATION lib/cmake/hip) diff --git a/projects/clr/hipamd/src/device_util.cpp b/projects/clr/hipamd/src/device_util.cpp deleted file mode 100644 index c86e52617b..0000000000 --- a/projects/clr/hipamd/src/device_util.cpp +++ /dev/null @@ -1,33 +0,0 @@ -/* -Copyright (c) 2015-2017 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. -*/ - -#include -#include "device_util.h" -#include "hip/hcc_detail/device_functions.h" -#include "hip/hcc_detail/grid_launch.h" -#include "hip/hip_runtime.h" -#include - -// abort -__device__ void abort() { return hc::abort(); } - -__host__ void* __get_dynamicgroupbaseptr() { return nullptr; }