Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

SWDEV-506793 - segfault in ihipMallocManaged when no devices are available #122

Open
wants to merge 2 commits into
base: amd-staging
Choose a base branch
from
Open

SWDEV-506793 - segfault in ihipMallocManaged when no devices are available #122

Show file tree
Hide file tree
Changes from 1 commit
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
af828bb
Do not allocate managed memory if no devices available
benrichard-amd Jan 13, 2025
d7ff152
Check device count in _hipRegisterManagedVar
benrichard-amd Jan 13, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
4 changes: 4 additions & 0 deletions hipamd/src/hip_hmm.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -287,6 +287,10 @@ hipError_t ihipMallocManaged(void** ptr, size_t size, unsigned int align) {
assert((hip::host_context != nullptr) && "Current host context must be valid");
amd::Context& ctx = *hip::host_context;

if (ctx.devices().size() == 0) {
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

HIP_INIT_API macro already validates the active devices. I don't see why we need extra validation, unless something else is broken.

#define HIP_INIT_API(cid, ...)
HIP_INIT_API_INTERNAL(0, cid, VA_ARGS)
if (hip::g_devices.size() == 0) {
HIP_RETURN(hipErrorNoDevice);
}

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Just validate the devices in __hipRegisterManagedVar() itself.

void __hipRegisterManagedVar(
. . .
HIP_INIT_VOID();
hipError_t status = ihipMallocManaged(pointer, size, align);

Copy link
Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Moved device count check to __hipRegisterManagedVar().

return hipErrorNoDevice;
}

const amd::Device& dev = *ctx.devices()[0];

// Allocate SVM fine grain buffer with the forced host pointer, avoiding explicit memory
Expand Down
3 changes: 3 additions & 0 deletions hipamd/src/hip_platform.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -163,6 +163,9 @@ void __hipRegisterManagedVar(
} else {
ClPrint(amd::LOG_ERROR, amd::LOG_API, "Host Queue is NULL");
}
} else if (status == hipErrorNoDevice) {
*pointer = nullptr;
size = 0;
} else {
guarantee(false, "Error during allocation of managed memory!, error: %d", status);
}
Expand Down