Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Jump to bottom

Enable building to target AMD GPUs #1731

Draft
wants to merge 3 commits into
base: master
Choose a base branch
from
Draft

Enable building to target AMD GPUs #1731

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
3 commits
Select commit Hold shift + click to select a range
5ff2c99
Enable CMake options for building for AMD
ndgrigorian Jul 13, 2024
15e0def
Refactor cmake for setting AMD targets and fix incorrect logic
ndgrigorian Jul 13, 2024
c53fb26
Use `sycl::log1p` in `math_utils.hpp`
ndgrigorian Aug 15, 2024
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
45 changes: 42 additions & 3 deletions CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -29,24 +29,63 @@ option(DPCTL_TARGET_CUDA
"Build DPCTL to target CUDA devices"
OFF
)
set(DPCTL_TARGET_AMD
""
CACHE STRING
"Build DPCTL to target an AMD device architecture"
)

find_package(IntelSYCL REQUIRED PATHS ${CMAKE_SOURCE_DIR}/cmake NO_DEFAULT_PATH)

set(_dpctl_sycl_target_compile_options)
set(_dpctl_sycl_target_link_options)

set(_dpctl_sycl_targets)
set(_dpctl_amd_targets)
if ("x${DPCTL_SYCL_TARGETS}" STREQUAL "x")
if(DPCTL_TARGET_CUDA)
if (DPCTL_TARGET_CUDA)
set(_dpctl_sycl_targets "nvptx64-nvidia-cuda,spir64-unknown-unknown")
else()
if(DEFINED ENV{DPCTL_TARGET_CUDA})
if (DEFINED ENV{DPCTL_TARGET_CUDA})
set(_dpctl_sycl_targets "nvptx64-nvidia-cuda,spir64-unknown-unknown")
endif()
endif()
if (NOT "x${DPCTL_TARGET_AMD}" STREQUAL "x")
set(_dpctl_amd_targets ${DPCTL_TARGET_AMD})
if(_dpctl_sycl_targets)
set(_dpctl_sycl_targets "amdgcn-amd-amdhsa,${_dpctl_sycl_targets}")
else()
set(_dpctl_sycl_targets "amdgcn-amd-amdhsa,spir64-unknown-unknown")
endif()
else()
if (DEFINED ENV{DPCTL_TARGET_AMD})
set(_dpctl_amd_targets $ENV{DPCTL_TARGET_AMD})
if(_dpctl_sycl_targets)
set(_dpctl_sycl_targets "amdgcn-amd-amdhsa,${_dpctl_sycl_targets}")
else()
set(_dpctl_sycl_targets "amdgcn-amd-amdhsa,spir64-unknown-unknown")
endif()
endif()
endif()
else()
set(_dpctl_sycl_targets ${DPCTL_SYCL_TARGETS})
if (NOT "x${DPCTL_TARGET_AMD}" STREQUAL "x")
set(_dpctl_amd_targets ${DPCTL_TARGET_AMD})
else()
if (DEFINED ENV{DPCTL_TARGET_AMD})
set(_dpctl_amd_targets $ENV{DPCTL_TARGET_AMD})
endif()
endif()
endif()

if(_dpctl_sycl_targets)
if (_dpctl_sycl_targets)
message(STATUS "Compiling for -fsycl-targets=${_dpctl_sycl_targets}")
list(APPEND _dpctl_sycl_target_compile_options -fsycl-targets=${_dpctl_sycl_targets})
list(APPEND _dpctl_sycl_target_link_options -fsycl-targets=${_dpctl_sycl_targets})
if(_dpctl_amd_targets)
list(APPEND _dpctl_sycl_target_compile_options -Xsycl-target-backend=amdgcn-amd-amdhsa --offload-arch=${_dpctl_amd_targets})
list(APPEND _dpctl_sycl_target_link_options -Xsycl-target-backend=amdgcn-amd-amdhsa --offload-arch=${_dpctl_amd_targets})
endif()
endif()

