AMD General cc @Zhao, Huasha @Gu, Zhenyu
-----Original Message----- From: Yang, Philip <[email protected]> Sent: Monday, November 9, 2026 1:15 PM To: [email protected]; Kuehling, Felix <[email protected]>; Gandham, Shashidhar <[email protected]>; Han, Qiang <[email protected]> Cc: Han, Qiang <[email protected]>; Yang, Philip <[email protected]> Subject: [PATCH] drm/amdkfd: wait for the queue quiesce in concurrent SVM evictions From: Qiang Han <[email protected]> With XNACK off, svm_range_evict() leaves the GPU page tables in place and relies on the process's queues being stopped before the MMU notifier returns. Only the first eviction (evicted_ranges == 1) calls kgd2kfd_quiesce_mm(). The quiesce preempts every GPU the process has opened and takes milliseconds. Any other range invalidated in the meantime sees evicted_ranges > 1 and returns at once. The core mm then frees, migrates or reuses the page while the queues still run, and an in-flight SDMA or blit copy accesses it through the old GPU PTE. The result is silent corruption of the copy, or of whatever process now owns the page. Hold a new svms->evict_lock across the first evictor's quiesce, so that concurrent evictions return only after the queues have stopped. This is the scheme amdgpu_amdkfd_evict_userptr() uses with process_info->notifier_lock. evict_lock nests inside prange->lock and is taken nowhere else. Under it, the quiesce path takes the dqm lock, which is already used in MMU notifiers (memalloc_noreclaim). One way to reach this: a pageable hipMemcpy() larger than 1 MiB. ROCclr pins the buffer with hsa_amd_memory_lock(), libhsakmt registers it as a KFD SVM range, and host memory pressure (reclaim, compaction, allocator purge) invalidates the range. Fixes: 8a7c184a16d7 ("drm/amdkfd: svm range eviction and restore") Signed-off-by: Qing Hang <[email protected]> Signed-off-by: Philip Yang <[email protected]> --- drivers/gpu/drm/amd/amdkfd/kfd_priv.h | 2 ++ drivers/gpu/drm/amd/amdkfd/kfd_svm.c | 35 ++++++++++++++++++--------- 2 files changed, 25 insertions(+), 12 deletions(-) diff --git a/drivers/gpu/drm/amd/amdkfd/kfd_priv.h b/drivers/gpu/drm/amd/amdkfd/kfd_priv.h index 158fa70a1a8e..581fb7ceba40 100644 --- a/drivers/gpu/drm/amd/amdkfd/kfd_priv.h +++ b/drivers/gpu/drm/amd/amdkfd/kfd_priv.h @@ -909,6 +909,8 @@ struct svm_range_list { struct list_head deferred_range_list; struct list_head criu_svm_metadata_list; spinlock_t deferred_list_lock; + /* serializes XNACK-off evictions with the queue quiesce */ + struct mutex evict_lock; atomic_t evicted_ranges; atomic_t drain_pagefaults; struct delayed_work restore_work; diff --git a/drivers/gpu/drm/amd/amdkfd/kfd_svm.c b/drivers/gpu/drm/amd/amdkfd/kfd_svm.c index 3688ae7ba35e..8b84333a85b7 100644 --- a/drivers/gpu/drm/amd/amdkfd/kfd_svm.c +++ b/drivers/gpu/drm/amd/amdkfd/kfd_svm.c @@ -2141,21 +2141,30 @@ svm_range_evict(struct svm_range *prange, struct mm_struct *mm, if (prange->start <= last && prange->last >= start) atomic_inc(&prange->invalid); + /* The MMU notifier must not return before the GPU has stopped + * accessing the invalidated pages. Only the first eviction stops + * the queues, which takes milliseconds (HWS preemption on every + * GPU of the process). Serialize all evictions of this process so + * that a concurrent invalidation of another range waits until + * that quiesce has completed instead of returning while queues + * still run on stale PTEs. + */ + mutex_lock(&svms->evict_lock); evicted_ranges = atomic_inc_return(&svms->evicted_ranges); - if (evicted_ranges != 1) - return r; - - pr_debug("evicting svms 0x%p range [0x%lx 0x%lx]\n", - prange->svms, prange->start, prange->last); + if (evicted_ranges == 1) { + pr_debug("evicting svms 0x%p range [0x%lx 0x%lx]\n", + prange->svms, prange->start, prange->last); - /* First eviction, stop the queues */ - r = kgd2kfd_quiesce_mm(mm, KFD_QUEUE_EVICTION_TRIGGER_SVM); - if (r) - pr_debug("failed to quiesce KFD\n"); + /* First eviction, stop the queues */ + r = kgd2kfd_quiesce_mm(mm, KFD_QUEUE_EVICTION_TRIGGER_SVM); + if (r) + pr_debug("failed to quiesce KFD\n"); - pr_debug("schedule to restore svm %p ranges\n", svms); - queue_delayed_work(system_freezable_wq, &svms->restore_work, - msecs_to_jiffies(AMDGPU_SVM_RANGE_RESTORE_DELAY_MS)); + pr_debug("schedule to restore svm %p ranges\n", svms); + queue_delayed_work(system_freezable_wq, &svms->restore_work, + msecs_to_jiffies(AMDGPU_SVM_RANGE_RESTORE_DELAY_MS)); + } + mutex_unlock(&svms->evict_lock); } else { unsigned long s, l; uint32_t trigger; @@ -3437,6 +3446,7 @@ void svm_range_list_fini(struct kfd_process *p) svm_range_free(prange, true); } + mutex_destroy(&p->svms.evict_lock); mutex_destroy(&p->svms.lock); pr_debug("process pid %d svms 0x%p done\n", @@ -3450,6 +3460,7 @@ int svm_range_list_init(struct kfd_process *p) svms->objects = RB_ROOT_CACHED; mutex_init(&svms->lock); + mutex_init(&svms->evict_lock); INIT_LIST_HEAD(&svms->list); atomic_set(&svms->evicted_ranges, 0); atomic_set(&svms->drain_pagefaults, 0); -- 2.50.1
