AI/rocm-systems
2
0
Garpu 0
Kode Masalah Tarik Permintaan Actions Packages Projects Rilis Wiki Kegiatan

test update

Telusuri Sumber
This commit is contained in:
Ben Sander
2016-04-17 07:54:39 -05:00
orang tua f9a31e28ad
melakukan 3301deb07c
2 mengubah file dengan 46 tambahan dan 26 penghapusan
Download Patch File Download Diff File Expand all files Collapse all files
hipamd/CMakeLists.txt
+1
Melihat File
@@ -92,6 +92,7 @@ if(HIP_PLATFORM STREQUAL "hcc")
set(CMAKE_CXX_COMPILER "${HCC_HOME}/bin/hcc")
set(CMAKE_C_COMPILER "${HCC_HOME}/bin/hcc")
# Set HIP_HCC so we know this is HIP compile, some files are shared with HCC (staging_buffer).
set(CMAKE_CXX_FLAGS " -hc -I${HCC_HOME}/include -I${HSA_PATH}/include -stdlib=libc++ -DHIP_HCC")
set(CMAKE_C_FLAGS " -hc -I${HCC_HOME}/include -I${HSA_PATH}/include -stdlib=libc++ -DHIP_HCC")
hipamd/tests/src/hipPeerToPeer_simple.cpp
+45 -26
Melihat File
@@ -99,44 +99,61 @@ void enablePeerFirst()
assert(canAccessPeer);
HIPCHECK(hipSetDevice(g_peerDevice));
HIPCHECK(hipDeviceEnablePeerAccess(g_currentDevice, 0));
}
HIPCHECK(hipDeviceEnablePeerAccess(g_currentDevice, 0));
}
size_t Nbytes = N*sizeof(char);
size_t Nbytes = N*sizeof(char);
char *A_d0, *A_d1;
char *A_h;
char *A_d0, *A_d1;
char *A_h;
A_h = (char*)malloc(Nbytes);
A_h = (char*)malloc(Nbytes);
// allocate and initialize memory on device0
// allocate and initialize memory on device0
HIPCHECK (hipSetDevice(g_currentDevice));
HIPCHECK (hipMalloc(&A_d0, Nbytes) );
HIPCHECK (hipMemset(A_d0, memsetval, Nbytes) );
// TODO - remove me:
HIPCHECK (hipDeviceSynchronize());
// allocate and initialize memory on peer device
HIPCHECK (hipSetDevice(g_peerDevice));
HIPCHECK (hipMalloc(&A_d1, Nbytes) );
HIPCHECK (hipMemset(A_d1, 0x13, Nbytes) );
// TODO - remove me:
HIPCHECK (hipDeviceSynchronize());
// Device0 push to device1, using P2P:
HIPCHECK (hipSetDevice(p_memcpyWithPeer ? g_peerDevice : g_currentDevice));
HIPCHECK (hipMemcpy(A_d1, A_d0, Nbytes, hipMemcpyDefault)); // This is P2P copy.
// TODO - remove me:
if (1) {
HIPCHECK (hipSetDevice(g_currentDevice));
HIPCHECK (hipMalloc(&A_d0, Nbytes) );
HIPCHECK (hipMemset(A_d0, memsetval, Nbytes) );
// allocate and initialize memory on peer device
HIPCHECK (hipDeviceSynchronize());
HIPCHECK (hipSetDevice(g_peerDevice));
HIPCHECK (hipMalloc(&A_d1, Nbytes) );
HIPCHECK (hipMemset(A_d1, 0x13, Nbytes) );
HIPCHECK (hipDeviceSynchronize());
}
// Copy data back to host:
HIPCHECK (hipSetDevice(g_peerDevice));
HIPCHECK (hipMemcpy(A_h, A_d1, Nbytes, hipMemcpyDeviceToHost));
// TODO - remove me:
HIPCHECK (hipDeviceSynchronize());
// Device0 push to device1, using P2P:
HIPCHECK (hipSetDevice(p_memcpyWithPeer ? g_peerDevice : g_currentDevice));
HIPCHECK (hipMemcpy(A_d1, A_d0, Nbytes, hipMemcpyDefault));
// Copy data back to host:
HIPCHECK (hipSetDevice(g_peerDevice));
HIPCHECK (hipMemcpy(A_h, A_d1, Nbytes, hipMemcpyDeviceToHost));
// Check host data:
for (int i=0; i<N; i++) {
if (A_h[i] != memsetval) {
failed("mismatch at index:%d computed:0x%02x, golden memsetval:0x%02x\n", i, (int)A_h[i], (int)memsetval);
}
// Check host data:
for (int i=0; i<N; i++) {
if (A_h[i] != memsetval) {
failed("mismatch at index:%d computed:0x%02x, golden memsetval:0x%02x\n", i, (int)A_h[i], (int)memsetval);
}
}
printf ("==done: %s\n\n", __func__);
}
//---
// Test which allocated memory first, then enables peer2peer.
@@ -199,6 +216,7 @@ void allocMemoryFirst()
failed("mismatch at index:%d computed:0x%02x, golden memsetval:0x%02x\n", i, (int)A_h[i], (int)memsetval);
}
}
printf ("==done: %s\n\n", __func__);
}
void simpleNegative()
@@ -233,6 +251,7 @@ void simpleNegative()
// More tests here:
printf ("==done: %s\n\n", __func__);
}