[ROCm] Build CK GEMM/SDPA per-arch and disable them for gfx1250

[pragupta]

composable_kernel has no gfx1250 support, so its CK GEMM and CK SDPA kernels fail to compile when targeting that arch. Previously the only lever was the global USE_ROCM_CK_GEMM / USE_ROCM_CK_SDPA options, so disabling CK for gfx1250 also disabled it for every other arch in the same (commonly multi-arch) build.

Instead, build the CK kernels for every requested arch except gfx1250 by filtering --offload-arch on the CK targets, and only fall back to fully disabling CK when gfx1250 is the sole arch.

1. aten/src/ATen/CMakeLists.txt

   • gfx1250-only guard: when gfx1250 is the only entry in PYTORCH_ROCM_ARCH there is no supported arch left to build CK for, so turn USE_ROCM_CK_GEMM and USE_ROCM_CK_SDPA off (via caffe2_update_option, which writes the cache and is therefore seen by caffe2/CMakeLists.txt). This runs before the CK blocks below.
   • ck_sdpa: filter --offload-arch to drop gfx1250.
   • CK GEMM: move the ck*.hip / bgemm_kernels sources out of the main torch_hip sources into a dedicated, --offload-arch-filtered ck_gemm library (mirrors the existing mslk/ck_sdpa pattern). The library gets USE_ROCM_CK_GEMM defined on the target since it is no longer a global flag.

2. cmake/Dependencies.cmake

   • Stop adding -DUSE_ROCM_CK_GEMM to the global HIP flags. Applied unconditionally here it ran before aten/src/ATen/CMakeLists.txt, so the gfx1250-only guard could not retract it.

3. caffe2/CMakeLists.txt

   • Apply USE_ROCM_CK_GEMM as a conditional target definition on torch_hip and link ck_gemm, mirroring how USE_ROCM_CK_SDPA is already handled. This makes the define honor the gfx1250-only guard.

Net behavior: a multi-arch build containing gfx1250 keeps CK GEMM/SDPA for the other archs and skips only gfx1250; a gfx1250-only build cleanly builds with CK GEMM/SDPA disabled (no empty --offload-arch compile and no unresolved CK symbols); builds without gfx1250 are unchanged.

This change was co-authored with Claude.

(cherry picked from commit 92e6d4e)

Motivation

Technical Details

Test Plan

Test Result

Submission Checklist

• Look over the contributing guidelines at https://github.com/ROCm/ROCm/blob/develop/CONTRIBUTING.md#pull-requests.