intel/llvm
1
0
Fork 0
mirror of https://github.com/intel/llvm.git synced 2026-02-04 20:00:11 +08:00
Code Issues Packages Projects Releases Wiki Activity

[libc] Free the GPU memory allocated in the device loaders

Browse Source 
Summary:
This part was ignored and we just hoped that shutting down the runtime
freed these correctly. But it's best to be specific and free the memory
we've allocated.
This commit is contained in:
Joseph Huber
2023-04-03 11:54:48 -05:00
parent d891968f10
commit dfc162ad3f
2 changed files with 28 additions and 0 deletions
Download Patch File Download Diff File Expand all files Collapse all files

16
libc/utils/gpu/loader/amdgpu/Loader.cpp
View File

@@ -417,6 +417,22 @@ int load(int argc, char **argv, char **envp, void *image, size_t size) {
// Save the return value and perform basic clean-up.
int ret = *static_cast<int *>(host_ret);
// Free the memory allocated for the device.
if (hsa_status_t err = hsa_amd_memory_pool_free(args))
handle_error(err);
if (hsa_status_t err = hsa_amd_memory_pool_free(dev_argv))
handle_error(err);
if (hsa_status_t err = hsa_amd_memory_pool_free(dev_ret))
handle_error(err);
if (hsa_status_t err = hsa_amd_memory_pool_free(server_inbox))
handle_error(err);
if (hsa_status_t err = hsa_amd_memory_pool_free(server_outbox))
handle_error(err);
if (hsa_status_t err = hsa_amd_memory_pool_free(buffer))
handle_error(err);
if (hsa_status_t err = hsa_amd_memory_pool_free(host_ret))
handle_error(err);
if (hsa_status_t err = hsa_signal_destroy(memory_signal))
handle_error(err);

12
libc/utils/gpu/loader/nvptx/Loader.cpp
View File

@@ -176,6 +176,18 @@ int load(int argc, char **argv, char **envp, void *image, size_t size) {
if (CUresult err = cuStreamSynchronize(stream))
handle_error(err);
// Free the memory allocated for the device.
if (CUresult err = cuMemFree(dev_ret))
handle_error(err);
if (CUresult err = cuMemFreeHost(dev_argv))
handle_error(err);
if (CUresult err = cuMemFreeHost(server_inbox))
handle_error(err);
if (CUresult err = cuMemFreeHost(server_outbox))
handle_error(err);
if (CUresult err = cuMemFreeHost(buffer))
handle_error(err);
// Destroy the context and the loaded binary.
if (CUresult err = cuModuleUnload(binary))
handle_error(err);