[CK_TILE] Fix Windows build in FMHA head grouping#5977
Merged
brockhargreaves-amd merged 2 commits intodevelopfrom Mar 30, 2026
Merged
[CK_TILE] Fix Windows build in FMHA head grouping#5977brockhargreaves-amd merged 2 commits intodevelopfrom
brockhargreaves-amd merged 2 commits intodevelopfrom
Conversation
Contributor
Author
|
@brockhargreaves-amd This PR will fix the Windows build. Could you take a look? |
Contributor
|
Windows Strix Halo should support this, no? |
|
This PR was created after building FlashAttention 2's CK backend on Windows. @astrelsky, do you think these changes are also necessary, or are the ones in this PR sufficient? We could get this one merged instead since @brockhargreaves-amd is already here. |
1 task
Contributor
Author
Yes it should support Windows Strix Halo. I think the previous CI error is coming from the other component like miopen, not related to this pr. |
brockhargreaves-amd
approved these changes
Mar 30, 2026
|
@hyoon1 Thanks for your work on Dao-AILab/flash-attention#2400 and #5018 as well! Also, thx to the maintainers for the reviews. After this PR is merged, the only remaining blocker for building FlashAttention's CK backend on Windows will be pypa/distutils#406. |
poyenc
approved these changes
Mar 30, 2026
assistant-librarian bot
pushed a commit
to ROCm/composable_kernel
that referenced
this pull request
Mar 30, 2026
[CK_TILE] Fix Windows build in FMHA head grouping ## Motivation This is a follow-up fix for [PR #5018](ROCm/rocm-libraries#5018). [PR #5018](ROCm/rocm-libraries#5018) added LLC-aware FMHA head grouping / head-major scheduling on RDNA, but it also introduced Linux-only code paths, including `<dirent.h>`, which break Windows builds. This change fixes that by guarding the Linux-specific LLC probing logic so non-Linux platforms can still build correctly. ## Technical Details - Guard `<dirent.h>` with `#ifdef __linux__` - Guard KFD sysfs traversal logic with `#if defined(__linux__)` - On non-Linux platforms, return `0` from `get_kfd_sysfs_llc_cache_bytes()` - Preserve existing fallback behavior through: - `CK_TILE_FMHA_LLC_CACHE_MB` - arch-based default LLC sizes - no head grouping when no LLC size can be resolved ## Test Plan <!-- Explain any relevant testing done to verify this PR. --> ## Test Result <!-- Briefly summarize test outcomes. --> ## Submission Checklist - [ ] Look over the contributing guidelines at https://github.com/ROCm/ROCm/blob/develop/CONTRIBUTING.md#pull-requests.
vidyasagar-amd
pushed a commit
that referenced
this pull request
Apr 9, 2026
## Motivation This is a follow-up fix for [PR #5018](#5018). [PR #5018](#5018) added LLC-aware FMHA head grouping / head-major scheduling on RDNA, but it also introduced Linux-only code paths, including `<dirent.h>`, which break Windows builds. This change fixes that by guarding the Linux-specific LLC probing logic so non-Linux platforms can still build correctly. ## Technical Details - Guard `<dirent.h>` with `#ifdef __linux__` - Guard KFD sysfs traversal logic with `#if defined(__linux__)` - On non-Linux platforms, return `0` from `get_kfd_sysfs_llc_cache_bytes()` - Preserve existing fallback behavior through: - `CK_TILE_FMHA_LLC_CACHE_MB` - arch-based default LLC sizes - no head grouping when no LLC size can be resolved ## Test Plan <!-- Explain any relevant testing done to verify this PR. --> ## Test Result <!-- Briefly summarize test outcomes. --> ## Submission Checklist - [ ] Look over the contributing guidelines at https://github.com/ROCm/ROCm/blob/develop/CONTRIBUTING.md#pull-requests.
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
4 participants
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
Motivation
This is a follow-up fix for PR #5018.
PR #5018 added LLC-aware FMHA head grouping / head-major scheduling on RDNA, but it also introduced Linux-only code paths, including
<dirent.h>, which break Windows builds. This change fixes that by guarding the Linux-specific LLC probing logic so non-Linux platforms can still build correctly.Technical Details
<dirent.h>with#ifdef __linux__#if defined(__linux__)0fromget_kfd_sysfs_llc_cache_bytes()CK_TILE_FMHA_LLC_CACHE_MBTest Plan
Test Result
Submission Checklist