More tests for memory,stream & error APIs

[ROCm/hip commit: 093279b10b]
2017-11-06 11:31:52 +05:30
@@ -0,0 +1,38 @@
+/*
+Copyright (c) 2015-2017 Advanced Micro Devices, Inc. All rights reserved.
+Permission is hereby granted, free of charge, to any person obtaining a copy
+of this software and associated documentation files (the "Software"), to deal
+in the Software without restriction, including without limitation the rights
+to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+copies of the Software, and to permit persons to whom the Software is
+furnished to do so, subject to the following conditions:
+The above copyright notice and this permission notice shall be included in
+all copies or substantial portions of the Software.
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANNTY OF ANY KIND, EXPRESS OR
+IMPLIED, INNCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+FITNNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL THE
+AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANNY CLAIM, DAMAGES OR OTHER
+LIABILITY, WHETHER INN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+OUT OF OR INN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
+THE SOFTWARE.
+*/
+
+/*
+ * Conformance test for checking functionality of
+ * hipError_t hipGetDevice(int *device);
+ */
+
+/* HIT_START
+ * BUILD: %t %s ../../test_common.cpp
+ * RUN: %t
+ * HIT_END
+ */
+
+#include "test_common.h"
+
+int main()
+{
+        hipSetDevice(-1);
+      if( hipPeekAtLastError() != hipSuccess)
+        passed();
+}
@@ -0,0 +1,77 @@
+/*
+Copyright (c) 2015-2017 Advanced Micro Devices, Inc. All rights reserved.
+Permission is hereby granted, free of charge, to any person obtaining a copy
+of this software and associated documentation files (the "Software"), to deal
+in the Software without restriction, including without limitation the rights
+to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+copies of the Software, and to permit persons to whom the Software is
+furnished to do so, subject to the following conditions:
+The above copyright notice and this permission notice shall be included in
+all copies or substantial portions of the Software.
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANNTY OF ANY KIND, EXPRESS OR
+IMPLIED, INNCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+FITNNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL THE
+AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANNY CLAIM, DAMAGES OR OTHER
+LIABILITY, WHETHER INN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+OUT OF OR INN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
+THE SOFTWARE.
+*/
+
+/*
+ * Conformance test for checking functionality of
+ * hipError_t hipMemcpyPeer(void* dst, int dstDeviceId, const void* src, int srcDeviceId, size_t sizeBytes);
+ */
+
+/* HIT_START
+ * BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvcc
+ * RUN: %t
+ * HIT_END
+ */
+
+#include "test_common.h"
+
+int main()
+{
+    hipDevice_t device;
+    size_t Nbytes = N*sizeof(int);
+    int numDevices = 0;
+    int *A_d, *B_d, *C_d, *X_d, *Y_d, *Z_d;
+    int *A_h, *B_h, *C_h ;
+
+    HIPCHECK(hipGetDeviceCount(&numDevices));
+     if(numDevices > 1)
+     {
+        HIPCHECK(hipSetDevice(0));
+        unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N);
+        HipTest::initArrays(&A_d, &B_d, &C_d, &A_h, &B_h, &C_h, N, false);
+        HIPCHECK(hipSetDevice(1));
+        HIPCHECK(hipMalloc(&X_d,Nbytes));
+        HIPCHECK(hipMalloc(&Y_d,Nbytes));
+        HIPCHECK(hipMalloc(&Z_d,Nbytes));
+
+        
+        HIPCHECK(hipSetDevice(0));
+        HIPCHECK ( hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice));
+        HIPCHECK ( hipMemcpy(B_d, B_h, Nbytes, hipMemcpyHostToDevice));
+        hipLaunchKernel(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock), 0, 0, A_d,B_d, C_d, N);
+        HIPCHECK ( hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost));
+        HIPCHECK (hipDeviceSynchronize());
+        HipTest::checkVectorADD(A_h, B_h, C_h, N);
+    
+         
+        HIPCHECK(hipSetDevice(1));
+        HIPCHECK(hipMemcpyDtoD(X_d, A_d,  Nbytes)); 
+        HIPCHECK(hipMemcpyDtoD(Y_d, B_d,  Nbytes));
+
+        hipLaunchKernel(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock), 0, 0, X_d,Y_d, Z_d, N);
+        HIPCHECK(hipMemcpyDtoH(C_h, Z_d, Nbytes));
+        HIPCHECK(hipDeviceSynchronize());
+        HipTest::checkVectorADD(A_h, B_h, C_h, N);
+
+        HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false);
+      }
+     
+        passed();
+     
+}
+
@@ -0,0 +1,81 @@
+/*
+Copyright (c) 2015-2017 Advanced Micro Devices, Inc. All rights reserved.
+Permission is hereby granted, free of charge, to any person obtaining a copy
+of this software and associated documentation files (the "Software"), to deal
+in the Software without restriction, including without limitation the rights
+to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+copies of the Software, and to permit persons to whom the Software is
+furnished to do so, subject to the following conditions:
+The above copyright notice and this permission notice shall be included in
+all copies or substantial portions of the Software.
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANNTY OF ANY KIND, EXPRESS OR
+IMPLIED, INNCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+FITNNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL THE
+AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANNY CLAIM, DAMAGES OR OTHER
+LIABILITY, WHETHER INN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+OUT OF OR INN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
+THE SOFTWARE.
+*/
+
+/*
+ * Conformance test for checking functionality of
+ * hipError_t hipMemcpyPeer(void* dst, int dstDeviceId, const void* src, int srcDeviceId, size_t sizeBytes);
+ */
+
+/* HIT_START
+ * BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvcc
+ * RUN: %t
+ * HIT_END
+ */
+
+#include "test_common.h"
+
+int main()
+{
+    hipDevice_t device;
+    size_t Nbytes = N*sizeof(int);
+    int numDevices = 0;
+    int *A_d, *B_d, *C_d, *X_d, *Y_d, *Z_d;
+    int *A_h, *B_h, *C_h ;
+    hipStream_t s;
+
+    HIPCHECK(hipGetDeviceCount(&numDevices));
+     if(numDevices > 1)
+     {
+        HIPCHECK(hipSetDevice(0));
+        unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N);
+        HipTest::initArrays(&A_d, &B_d, &C_d, &A_h, &B_h, &C_h, N, false);
+        HIPCHECK(hipSetDevice(1));
+        HIPCHECK(hipMalloc(&X_d,Nbytes));
+        HIPCHECK(hipMalloc(&Y_d,Nbytes));
+        HIPCHECK(hipMalloc(&Z_d,Nbytes));
+
+        
+        HIPCHECK(hipSetDevice(0));
+        HIPCHECK ( hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice));
+        HIPCHECK ( hipMemcpy(B_d, B_h, Nbytes, hipMemcpyHostToDevice));
+        hipLaunchKernel(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock), 0, 0, A_d,B_d, C_d, N);
+        HIPCHECK ( hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost));
+        HIPCHECK (hipDeviceSynchronize());
+        HipTest::checkVectorADD(A_h, B_h, C_h, N);
+    
+        HIPCHECK(hipStreamCreate(&s));    
+        HIPCHECK(hipSetDevice(1));
+        HIPCHECK(hipMemcpyDtoDAsync(X_d, A_d, Nbytes, s)); 
+        HIPCHECK(hipMemcpyDtoDAsync(Y_d, B_d, Nbytes, s));
+
+        hipLaunchKernel(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock), 0, 0, X_d,Y_d, Z_d, N);
+        HIPCHECK(hipMemcpyDtoHAsync(C_h, Z_d, Nbytes, s));
+        HIPCHECK(hipStreamSynchronize(s));
+        HIPCHECK(hipDeviceSynchronize());
+        
+        HipTest::checkVectorADD(A_h, B_h, C_h, N);
+
+        HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false);
+     }
+
+        passed();
+     
+
+}
+
@@ -0,0 +1,77 @@
+/*
+Copyright (c) 2015-2017 Advanced Micro Devices, Inc. All rights reserved.
+Permission is hereby granted, free of charge, to any person obtaining a copy
+of this software and associated documentation files (the "Software"), to deal
+in the Software without restriction, including without limitation the rights
+to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+copies of the Software, and to permit persons to whom the Software is
+furnished to do so, subject to the following conditions:
+The above copyright notice and this permission notice shall be included in
+all copies or substantial portions of the Software.
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANNTY OF ANY KIND, EXPRESS OR
+IMPLIED, INNCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+FITNNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL THE
+AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANNY CLAIM, DAMAGES OR OTHER
+LIABILITY, WHETHER INN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+OUT OF OR INN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
+THE SOFTWARE.
+*/
+
+/*
+ * Conformance test for checking functionality of
+ * hipError_t hipMemcpyPeer(void* dst, int dstDeviceId, const void* src, int srcDeviceId, size_t sizeBytes);
+ */
+
+/* HIT_START
+ * BUILD: %t %s ../../test_common.cpp
+ * RUN: %t
+ * HIT_END
+ */
+
+#include "test_common.h"
+
+int main()
+{
+    hipDevice_t device;
+    size_t Nbytes = N*sizeof(int);
+    int numDevices = 0;
+    int *A_d, *B_d, *C_d, *X_d, *Y_d, *Z_d;
+    int *A_h, *B_h, *C_h ;
+
+    HIPCHECK(hipGetDeviceCount(&numDevices));
+     if(numDevices > 1)
+     {
+        HIPCHECK(hipSetDevice(0));
+        unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N);
+        HipTest::initArrays(&A_d, &B_d, &C_d, &A_h, &B_h, &C_h, N, false);
+        HIPCHECK(hipSetDevice(1));
+        HIPCHECK(hipMalloc(&X_d,Nbytes));
+        HIPCHECK(hipMalloc(&Y_d,Nbytes));
+        HIPCHECK(hipMalloc(&Z_d,Nbytes));
+
+        
+        HIPCHECK(hipSetDevice(0));
+        HIPCHECK ( hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice));
+        HIPCHECK ( hipMemcpy(B_d, B_h, Nbytes, hipMemcpyHostToDevice));
+        hipLaunchKernel(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock), 0, 0, A_d,B_d, C_d, N);
+        HIPCHECK ( hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost));
+        HIPCHECK (hipDeviceSynchronize());
+        HipTest::checkVectorADD(A_h, B_h, C_h, N);
+    
+         
+        HIPCHECK(hipSetDevice(1));
+        hipMemcpyPeer(X_d, 1, A_d, 0, Nbytes); //this call is eqv to hipMemcpy(hipMemcpyD2D) which goes via stg bufs.
+        hipMemcpyPeer(Y_d, 1, B_d, 0, Nbytes);
+
+        hipLaunchKernel(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock), 0, 0, X_d,Y_d, Z_d, N);
+        HIPCHECK ( hipMemcpy(C_h, Z_d, Nbytes, hipMemcpyDeviceToHost));
+        HIPCHECK (hipDeviceSynchronize());
+        HipTest::checkVectorADD(A_h, B_h, C_h, N);
+     }
+        passed();
+
+
+     
+
+}
+
@@ -0,0 +1,79 @@
+/*
+Copyright (c) 2015-2017 Advanced Micro Devices, Inc. All rights reserved.
+Permission is hereby granted, free of charge, to any person obtaining a copy
+of this software and associated documentation files (the "Software"), to deal
+in the Software without restriction, including without limitation the rights
+to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+copies of the Software, and to permit persons to whom the Software is
+furnished to do so, subject to the following conditions:
+The above copyright notice and this permission notice shall be included in
+all copies or substantial portions of the Software.
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANNTY OF ANY KIND, EXPRESS OR
+IMPLIED, INNCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+FITNNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL THE
+AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANNY CLAIM, DAMAGES OR OTHER
+LIABILITY, WHETHER INN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+OUT OF OR INN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
+THE SOFTWARE.
+*/
+
+/*
+ * Conformance test for checking functionality of
+ * hipError_t hipMemcpyPeer(void* dst, int dstDeviceId, const void* src, int srcDeviceId, size_t sizeBytes);
+ */
+
+/* HIT_START
+ * BUILD: %t %s ../../test_common.cpp
+ * RUN: %t
+ * HIT_END
+ */
+
+#include "test_common.h"
+
+int main()
+{
+    hipDevice_t device;
+    size_t Nbytes = N*sizeof(int);
+    int numDevices = 0;
+    int *A_d, *B_d, *C_d, *X_d, *Y_d, *Z_d;
+    int *A_h, *B_h, *C_h ;
+    hipStream_t s;
+
+
+    HIPCHECK(hipGetDeviceCount(&numDevices));
+     if(numDevices > 1)
+     {
+        HIPCHECK(hipSetDevice(0));
+        unsigned blocks = HipTest::setNumBlocks(blocksPerCU, threadsPerBlock, N);
+        HipTest::initArrays(&A_d, &B_d, &C_d, &A_h, &B_h, &C_h, N, false);
+        HIPCHECK(hipSetDevice(1));
+        HIPCHECK(hipMalloc(&X_d,Nbytes));
+        HIPCHECK(hipMalloc(&Y_d,Nbytes));
+        HIPCHECK(hipMalloc(&Z_d,Nbytes));
+
+        
+        HIPCHECK(hipSetDevice(0));
+        HIPCHECK ( hipMemcpy(A_d, A_h, Nbytes, hipMemcpyHostToDevice));
+        HIPCHECK ( hipMemcpy(B_d, B_h, Nbytes, hipMemcpyHostToDevice));
+        hipLaunchKernel(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock), 0, 0, A_d,B_d, C_d, N);
+        HIPCHECK ( hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost));
+        HIPCHECK (hipDeviceSynchronize());
+        HipTest::checkVectorADD(A_h, B_h, C_h, N);
+    
+        HIPCHECK(hipStreamCreate(&s));  
+        HIPCHECK(hipSetDevice(1));
+        HIPCHECK(hipMemcpyPeerAsync(X_d, 1, A_d, 0, Nbytes, s));
+        HIPCHECK(hipMemcpyPeerAsync(Y_d, 1, B_d, 0, Nbytes, s));
+
+        hipLaunchKernel(HipTest::vectorADD, dim3(blocks), dim3(threadsPerBlock), 0, 0, X_d,Y_d, Z_d, N);
+        HIPCHECK ( hipMemcpy(C_h, Z_d, Nbytes, hipMemcpyDeviceToHost));
+        HIPCHECK (hipDeviceSynchronize());
+        HIPCHECK (hipStreamSynchronize(s));
+        HipTest::checkVectorADD(A_h, B_h, C_h, N);
+     }
+
+        passed();
+    
+
+}
+
@@ -0,0 +1,44 @@
+/*
+Copyright (c) 2015-2016 Advanced Micro Devices, Inc. All rights reserved.
+Permission is hereby granted, free of charge, to any person obtaining a copy
+of this software and associated documentation files (the "Software"), to deal
+in the Software without restriction, including without limitation the rights
+to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
+copies of the Software, and to permit persons to whom the Software is
+furnished to do so, subject to the following conditions:
+The above copyright notice and this permission notice shall be included in
+all copies or substantial portions of the Software.
+THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT.  IN NO EVENT SHALL THE
+AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
+THE SOFTWARE.
+*/
+
+/* HIT_START
+ * BUILD: %t %s ../../test_common.cpp EXCLUDE_HIP_PLATFORM nvcc
+ * RUN: %t
+ * HIT_END
+ */
+
+#include "test_common.h"
+
+
+int main(int argc, char *argv[])
+{
+    hipStream_t stream;
+    unsigned int flags;
+    HIPCHECK(hipStreamCreateWithFlags(&stream, hipStreamDefault));
+    HIPCHECK(hipStreamGetFlags(stream, &flags));
+    HIPASSERT(flags == 0); 
+    HIPCHECK(hipStreamDestroy(stream));
+
+    HIPCHECK(hipStreamCreateWithFlags(&stream, hipStreamNonBlocking));
+    HIPCHECK(hipStreamGetFlags(stream, &flags));
+    HIPASSERT(flags == 1); 
+    HIPCHECK(hipStreamDestroy(stream));
+
+    passed();
+}