diff --git a/projects/hip-tests/catch/unit/module/hipModuleGetFunction.cc b/projects/hip-tests/catch/unit/module/hipModuleGetFunction.cc index 676a61c33d..c4decba9d1 100644 --- a/projects/hip-tests/catch/unit/module/hipModuleGetFunction.cc +++ b/projects/hip-tests/catch/unit/module/hipModuleGetFunction.cc @@ -71,4 +71,21 @@ TEST_CASE("Unit_hipModuleGetFunction_Negative_Parameters") { SECTION("kname == __device__ kernel") { HIP_CHECK_ERROR(hipModuleGetFunction(&kernel, GetModule(), "DeviceKernel"), hipErrorNotFound); } -} \ No newline at end of file +} + +// Test description: Loading kernel function from different device than the one on which the module +// is loaded +TEST_CASE("Unit_hipModuleGetFunction_DiffDevice") { + int numDevices = 0; + HIP_CHECK(hipGetDeviceCount(&numDevices)); + if (numDevices < 2) { + SUCCEED("skipped the testcase as no of devices is less than 2"); + return; + } + + hipFunction_t kernel = nullptr; + auto module = GetModule(); + HIP_CHECK(hipSetDevice(1)); + HIP_CHECK(hipModuleGetFunction(&kernel, module, "GlobalKernel")); + REQUIRE(kernel != nullptr); +} diff --git a/projects/hip-tests/catch/unit/module/hipModuleGetGlobal.cc b/projects/hip-tests/catch/unit/module/hipModuleGetGlobal.cc index 32f46f0caf..a4a5423d14 100644 --- a/projects/hip-tests/catch/unit/module/hipModuleGetGlobal.cc +++ b/projects/hip-tests/catch/unit/module/hipModuleGetGlobal.cc @@ -142,4 +142,22 @@ TEST_CASE("Unit_hipModuleGetGlobal_Negative_Name_Is_Empty_String") { TEST_CASE("Unit_hipModuleGetGlobal_Negative_Dptr_And_Bytes_Are_Nullptr") { hipModule_t module = GetModule(); HIP_CHECK_ERROR(hipModuleGetGlobal(nullptr, nullptr, module, "int_var"), hipErrorInvalidValue); -} \ No newline at end of file +} + +// Test description: Loading device ptr from different device than the one on which the module +// is loaded +TEST_CASE("Unit_hipModuleGetGlobal_DiffDevice") { + int numDevices = 0; + HIP_CHECK(hipGetDeviceCount(&numDevices)); + if (numDevices < 2) { + SUCCEED("skipped the testcase as no of devices is less than 2"); + return; + } + auto module = GetModule(); + HIP_CHECK(hipSetDevice(1)); + hipDeviceptr_t global; + size_t global_size = 0; + HIP_CHECK(hipModuleGetGlobal(&global, &global_size, module, "int_var")); + REQUIRE(global != 0); + REQUIRE(sizeof(int) == global_size); +}