Re: [PATCH] drm/amdkfd: don't widen the validate GPU set past one page owner

[Philip Yang <yangp@amd.com>]

[-- Attachment #1: Type: text/plain, Size: 5834 bytes --]



On 2026-09-15 10:58, Andrew Martin wrote:
> Commit 6b2969831f7e ("drm/amdkfd: Unmap svm range from GPU set to
> no-access") made svm_range_validate_and_map() OR bitmap_access into
> ctx->bitmap unconditionally for a range carrying
> KFD_IOCTL_SVM_FLAG_GPU_ALWAYS_MAPPED.  Previously that only happened when
> ctx->bitmap was still empty:
>
> 	if (bitmap_empty(ctx->bitmap, MAX_GPU_INSTANCE)) {
> 		if (prange->mapped_to_gpu ||
> 		    prange->flags & KFD_IOCTL_SVM_FLAG_GPU_ALWAYS_MAPPED)
> 			bitmap_copy(ctx->bitmap, prange->bitmap_access,
> 				    MAX_GPU_INSTANCE);
> 	}
>
> The block just above already sets the bit for prange->actual_loc when the
> range has been migrated to VRAM, so under the old code ctx->bitmap was
> left holding that single GPU and the OR was skipped.  Without the guard
> it now also picks up every other GPU with the ACCESS attribute.
>
> That matters because ctx->bitmap is not only the set of GPUs to map.  It
> also selects the HMM owner: owner is taken from the first set bit and
> reset to NULL as soon as two set bits disagree.  On a two GPU system with
> ACCESS on both, owner becomes NULL, and amdgpu_hmm_range_get_pages() then
> no longer recognises the process's own device-private pages.  HMM calls
> migrate_to_ram on each of them, svm_migrate_to_ram() sees
> p->svms.faulting_task == current and returns 0 without migrating, and
> hmm_range_fault() retries the same page forever.
>
> Which GPUs disagree depends on the interconnect.  SVM_ADEV_PGMAP_OWNER()
> is the XGMI hive when the device has one and the adev otherwise, so GPUs
> sharing a hive also share an owner and never trip this.  It takes two or
> more GPUs with ACCESS that are not in a common hive, which is why this
> only shows up on multi-GPU machines with no XGMI link between the cards.
>
> On a 2x Instinct MI350P (gfx950) this hangs
>
>    HSA_XNACK=1 hipblaslt-test \
>      --gtest_filter=aux_handle_test.set_sm_count_target_default_is_zero
>
> The application never completes and burns 100% of a CPU in the kernel.
> /proc/<pid>/stat shows utime frozen while stime climbs and minflt advances
> at roughly 12 million faults per second, reaching billions; majflt stays
> at 0 and the GPU sits at 0% busy.  perf on the stuck task is 24%
> svm_migrate_to_ram, 23% __handle_mm_fault, 5% do_swap_page.  The process
> cannot be killed and the machine has to be rebooted.
>
> Restore the bitmap_empty() guard.  The rest of that commit is kept: the
> bitmap_mapped tracking, bitmap_needs_unmap, mapping_done and the unmap of
> ranges set to no-access are all unaffected, and bitmap_mapped is still
> ORed in at the top of the branch, so a range already mapped on a GPU
> still has its mapping updated there.
>
> The instrumentation, the hardware testing and this changelog were done
> with the assistance of an AI coding agent; the resulting change has been
> reviewed by hand.
>
> Fixes: 6b2969831f7e ("drm/amdkfd: Unmap svm range from GPU set to no-access")
The fix tag is correct, suggest to change patch title and commit message

drm/amdkfd: Fix always mapped range mapping to all ACCESS GPUs

     CWSR ranges use the always mapped flag and only need a mapping on the
     GPU running the queue. However, svm_range_validate_and_map now updates
     the mapping on every GPU with the ACCESS attribute, which is all
     supported GPUs when XNACK is on. After the range is prefetched to VRAM,
     those GPUs may span different XGMI hives, so the page owner mismatches
     and is set to NULL. hmm_range_fault then tries to migrate the VRAM
     and is set to NULL. hmm_range_fault then tries to migrate the VRAM
     pages back to system memory, but svm_migrate_to_ram skips it because
     migrating pages back to system memory is not allowed while the GPU
     mapping is being updated, so the CPU page fault is never resolved.

     Only map to the ACCESS GPUs if the range is not mapped to any GPU and
     is not in VRAM.

> Signed-off-by: Andrew Martin<andrew.martin@amd.com>
> Assisted-by: Claude:claude-opus-5
> ---
>   drivers/gpu/drm/amd/amdkfd/kfd_svm.c | 9 +++++++--
>   1 file changed, 7 insertions(+), 2 deletions(-)
>
> diff --git a/drivers/gpu/drm/amd/amdkfd/kfd_svm.c b/drivers/gpu/drm/amd/amdkfd/kfd_svm.c
> index c521ec07bf3c..75e25f075da3 100644
> --- a/drivers/gpu/drm/amd/amdkfd/kfd_svm.c
> +++ b/drivers/gpu/drm/amd/amdkfd/kfd_svm.c
> @@ -1758,9 +1758,14 @@ static int svm_range_validate_and_map(struct mm_struct *mm,
>   
>   		/*
>   		 * If prange with always mapped flag, update mapping on GPUs with
> -		 * ACCESS attribute
> +		 * ACCESS attribute.  Only widen the set when nothing else has
> +		 * selected a GPU: ctx->bitmap also picks the owner passed to
> +		 * amdgpu_hmm_range_get_pages() below, and spanning GPUs with
> +		 * different page owners forces that owner to NULL, which makes
> +		 * HMM fault on the process's own migrated pages.
>   		 */
> -		if (prange->flags & KFD_IOCTL_SVM_FLAG_GPU_ALWAYS_MAPPED)
> +		if (bitmap_empty(ctx->bitmap, MAX_GPU_INSTANCE) &&
> +		    prange->flags & KFD_IOCTL_SVM_FLAG_GPU_ALWAYS_MAPPED)
>   			bitmap_or(ctx->bitmap, ctx->bitmap, prange->bitmap_access,
>   				  MAX_GPU_INSTANCE);
         /*
          * If prange with always mapped flag is not mapped or migrated
          * yet, update mapping on GPUs with ACCESS attribute. Don't add
          * them otherwise, ACCESS GPUs may be on different XGMI hives,
          * that would force hmm_range_fault to migrate VRAM back to ram.
          */
         if (bitmap_empty(ctx->bitmap, MAX_GPU_INSTANCE) &&
             prange->flags & KFD_IOCTL_SVM_FLAG_GPU_ALWAYS_MAPPED)
             bitmap_copy(ctx->bitmap, prange->bitmap_access,
                     MAX_GPU_INSTANCE);

Regards,
Philip
>   	} else {

[-- Attachment #2: Type: text/html, Size: 7211 bytes --]