Make hipMemset*Async* tests pass in static lib

3 hipMemset*Async* direct tests will crash in static lib because
some hsa objects are not initialized when hip apis call them.
The fix will make the hip apis called in main() so that hsa objects
have been initialized at that stage.

Change-Id: I41eb29b8c2588acac0cd3ae6d96b14fb1abd235a


[ROCm/clr commit: 272f89a966]
Bu işleme şunda yer alıyor:
Tao Sang
2020-11-02 09:39:56 -05:00
işlemeyi yapan: Tao Sang
ebeveyn 69b62789fb
işleme e6a4e455e5
3 değiştirilmiş dosya ile 14 ekleme ve 7 silme
@@ -24,7 +24,7 @@
// //
/* HIT_START /* HIT_START
* BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvidia EXCLUDE_HIP_LIB_TYPE static * BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvidia
* TEST: %t * TEST: %t
* HIT_END * HIT_END
*/ */
@@ -35,7 +35,7 @@
#define NUM_H 256 #define NUM_H 256
#define NUM_W 256 #define NUM_W 256
unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N); unsigned blocks = 0;
hipStream_t stream; hipStream_t stream;
bool testResult = true; bool testResult = true;
@@ -150,6 +150,9 @@ bool testhipMemset2DAsyncMultiThread() {
int main() { int main() {
bool testResult = true; bool testResult = true;
blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N);
printf("blocks: %u\n", blocks);
testResult &= testhipMemset2DAsyncWithKernel(); testResult &= testhipMemset2DAsyncWithKernel();
if (testResult) { if (testResult) {
printf("Kernel and hipMemset2DAsync executed in correct order!\n"); printf("Kernel and hipMemset2DAsync executed in correct order!\n");
+5 -2
Dosyayı Görüntüle
@@ -23,7 +23,7 @@
*/ */
/* HIT_START /* HIT_START
* BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvidia EXCLUDE_HIP_LIB_TYPE static * BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvidia
* TEST: %t * TEST: %t
* HIT_END * HIT_END
*/ */
@@ -32,7 +32,7 @@
#define ITER 10 #define ITER 10
#define N 1024 * 1024 #define N 1024 * 1024
unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N); unsigned blocks = 0;
template <typename T> template <typename T>
__global__ void vector_square(T* B_d, T* C_d, size_t M) { __global__ void vector_square(T* B_d, T* C_d, size_t M) {
@@ -153,6 +153,9 @@ bool testhipMemsetD8AsyncWithKernel() {
int main() { int main() {
bool testResult = true; bool testResult = true;
int numDevices = 0; int numDevices = 0;
blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N);
printf("blocks: %u\n", blocks);
HIPCHECK(hipGetDeviceCount(&numDevices)); HIPCHECK(hipGetDeviceCount(&numDevices));
printf("total number of gpus in the system: %d\n", numDevices); printf("total number of gpus in the system: %d\n", numDevices);
+4 -3
Dosyayı Görüntüle
@@ -22,7 +22,7 @@
*/ */
/* HIT_START /* HIT_START
* BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvidia EXCLUDE_HIP_LIB_TYPE static * BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvidia
* TEST: %t * TEST: %t
* HIT_END * HIT_END
*/ */
@@ -31,8 +31,6 @@
#define NUM_THREADS 50 #define NUM_THREADS 50
#define ITER 50 #define ITER 50
unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N);
template <typename T> template <typename T>
class MemSetTest { class MemSetTest {
public: public:
@@ -219,6 +217,9 @@ int main() {
printf("Queueing up hipMemSetAsync jobs on multiple threads" printf("Queueing up hipMemSetAsync jobs on multiple threads"
"and checking results\n"); "and checking results\n");
unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N);
printf("blocks: %u\n", blocks);
testResult &= testhipMemsetAsyncWithMultiThread(); testResult &= testhipMemsetAsyncWithMultiThread();
if (!(testResult)) { if (!(testResult)) {
printf("Thread execution did not complete for hipMemsetAsync\n"); printf("Thread execution did not complete for hipMemsetAsync\n");