SWDEV-294596 - Fix module test failure on NV

1.Fix hipModuleNegative failure on all NV GPUs
a.Add signal handler for signal sent by cuda functions.
b.Make hipModuleGetGlobal match cuModuleGetGlobal behavour.
That is, if one of the first two parameters is nullptr, ignore it.

2.Fix hipModuleLoadDataMultThreaded failure on NV RTX5000
Improve lamda function.

Change-Id: I3fe6dbc35a7a14aa9119df197b7885df83d28047


[ROCm/hip commit: ae30c5cd6b]
This commit is contained in:
Tao Sang
2021-07-22 00:06:45 -04:00
committed by Tao Sang
orang tua d3fbcd81ee
melakukan 927bc28565
2 mengubah file dengan 51 tambahan dan 10 penghapusan
@@ -18,7 +18,7 @@ THE SOFTWARE.
*/
/* HIT_START
* BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11 EXCLUDE_HIP_PLATFORM nvidia
* BUILD: %t %s ../../test_common.cpp NVCC_OPTIONS -std=c++11
* TEST: %t
* HIT_END
*/
@@ -90,6 +90,8 @@ void run(const std::vector<char>& buffer) {
HIP_LAUNCH_PARAM_END};
HIPCHECK(hipModuleLaunchKernel(Function, 1, 1, 1, LEN, 1, 1, 0, stream, NULL, (void**)&config));
HIPCHECK(hipStreamSynchronize(stream));
HIPCHECK(hipStreamDestroy(stream));
HIPCHECK(hipModuleUnload(Module));
@@ -122,7 +124,7 @@ struct joinable_thread : std::thread {
void run_multi_threads(uint32_t n, const std::vector<char>& buffer) {
std::vector<joinable_thread> threads;
for (uint32_t i = 0; i < n; i++) {
threads.emplace_back(std::thread{[&, buffer] {
threads.emplace_back(std::thread{[&] {
run(buffer);
}});
}