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: ba43d1ce1b]
이 커밋은 다음에 포함됨:
Aaron Enye Shi
@@ -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 $<INSTALL_INTERFACE:$<INSTALL_PREFIX>/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)
|
||||
|
||||
@@ -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 ";
|
||||
}
|
||||
}
|
||||
|
||||
|
||||
@@ -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__
|
||||
|
||||
@@ -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
|
||||
|
||||
|
||||
@@ -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)
|
||||
|
||||
@@ -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()
|
||||
|
||||
|
||||
@@ -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)
|
||||
|
||||
@@ -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 <hc.hpp>
|
||||
#include "device_util.h"
|
||||
#include "hip/hcc_detail/device_functions.h"
|
||||
#include "hip/hcc_detail/grid_launch.h"
|
||||
#include "hip/hip_runtime.h"
|
||||
#include <atomic>
|
||||
|
||||
// abort
|
||||
__device__ void abort() { return hc::abort(); }
|
||||
|
||||
__host__ void* __get_dynamicgroupbaseptr() { return nullptr; }
|
||||
새 이슈에서 참조
사용자 차단