gnu: Add composable-kernel.

* gnu/packages/rocm-libs.scm (composable-kernel): New variable.
* gnu/packages/patches/composable-kernel-conditional-kernels.patch: New patch.
* gnu/local.mk (dist_patch_DATA): Add it.

Signed-off-by: Ludovic Courtès <ludo@gnu.org>
Modified-by: Ludovic Courtès <ludo@gnu.org>

3 files changed, 62 insertions, 0 deletions

diff --git a/gnu/local.mk b/gnu/local.mk
index 576cc1ee4b0..24425d0281b 100644
--- a/gnu/local.mk
+++ b/gnu/local.mk
@@ -1129,6 +1129,7 @@ dist_patch_DATA =						\
   %D%/packages/patches/collectd-5.11.0-python-3.11.patch	\
   %D%/packages/patches/combinatorial-blas-awpm.patch		\
   %D%/packages/patches/combinatorial-blas-io-fix.patch		\
+  %D%/packages/patches/composable-kernel-conditional-kernels.patch \
   %D%/packages/patches/compsize-fix-btrfs-progs-compatibility.patch	\
   %D%/packages/patches/containerd-create-pid-file.patch	\
   %D%/packages/patches/containerd-fix-includes.patch            \
diff --git a/gnu/packages/patches/composable-kernel-conditional-kernels.patch b/gnu/packages/patches/composable-kernel-conditional-kernels.patch
new file mode 100644
index 00000000000..4182bd0fb4c
--- /dev/null
+++ b/gnu/packages/patches/composable-kernel-conditional-kernels.patch
@@ -0,0 +1,18 @@
+Taken from: https://github.com/gentoo/gentoo/blob/91670e3163c4f7daa9821da683fb16875845f263/sci-libs/composable-kernel/files/composable-kernel-6.3.0-conditional-kernels.patch
+
+Fix component discovery, when some targets are not installed.
+Upstream bug: https://github.com/ROCm/composable_kernel/issues/1646
+--- a/Config.cmake.in
++++ b/Config.cmake.in
+@@ -6,6 +6,10 @@ foreach(_comp ${composable_kernel_FIND_COMPONENTS})
+ 	if(NOT _comp IN_LIST _composable_kernel_supported_components)
+ 		set(composable_kernel_FOUND False)
+ 		set(composable_kernel_NOT_FOUND_MESSAGE "Unsupported component: ${_comp}")
++	elseif(NOT EXISTS "${CMAKE_CURRENT_LIST_DIR}/composable_kernel${_comp}Targets.cmake")
++		set(composable_kernel_FOUND False)
++		set(composable_kernel_NOT_FOUND_MESSAGE "Unsupported component for current targets: ${_comp}")
++	else()
++		include("${CMAKE_CURRENT_LIST_DIR}/composable_kernel${_comp}Targets.cmake")
+ 	endif()
+-	include("${CMAKE_CURRENT_LIST_DIR}/composable_kernel${_comp}Targets.cmake")
+ endforeach()
diff --git a/gnu/packages/rocm-libs.scm b/gnu/packages/rocm-libs.scm
index 5432d96c591..124c5bac497 100644
--- a/gnu/packages/rocm-libs.scm
+++ b/gnu/packages/rocm-libs.scm
@@ -729,3 +729,46 @@ primitives, in particular via rocPRIM for AMD GPUs.")
     (description "@code{rocThrust} is a parallel algorithm library for ROCm,
 based on @code{rocPRIM}.  It is a port of the CUDA Thrust library.")
     (license license:asl2.0)))
+
+(define-public composable-kernel
+  (package
+    (name "composable-kernel")
+    (version %rocm-version)
+    (source
+     (rocm-library-source
+      "composablekernel"
+      #:patches (search-patches "composable-kernel-conditional-kernels.patch")))
+    (build-system cmake-build-system)
+    (arguments
+     (list
+      #:tests? #f
+      #:build-type "Release"
+      #:configure-flags
+      #~(list
+         "-DCMAKE_CXX_COMPILER=hipcc"
+         "-DBUILD_TESTING=OFF"
+         #$(string-append "-DGPU_TARGETS=" (current-amd-gpu-targets-string)))
+      #:phases
+      #~(modify-phases %standard-phases
+          (add-after 'unpack 'patch-cmake
+            (lambda _
+              (substitute* "CMakeLists.txt"
+                (("find_package\\(Git.*") "")
+                (("execute_process\\(COMMAND.*GIT_EXECUTABLE.*")
+                 (string-append
+                  "set(COMMIT_ID " #$%rocm-libraries-commit ")\n"))
+                ;; Reduce memory consumption:
+                ;; https://gitweb.gentoo.org/repo/gentoo.git/tree/sci-libs/composable-kernel/composable-kernel-7.1.0.ebuild?id=a772170f741d75565652cb319a4e37eeb7efe6e9#n86
+                ((".*(-amdgpu-early-inline-all|-amdgpu-function-calls).*") ""))))
+          (add-after 'unpack 'setenv
+            (lambda _
+              (setenv "HIP_PATH" #$(this-package-input "rocm-hip-runtime")))))))
+    (inputs (list rocm-hip-runtime))
+    (native-inputs (list python rocm-cmake rocm-toolchain))
+    (properties `((amd-gpu-targets . ,%default-amd-gpu-targets)))
+    (home-page %rocm-libraries-url)
+    (synopsis "Library for writing performance-portable kernels")
+    (description "Composable Kernel is a library which provides a tile-based
+programming model for writing performance-portable kernels, in particular for
+GPUs using the HIP programming language.")
+    (license license:expat)))