AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

Merge branch 'amd-develop' into amd-master

Browse Source 
Change-Id: I0e856db61fa4a50e190bd1d4c464ceb4a709b550
This commit is contained in:
Maneesh Gupta
2017-02-23 11:19:06 +05:30
parent 008499801b b5e3f59377
commit 518967319c
71 changed files with 598 additions and 224 deletions
Download Patch File Download Diff File Expand all files Collapse all files
samples/0_Intro/square/square.hipref.cpp
+1 -1
View File
@@ -81,7 +81,7 @@ int main(int argc, char *argv[])
const unsigned threadsPerBlock = 256;
printf ("info: launch 'vector_square' kernel\n");
hipLaunchKernel(HIP_KERNEL_NAME(vector_square), dim3(blocks), dim3(threadsPerBlock), 0, 0, C_d, A_d, N);
hipLaunchKernel(vector_square, dim3(blocks), dim3(threadsPerBlock), 0, 0, C_d, A_d, N);
printf ("info: copy Device2Host\n");
CHECK ( hipMemcpy(C_h, C_d, Nbytes, hipMemcpyDeviceToHost));
samples/1_Utils/hipCommander/hipCommander.cpp
+1 -1
View File
@@ -23,7 +23,7 @@ bool g_printedTiming = false;
// Cmdline parms:
int p_device = 0;
const char* p_command = "H2D; NullKernel; D2H";
const char* p_command = "setstream(1); H2D; NullKernel; D2H;";
const char* p_file = nullptr;
unsigned p_verbose = 0x0;
unsigned p_db = 0x0;
samples/1_Utils/hipCommander/perf/scripts/latency_2_d2h_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; H2D; streamsync;D2H; H2D; streamsync; endloop(1);
loop(10); D2H; H2D; streamsync;D2H; H2D; streamsync; endloop(1);
loop(100); D2H; H2D; streamsync;D2H; H2D; streamsync; endloop(1);
loop(100); D2H; H2D; streamsync;D2H; H2D; streamsync; endloop(1);
loop(1000); D2H; H2D; streamsync;D2H; H2D; streamsync; endloop(1);
loop(1000); D2H; H2D; streamsync;D2H; H2D; streamsync; endloop(1);
loop(10000); D2H; H2D; streamsync;D2H; H2D; streamsync; endloop(1);
loop(10000); D2H; H2D; streamsync;D2H; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_d2h_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync; endloop(1);
loop(10); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync; endloop(1);
loop(100); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync; endloop(1);
loop(100); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync; endloop(1);
loop(1000); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync; endloop(1);
loop(1000); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync; endloop(1);
loop(10000); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync; endloop(1);
loop(10000); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_d2h_sync_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; endloop(1);
loop(10); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; endloop(1);
loop(100); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; endloop(1);
loop(100); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; endloop(1);
loop(1000); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; endloop(1);
loop(1000); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; endloop(1);
loop(10000); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; endloop(1);
loop(10000); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_d2h_sync_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; endloop(1);
loop(10); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; endloop(1);
loop(100); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; endloop(1);
loop(100); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_h2d_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; D2H; streamsync;H2D; D2H; streamsync; endloop(1);
loop(10); H2D; D2H; streamsync;H2D; D2H; streamsync; endloop(1);
loop(100); H2D; D2H; streamsync;H2D; D2H; streamsync; endloop(1);
loop(100); H2D; D2H; streamsync;H2D; D2H; streamsync; endloop(1);
loop(1000); H2D; D2H; streamsync;H2D; D2H; streamsync; endloop(1);
loop(1000); H2D; D2H; streamsync;H2D; D2H; streamsync; endloop(1);
loop(10000); H2D; D2H; streamsync;H2D; D2H; streamsync; endloop(1);
loop(10000); H2D; D2H; streamsync;H2D; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_h2d_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync; endloop(1);
loop(10); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync;endloop(1);
loop(100); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync;endloop(1);
loop(100); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_h2d_sync_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; endloop(1);
loop(10); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; endloop(1);
loop(100); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; endloop(1);
loop(100); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; endloop(1);
loop(1000); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; endloop(1);
loop(1000); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; endloop(1);
loop(10000); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; endloop(1);
loop(10000); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_h2d_sync_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(10); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_kernel_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; endloop(1);
loop(10); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; endloop(1);
loop(100); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; endloop(1);
loop(100); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; endloop(1);
loop(1000); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; endloop(1);
loop(1000); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; endloop(1);
loop(10000); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; endloop(1);
loop(10000); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_kernel_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; endloop(1);
loop(10); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; endloop(1);
loop(100); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; endloop(1);
loop(100); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; endloop(1);
loop(1000); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; endloop(1);
loop(1000); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; endloop(1);
loop(10000); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; endloop(1);
loop(10000); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_kernel_sync_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(100); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(100); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_kernel_sync_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(100); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(100); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2_sync.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); streamsync; streamsync; endloop(1);
loop(10); streamsync; streamsync; endloop(1);
loop(100); streamsync; streamsync; endloop(1);
loop(100); streamsync; streamsync; endloop(1);
loop(1000); streamsync; streamsync; endloop(1);
loop(1000); streamsync; streamsync; endloop(1);
loop(10000); streamsync; streamsync; endloop(1);
loop(10000); streamsync; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2d2h.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); D2H; streamsync; D2H; streamsync; endloop(1);
loop(10); D2H; streamsync; D2H; streamsync; endloop(1);
loop(100); D2H; streamsync; D2H; streamsync; endloop(1);
loop(100); D2H; streamsync; D2H; streamsync; endloop(1);
loop(1000); D2H;streamsync; D2H; streamsync; endloop(1);
loop(1000); D2H; streamsync; D2H; streamsync; endloop(1);
loop(1000); D2H; streamsync; D2H; streamsync; endloop(1);
loop(10000); D2H; streamsync; D2H; streamsync; endloop(1);
loop(10000); D2H; streamsync; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2d2h_wosync.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); D2H; D2H; streamsync; endloop(1);
loop(10); D2H; D2H; streamsync; endloop(1);
loop(100); D2H; D2H; streamsync; endloop(1);
loop(100); D2H; D2H; streamsync; endloop(1);
loop(1000); D2H; D2H; streamsync; endloop(1);
loop(1000); D2H; D2H; streamsync; endloop(1);
loop(1000); D2H; D2H; streamsync; endloop(1);
loop(10000); D2H; D2H; streamsync; endloop(1);
loop(10000); D2H; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2h2d.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); H2D; streamsync; H2D; streamsync; endloop(1);
loop(10); H2D; streamsync; H2D; streamsync; endloop(1);
loop(100); H2D; streamsync; H2D; streamsync; endloop(1);
loop(100); H2D; streamsync; H2D; streamsync; endloop(1);
loop(1000); H2D;streamsync; H2D; streamsync; endloop(1);
loop(1000); H2D; streamsync; H2D; streamsync; endloop(1);
loop(1000); H2D; streamsync; H2D; streamsync; endloop(1);
loop(10000); H2D; streamsync; H2D; streamsync; endloop(1);
loop(10000); H2D; streamsync; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2h2d_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel; streamsync;endloop(1);
loop(10); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel; streamsync;endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel; streamsync;endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel; streamsync;endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel; streamsync;endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel; streamsync;endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2h2d_kernel_wosync.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync; endloop(1);
loop(10); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync;endloop(1);
loop(100); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync;endloop(1);
loop(100); H2D; NullKernel; streamsync; H2D; NullKernel; streamsync;endloop(1);
loop(1000); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D ; NullKernel; streamsync;H2D; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2h2d_wosync.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); H2D; H2D; streamsync; endloop(1);
loop(10); H2D; H2D; streamsync; endloop(1);
loop(100); H2D; H2D; streamsync; endloop(1);
loop(100); H2D; H2D; streamsync; endloop(1);
loop(1000); H2D; H2D; streamsync; endloop(1);
loop(1000); H2D; H2D; streamsync; endloop(1);
loop(1000); H2D; H2D; streamsync; endloop(1);
loop(10000); H2D; H2D; streamsync; endloop(1);
loop(10000); H2D; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2kernels.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(10); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_2kernels_wosync.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); NullKernel; NullKernel; streamsync; endloop(1);
loop(10); NullKernel; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_d2h_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; H2D; streamsync;D2H; H2D; streamsync; D2H; H2D; streamsync; endloop(1);
loop(10); D2H; H2D; streamsync;D2H; H2D; streamsync; D2H; H2D; streamsync; endloop(1);
loop(100); D2H; H2D; streamsync;D2H; H2D; streamsync; D2H; H2D; streamsync; endloop(1);
loop(100); D2H; H2D; streamsync;D2H; H2D; streamsync; D2H; H2D; streamsync; endloop(1);
loop(1000); D2H; H2D; streamsync;D2H; H2D; streamsync; D2H; H2D; streamsync; endloop(1);
loop(1000); D2H; H2D; streamsync;D2H; H2D; streamsync; D2H; H2D; streamsync; endloop(1);
loop(10000); D2H; H2D; streamsync;D2H; H2D; streamsync; D2H; H2D; streamsync; endloop(1);
loop(10000); D2H; H2D; streamsync;D2H; H2D; streamsync; D2H; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_d2h_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync;streamsync; D2H; NullKernel; streamsync; endloop(1);
loop(10); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync;streamsync; D2H; NullKernel; streamsync; endloop(1);
loop(100); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync;streamsync; D2H; NullKernel; streamsync; endloop(1);
loop(100); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync;streamsync; D2H; NullKernel; streamsync; endloop(1);
loop(1000); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync;streamsync; D2H; NullKernel; streamsync; endloop(1);
loop(1000); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync;streamsync; D2H; NullKernel; streamsync; endloop(1);
loop(10000); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync;streamsync; D2H; NullKernel; streamsync; endloop(1);
loop(10000); D2H; NullKernel; streamsync;D2H; NullKernel; streamsync;streamsync; D2H; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_d2h_sync_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; D2H; streamsync; H2D;streamsync; endloop(1);
loop(10); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; D2H; streamsync; H2D;streamsync; endloop(1);
loop(100); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; D2H; streamsync; H2D;streamsync; endloop(1);
loop(100); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; D2H; streamsync; H2D;streamsync; endloop(1);
loop(1000); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; D2H; streamsync; H2D;streamsync; endloop(1);
loop(1000); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; D2H; streamsync; H2D;streamsync; endloop(1);
loop(10000); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; D2H; streamsync; H2D;streamsync; endloop(1);
loop(10000); D2H; streamsync; H2D; streamsync;D2H; streamsync; H2D; streamsync; D2H; streamsync; H2D;streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_d2h_sync_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; D2H; streamsync; NullKernel;streamsync; endloop(1);
loop(10); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; D2H; streamsync; NullKernel;streamsync; endloop(1);
loop(100); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; D2H; streamsync; NullKernel;streamsync; endloop(1);
loop(100); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; D2H; streamsync; NullKernel;streamsync; endloop(1);
loop(1000); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; D2H; streamsync; NullKernel;streamsync; endloop(1);
loop(1000); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; D2H; streamsync; NullKernel;streamsync; endloop(1);
loop(10000); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; D2H; streamsync; NullKernel;streamsync; endloop(1);
loop(10000); D2H; streamsync; NullKernel; streamsync;D2H; streamsync; NullKernel; streamsync; D2H; streamsync; NullKernel;streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_h2d_d2h.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); H2D; D2H; streamsync;H2D; D2H; streamsync; H2D; D2H; streamsync; endloop(1);
loop(10); H2D; D2H; streamsync;H2D; D2H; streamsync; H2D; D2H; streamsync; endloop(1);
loop(100); H2D; D2H; streamsync;H2D; D2H; streamsync; H2D; D2H; streamsync; endloop(1);
loop(100); H2D; D2H; streamsync;H2D; D2H; streamsync; H2D; D2H; streamsync; endloop(1);
loop(1000); H2D; D2H; streamsync;H2D; D2H; streamsync; H2D; D2H; streamsync; endloop(1);
loop(1000); H2D; D2H; streamsync;H2D; D2H; streamsync; H2D; D2H; streamsync; endloop(1);
loop(10000); H2D; D2H; streamsync;H2D; D2H; streamsync; H2D; D2H; streamsync; endloop(1);
loop(10000); H2D; D2H; streamsync;H2D; D2H; streamsync; H2D; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_h2d_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync;streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(10); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync;streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(100); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync;streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(100); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync;streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync;streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync;streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync;streamsync; H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D; NullKernel; streamsync;H2D; NullKernel; streamsync;streamsync; H2D; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_h2d_sync_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; H2D; streamsync; D2H;streamsync; endloop(1);
loop(10); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; H2D; streamsync; D2H;streamsync; endloop(1);
loop(100); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; H2D; streamsync; D2H;streamsync; endloop(1);
loop(100); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; H2D; streamsync; D2H;streamsync; endloop(1);
loop(1000); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; H2D; streamsync; D2H;streamsync; endloop(1);
loop(1000); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; H2D; streamsync; D2H;streamsync; endloop(1);
loop(10000); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; H2D; streamsync; D2H;streamsync; endloop(1);
loop(10000); H2D; streamsync; D2H; streamsync;H2D; streamsync; D2H; streamsync; H2D; streamsync; D2H;streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_h2d_sync_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel;streamsync; endloop(1);
loop(10); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel;streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel;streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel;streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel;streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel;streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel;streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync;H2D; streamsync; NullKernel; streamsync; H2D; streamsync; NullKernel;streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_kernel_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; NullKernel; D2H; streamsync;endloop(1);
loop(10); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; NullKernel; D2H; streamsync;endloop(1);
loop(100); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; NullKernel; D2H; streamsync;endloop(1);
loop(100); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; NullKernel; D2H; streamsync;endloop(1);
loop(1000); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; NullKernel; D2H; streamsync;endloop(1);
loop(1000); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; NullKernel; D2H; streamsync;endloop(1);
loop(10000); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; NullKernel; D2H; streamsync;endloop(1);
loop(10000); NullKernel; D2H; streamsync;NullKernel; D2H; streamsync; NullKernel; D2H; streamsync;endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_kernel_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; NullKernel; H2D; streamsync;endloop(1);
loop(10); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; NullKernel; H2D; streamsync;endloop(1);
loop(100); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; NullKernel; H2D; streamsync;endloop(1);
loop(100); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; NullKernel; H2D; streamsync;endloop(1);
loop(1000); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; NullKernel; H2D; streamsync;endloop(1);
loop(1000); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; NullKernel; H2D; streamsync;endloop(1);
loop(10000); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; NullKernel; H2D; streamsync;endloop(1);
loop(10000); NullKernel; H2D; streamsync;NullKernel; H2D; streamsync; NullKernel; H2D; streamsync;endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_kernel_sync_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(100); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(100); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync;NullKernel; streamsync; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_kernel_sync_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(100); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(100); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync;NullKernel; streamsync; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3_sync.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10);streamsync; streamsync; streamsync; endloop(1);
loop(10);streamsync; streamsync; streamsync; endloop(1);
loop(100);streamsync; streamsync; streamsync; endloop(1);
loop(100);streamsync; streamsync; streamsync; endloop(1);
loop(1000);streamsync; streamsync; streamsync; endloop(1);
loop(1000);streamsync; streamsync; streamsync; endloop(1);
loop(10000);streamsync; streamsync; streamsync; endloop(1);
loop(10000);streamsync; streamsync; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; streamsync; D2H; streamsync; D2H; streamsync; endloop(1);
loop(10); D2H; streamsync; D2H; streamsync; D2H; streamsync; endloop(1);
loop(100); D2H; streamsync; D2H; streamsync; D2H; streamsync; endloop(1);
loop(100); D2H; streamsync; D2H; streamsync; D2H; streamsync; endloop(1);
loop(1000); D2H;streamsync; D2H; streamsync; D2H; streamsync; endloop(1);
loop(1000); D2H; streamsync; D2H; streamsync; D2H; streamsync; endloop(1);
loop(10000); D2H; streamsync; D2H; streamsync; D2H; streamsync; endloop(1);
loop(10000); D2H; streamsync; D2H; streamsync; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3d2h_wosync.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; D2H; D2H; streamsync; endloop(1);
loop(10); D2H; D2H; D2H; streamsync; endloop(1);
loop(100); D2H; D2H; D2H; streamsync; endloop(1);
loop(100); D2H; D2H; D2H; streamsync; endloop(1);
loop(1000); D2H; D2H; D2H; streamsync; endloop(1);
loop(1000); D2H; D2H; D2H; streamsync; endloop(1);
loop(10000); D2H; D2H; D2H; streamsync; endloop(1);
loop(10000); D2H; D2H; D2H;streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3h2d.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); H2D; streamsync;H2D;streamsync; H2D; streamsync; endloop(1);
loop(10); H2D; streamsync;H2D;streamsync; H2D; streamsync; endloop(1);
loop(100); H2D; streamsync;H2D; streamsync;H2D; streamsync; endloop(1);
loop(100); H2D;streamsync; H2D; streamsync;H2D; streamsync; endloop(1);
loop(1000); H2D;streamsync; H2D;streamsync; H2D; streamsync; endloop(1);
loop(1000); H2D;streamsync; H2D; streamsync;H2D; streamsync; endloop(1);
loop(1000); H2D;streamsync; H2D; streamsync;H2D; streamsync; endloop(1);
loop(10000); H2D;streamsync; H2D; streamsync;H2D; streamsync; endloop(1);
loop(10000); H2D;streamsync; H2D;streamsync; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3h2d_wosync.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); H2D; H2D; H2D; streamsync; endloop(1);
loop(10); H2D; H2D; H2D; streamsync; endloop(1);
loop(100); H2D; H2D; H2D; streamsync; endloop(1);
loop(100); H2D; H2D; H2D; streamsync; endloop(1);
loop(1000); H2D; H2D; H2D; streamsync; endloop(1);
loop(1000); H2D; H2D; H2D; streamsync; endloop(1);
loop(1000); H2D; H2D; H2D; streamsync; endloop(1);
loop(10000); H2D; H2D; H2D; streamsync; endloop(1);
loop(10000); H2D; H2D; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3kernels.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(10); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; NullKernel; streamsync; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_3kernels_wosync.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(10); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_4kernels.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(10); NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
loop(100); NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_5kernels.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); NullKernel; NullKernel; NullKernel; NullKernel;NullKernel; streamsync; endloop(1);
loop(10); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(100); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
loop(100); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_6kernels.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); NullKernel; NullKernel; NullKernel; NullKernel;NullKernel;NullKernel; streamsync; endloop(1);
loop(10); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
loop(100); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;NullKernel;streamsync; endloop(1);
loop(100); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
loop(1000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
loop(10000); NullKernel; NullKernel; NullKernel; NullKernel; NullKernel; NullKernel;streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; streamsync; endloop(1);
loop(10); D2H; streamsync; endloop(1);
loop(100); D2H; streamsync; endloop(1);
loop(100); D2H; streamsync; endloop(1);
loop(1000); D2H; streamsync; endloop(1);
loop(1000); D2H; streamsync; endloop(1);
loop(10000); D2H; streamsync; endloop(1);
loop(10000); D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_d2h_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; H2D; streamsync; endloop(1);
loop(10); D2H; H2D; streamsync; endloop(1);
loop(100); D2H; H2D; streamsync; endloop(1);
loop(100); D2H; H2D; streamsync; endloop(1);
loop(1000); D2H; H2D; streamsync; endloop(1);
loop(1000); D2H; H2D; streamsync; endloop(1);
loop(10000); D2H; H2D; streamsync; endloop(1);
loop(10000); D2H; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_d2h_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; NullKernel; streamsync; endloop(1);
loop(10); D2H; NullKernel; streamsync; endloop(1);
loop(100); D2H; NullKernel; streamsync; endloop(1);
loop(100); D2H; NullKernel; streamsync; endloop(1);
loop(1000); D2H; NullKernel; streamsync; endloop(1);
loop(1000); D2H; NullKernel; streamsync; endloop(1);
loop(10000); D2H; NullKernel; streamsync; endloop(1);
loop(10000); D2H; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_d2h_sync_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; streamsync; H2D; streamsync; endloop(1);
loop(10); D2H; streamsync; H2D; streamsync; endloop(1);
loop(100); D2H; streamsync; H2D; streamsync; endloop(1);
loop(100); D2H; streamsync; H2D; streamsync; endloop(1);
loop(1000); D2H; streamsync; H2D; streamsync; endloop(1);
loop(1000); D2H; streamsync; H2D; streamsync; endloop(1);
loop(10000); D2H; streamsync; H2D; streamsync; endloop(1);
loop(10000); D2H; streamsync; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_d2h_sync_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); D2H; streamsync;NullKernel; streamsync; endloop(1);
loop(10); D2H; streamsync;NullKernel; streamsync; endloop(1);
loop(100); D2H; streamsync;NullKernel; streamsync; endloop(1);
loop(100); D2H; streamsync;NullKernel; streamsync; endloop(1);
loop(1000); D2H; streamsync;NullKernel; streamsync; endloop(1);
loop(1000); D2H; streamsync;NullKernel; streamsync; endloop(1);
loop(10000); D2H; streamsync;NullKernel; streamsync; endloop(1);
loop(10000); D2H; streamsync;NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); H2D; streamsync; endloop(1);
loop(10); H2D; streamsync; endloop(1);
loop(100); H2D; streamsync; endloop(1);
loop(100); H2D; streamsync; endloop(1);
loop(1000); H2D; streamsync; endloop(1);
loop(1000); H2D; streamsync; endloop(1);
loop(1000); H2D; streamsync; endloop(1);
loop(10000); H2D; streamsync; endloop(1);
loop(10000); H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_10.hcm
+2
View File
@@ -0,0 +1,2 @@
setstream(1);
loop(10); H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; D2H; streamsync; endloop(1);
loop(10); H2D; D2H; streamsync; endloop(1);
loop(100); H2D; D2H; streamsync; endloop(1);
loop(100); H2D; D2H; streamsync; endloop(1);
loop(1000); H2D; D2H; streamsync; endloop(1);
loop(1000); H2D; D2H; streamsync; endloop(1);
loop(10000); H2D; D2H; streamsync; endloop(1);
loop(10000); H2D; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_kernel.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; NullKernel; streamsync; endloop(1);
loop(10); H2D; NullKernel; streamsync; endloop(1);
loop(100); H2D; NullKernel; streamsync; endloop(1);
loop(100); H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_kernel_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10);H2D; streamsync; NullKernel; D2H; streamsync;endloop(1);
loop(10); H2D; streamsync; NullKernel; D2H; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; D2H; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; D2H; streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; D2H; streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; D2H; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; D2H; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_kernel_d2h_wosync.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10);H2D; NullKernel; D2H; streamsync;endloop(1);
loop(10); H2D; NullKernel; D2H; streamsync; endloop(1);
loop(100); H2D; NullKernel; D2H; streamsync; endloop(1);
loop(100); H2D; NullKernel; D2H; streamsync; endloop(1);
loop(1000); H2D; NullKernel; D2H; streamsync; endloop(1);
loop(1000); H2D; NullKernel; D2H; streamsync; endloop(1);
loop(10000); H2D; NullKernel; D2H; streamsync; endloop(1);
loop(10000); H2D; NullKernel; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_kernel_wosync.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; NullKernel; streamsync; endloop(1);
loop(10); H2D; NullKernel; streamsync; endloop(1);
loop(100); H2D; NullKernel; streamsync; endloop(1);
loop(100); H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync; endloop(1);
loop(1000); H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D; NullKernel; streamsync; endloop(1);
loop(10000); H2D ; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_sync_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; streamsync; D2H; streamsync; endloop(1);
loop(10); H2D; streamsync; D2H; streamsync; endloop(1);
loop(100); H2D; streamsync; D2H; streamsync; endloop(1);
loop(100); H2D; streamsync; D2H; streamsync; endloop(1);
loop(1000); H2D; streamsync; D2H; streamsync; endloop(1);
loop(1000); H2D; streamsync; D2H; streamsync; endloop(1);
loop(10000); H2D; streamsync; D2H; streamsync; endloop(1);
loop(10000); H2D; streamsync; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_sync_kernel_sync.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(10); H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_h2d_sync_kernel_sync_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10);H2D; streamsync; NullKernel;streamsync; D2H; streamsync;endloop(1);
loop(10); H2D; streamsync; NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(100); H2D; streamsync; NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(1000); H2D; streamsync; NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10000); H2D; streamsync; NullKernel; streamsync; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_kernel.hcm
+10
View File
@@ -0,0 +1,10 @@
setstream(1);
loop(10); NullKernel; streamsync; endloop(1);
loop(10); NullKernel; streamsync; endloop(1);
loop(100); NullKernel; streamsync; endloop(1);
loop(100); NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_kernel_barrier.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; streamsync; streamsync; endloop(1);
loop(10); NullKernel; streamsync; streamsync; endloop(1);
loop(100); NullKernel; streamsync; streamsync; endloop(1);
loop(100); NullKernel; streamsync; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_kernel_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; D2H; streamsync; endloop(1);
loop(10); NullKernel; D2H; streamsync; endloop(1);
loop(100); NullKernel; D2H; streamsync; endloop(1);
loop(100); NullKernel; D2H; streamsync; endloop(1);
loop(1000); NullKernel; D2H; streamsync; endloop(1);
loop(1000); NullKernel; D2H; streamsync; endloop(1);
loop(10000); NullKernel; D2H; streamsync; endloop(1);
loop(10000); NullKernel; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_kernel_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; H2D; streamsync; endloop(1);
loop(10); NullKernel; H2D; streamsync; endloop(1);
loop(100); NullKernel; H2D; streamsync; endloop(1);
loop(100); NullKernel; H2D; streamsync; endloop(1);
loop(1000); NullKernel; H2D; streamsync; endloop(1);
loop(1000); NullKernel; H2D; streamsync; endloop(1);
loop(10000); NullKernel; H2D; streamsync; endloop(1);
loop(10000); NullKernel; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_kernel_sync_d2h.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10); NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(100); NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(100); NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; D2H; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; D2H; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_kernel_sync_h2d.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10); NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(100); NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(100); NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(1000); NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; H2D; streamsync; endloop(1);
loop(10000); NullKernel; streamsync; H2D; streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_streamcreate.hcm
+2
View File
@@ -0,0 +1,2 @@
setstream(1);
loop(10);setstream(1);setstream(2);setstream(3);setstream(4);setstream(5);streamsync; endloop(1);
samples/1_Utils/hipCommander/perf/scripts/latency_sync.hcm
+9
View File
@@ -0,0 +1,9 @@
setstream(1);
loop(10); streamsync; endloop(1);
loop(10); streamsync; endloop(1);
loop(100); streamsync; endloop(1);
loop(100); streamsync; endloop(1);
loop(1000); streamsync; endloop(1);
loop(1000); streamsync; endloop(1);
loop(10000); streamsync; endloop(1);
loop(10000); streamsync; endloop(1);
samples/7_Advanced/hipblas_saxpy/Makefile
-34
View File
@@ -1,34 +0,0 @@
HIP_PATH?= $(wildcard /opt/rocm/hip)
ifeq (,$(HIP_PATH))
HIP_PATH=../../..
endif
HIPCC=$(HIP_PATH)/bin/hipcc
HIPCC_FLAGS += -std=c++11
HIP_PLATFORM=$(shell $(HIP_PATH)/bin/hipconfig --platform)
ifeq (${HIP_PLATFORM}, nvcc)
LIBS = -lcublas
endif
ifeq (${HIP_PLATFORM}, hcc)
HCBLAS_ROOT?= $(wildcard /opt/rocm/hcblas)
HIPCC_FLAGS += -stdlib=libc++ -I$(HCBLAS_ROOT)/include
LIBS = -L$(HCBLAS_ROOT)/lib -lhipblas -rpath $(HIP_PATH)/lib
endif
all: saxpy.hipblas.out
saxpy.cublas.out : saxpy.cublas.cpp
nvcc -std=c++11 -I$(CUDA_HOME)/include saxpy.cublas.cpp -o $@ -L$(CUDA_HOME)/lib64 -lcublas
# $HIPBLAS_ROOT/bin/hipifyblas ./saxpy.cublas.cpp > ./saxpy.hipblas.cpp
# Then review & finish port in saxpy.hipblas.cpp
saxpy.hipblasref.o: saxpy.hipblasref.cpp
$(HIPCC) $(HIPCC_FLAGS) -c $< -o $@
saxpy.hipblas.out: saxpy.hipblasref.o
$(HIPCC) $< -o $@ $(LIBS)
clean:
rm -f *.o *.out
samples/7_Advanced/hipblas_saxpy/saxpy.cublas.cpp
-94
View File
@@ -1,94 +0,0 @@
#include <random>
#include <algorithm>
#include <iostream>
#include <cmath>
// header file for the GPU API
#include <cuda_runtime.h>
#include <cublas_v2.h>
#define N (1024 * 500)
#define CHECK(cmd) \
{\
cudaError_t error = cmd; \
if (error != cudaSuccess) { \
fprintf(stderr, "error: '%s'(%d) at %s:%d\n", cudaGetErrorString(error), error,__FILE__, __LINE__); \
exit(EXIT_FAILURE);\
}\
}
#define CHECK_BLAS(cmd) \
{\
cublasStatus_t error = cmd;\
if (error != CUBLAS_STATUS_SUCCESS) { \
fprintf(stderr, "error: (%d) at %s:%d\n", error,__FILE__, __LINE__); \
exit(EXIT_FAILURE);\
}\
}
int main() {
const float a = 100.0f;
float x[N];
float y[N], y_cpu_res[N], y_gpu_res[N];
// initialize the input data
std::default_random_engine random_gen;
std::uniform_real_distribution<float> distribution(-N, N);
std::generate_n(x, N, [&]() { return distribution(random_gen); });
std::generate_n(y, N, [&]() { return distribution(random_gen); });
std::copy_n(y, N, y_cpu_res);
// Explicit GPU code:
size_t Nbytes = N*sizeof(float);
float *x_gpu, *y_gpu;
cublasHandle_t handle;
cudaDeviceProp props;
CHECK(cudaGetDeviceProperties(&props, 0/*deviceID*/));
printf ("info: running on device %s\n", props.name);
printf ("info: allocate host mem (%6.2f MB)\n", 2*Nbytes/1024.0/1024.0);
printf ("info: allocate device mem (%6.2f MB)\n", 2*Nbytes/1024.0/1024.0);
CHECK(cudaMalloc(&x_gpu, Nbytes));
CHECK(cudaMalloc(&y_gpu, Nbytes));
// Initialize the blas library
CHECK_BLAS ( cublasCreate(&handle));
// copy n elements from a vector in host memory space to a vector in GPU memory space
printf ("info: copy Host2Device\n");
CHECK_BLAS ( cublasSetVector(N, sizeof(*x), x, 1, x_gpu, 1));
CHECK_BLAS ( cublasSetVector(N, sizeof(*y), y, 1, y_gpu, 1));
printf ("info: launch 'saxpy' kernel\n");
CHECK_BLAS ( cublasSaxpy(handle, N, &a, x_gpu, 1, y_gpu, 1));
cudaDeviceSynchronize();
printf ("info: copy Device2Host\n");
CHECK_BLAS ( cublasGetVector(N, sizeof(*y_gpu_res), y_gpu, 1, y_gpu_res, 1));
// CPU implementation of saxpy
for (int i = 0; i < N; i++) {
y_cpu_res[i] = a * x[i] + y[i];
}
// verify the results
int errors = 0;
for (int i = 0; i < N; i++) {
if (fabs(y_cpu_res[i] - y_gpu_res[i]) > fabs(y_cpu_res[i] * 0.0001f))
errors++;
}
std::cout << errors << " errors" << std::endl;
CHECK( cudaFree(x_gpu));
CHECK( cudaFree(y_gpu));
CHECK_BLAS( cublasDestroy(handle));
return errors;
}
samples/7_Advanced/hipblas_saxpy/saxpy.hipblasref.cpp
-94
View File
@@ -1,94 +0,0 @@
#include <random>
#include <algorithm>
#include <iostream>
#include <cmath>
// header file for the GPU API
#include "hip/hip_runtime.h"
#include <hipblas.h>
#define N (1024 * 500)
#define CHECK(cmd) \
{\
hipError_t error = cmd; \
if (error != hipSuccess) { \
fprintf(stderr, "error: '%s'(%d) at %s:%d\n", hipGetErrorString(error), error,__FILE__, __LINE__); \
exit(EXIT_FAILURE);\
}\
}
#define CHECK_BLAS(cmd) \
{\
hipblasStatus_t error = cmd;\
if (error != HIPBLAS_STATUS_SUCCESS) { \
fprintf(stderr, "error: (%d) at %s:%d\n", error,__FILE__, __LINE__); \
exit(EXIT_FAILURE);\
}\
}
int main() {
const float a = 100.0f;
float x[N];
float y[N], y_cpu_res[N], y_gpu_res[N];
// initialize the input data
std::default_random_engine random_gen;
std::uniform_real_distribution<float> distribution(-N, N);
std::generate_n(x, N, [&]() { return distribution(random_gen); });
std::generate_n(y, N, [&]() { return distribution(random_gen); });
std::copy_n(y, N, y_cpu_res);
// Explicit GPU code:
size_t Nbytes = N*sizeof(float);
float *x_gpu, *y_gpu;
hipblasHandle_t handle;
hipDeviceProp_t props;
CHECK(hipGetDeviceProperties(&props, 0/*deviceID*/));
printf ("info: running on device %s\n", props.name);
printf ("info: allocate host mem (%6.2f MB)\n", 2*Nbytes/1024.0/1024.0);
printf ("info: allocate device mem (%6.2f MB)\n", 2*Nbytes/1024.0/1024.0);
CHECK(hipMalloc(&x_gpu, Nbytes));
CHECK(hipMalloc(&y_gpu, Nbytes));
// Initialize the blas library
CHECK_BLAS ( hipblasCreate(&handle));
// copy n elements from a vector in host memory space to a vector in GPU memory space
printf ("info: copy Host2Device\n");
CHECK_BLAS ( hipblasSetVector(N, sizeof(*x), x, 1, x_gpu, 1));
CHECK_BLAS ( hipblasSetVector(N, sizeof(*y), y, 1, y_gpu, 1));
printf ("info: launch 'saxpy' kernel\n");
CHECK_BLAS ( hipblasSaxpy(handle, N, &a, x_gpu, 1, y_gpu, 1));
hipDeviceSynchronize();
printf ("info: copy Device2Host\n");
CHECK_BLAS ( hipblasGetVector(N, sizeof(*y_gpu_res), y_gpu, 1, y_gpu_res, 1));
// CPU implementation of saxpy
for (int i = 0; i < N; i++) {
y_cpu_res[i] = a * x[i] + y[i];
}
// verify the results
int errors = 0;
for (int i = 0; i < N; i++) {
if (fabs(y_cpu_res[i] - y_gpu_res[i]) > fabs(y_cpu_res[i] * 0.0001f))
errors++;
}
std::cout << errors << " errors" << std::endl;
CHECK( hipFree(x_gpu));
CHECK( hipFree(y_gpu));
CHECK_BLAS( hipblasDestroy(handle));
return errors;
}