add_subdirectory(libsyclinterface)
Expand Down
4 changes: 2 additions & 2 deletions dpctl/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -116,12 +116,12 @@ function(build_dpctl_ext _trgt _src _dest)
target_compile_options(
${_trgt}
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_compile_options}
)
target_link_options(
${_trgt}
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_link_options}
)
endif()
endif()
Expand Down
4 changes: 2 additions & 2 deletions dpctl/tensor/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -272,12 +272,12 @@ foreach(python_module_name ${_py_trgts})
target_compile_options(
${python_module_name}
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_compile_options}
)
target_link_options(
${python_module_name}
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_link_options}
)
endif()
# TODO: update source so they refernece individual libraries instead of
Expand Down
4 changes: 2 additions & 2 deletions dpctl/tensor/libtensor/include/utils/math_utils.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -126,10 +126,10 @@ template <typename T> T logaddexp(T x, T y)
// compiler segfault in CUDA build is fixed
const T tmp = x - y;
if (tmp > 0) {
return x + std::log1p(sycl::exp(-tmp));
return x + sycl::log1p(sycl::exp(-tmp));
}
else if (tmp <= 0) {
return y + std::log1p(sycl::exp(tmp));
return y + sycl::log1p(sycl::exp(tmp));
}
else {
return std::numeric_limits<T>::quiet_NaN();
Expand Down
31 changes: 17 additions & 14 deletions dpctl/utils/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -21,18 +21,21 @@ if(DPCTL_GENERATE_COVERAGE)
PRIVATE -fprofile-instr-generate -fcoverage-mapping
)
endif()
if(_dpctl_sycl_targets)
# make fat binary
target_compile_options(
${python_module_name}
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
)
target_link_options(
${python_module_name}
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
)
if (BUILD_DPCTL_EXT_SYCL)
add_sycl_to_target(TARGET ${_trgt} SOURCES ${_generated_src})
if(_dpctl_sycl_targets)
# make fat binary
target_compile_options(
${python_module_name}
PRIVATE
${_dpctl_sycl_target_compile_options}
)
target_link_options(
${python_module_name}
PRIVATE
${_dpctl_sycl_target_link_options}
)
endif()
endif()
target_link_libraries(${python_module_name} PRIVATE DpctlCAPI)
install(TARGETS ${python_module_name} DESTINATION "dpctl/utils")
Expand Down Expand Up @@ -60,12 +63,12 @@ if(_dpctl_sycl_targets)
target_compile_options(
${python_module_name}
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_compile_options}
)
target_link_options(
${python_module_name}
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_link_options}
)
endif()
target_link_libraries(${python_module_name} PRIVATE DpctlCAPI)
Expand Down
6 changes: 3 additions & 3 deletions libsyclinterface/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -244,13 +244,13 @@ if(_dpctl_sycl_targets)
target_compile_options(
DPCTLSyclInterface
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_compile_options}
)
target_link_options(
DPCTLSyclInterface
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
)
${_dpctl_sycl_target_link_options}
)
endif()

if(DPCTL_GENERATE_COVERAGE)
Expand Down
10 changes: 4 additions & 6 deletions libsyclinterface/tests/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -55,16 +55,15 @@ add_sycl_to_target(
)

if(_dpctl_sycl_targets)
# make fat binary
target_compile_options(
dpctl_c_api_tests
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_compile_options}
)
target_link_options(
dpctl_c_api_tests
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_link_options}
)
endif()

Expand All @@ -85,16 +84,15 @@ target_include_directories(dpctl_c_api_tests
)

if(_dpctl_sycl_targets)
# make fat binary
target_compile_options(
dpctl_c_api_tests
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_compile_options}
)
target_link_options(
dpctl_c_api_tests
PRIVATE
-fsycl-targets=${_dpctl_sycl_targets}
${_dpctl_sycl_target_link_options}
)
endif()

Expand Down
Loading