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

Fixing Readme & File structure & Typos

Browse Source 
Change-Id: I354b4e0e0448c3e8555d0b766b2c410c8049c2ff


[ROCm/rocprofiler commit: f308e36798]
This commit is contained in:
Ammar ELWazir authored and Ammar Elwazir committed 2023-05-31 18:00:16 -04:00
1 parent a305118408
commit f033633c04
12 files changed
+170 -450
Download Patch File Download Diff File Expand all files Collapse all files

No files matched your search

projects/rocprofiler/tests/featuretests/profiler/apps/hello_world_hip.cpp
-11
View File
@@ -42,16 +42,11 @@ __global__ void helloworld(char* in, char* out) {
int main(int argc, char* argv[]) {
hipDeviceProp_t devProp;
HIP_RC(hipGetDeviceProperties(&devProp, 0));
std::cout << " System minor " << devProp.minor << std::endl;
std::cout << " System major " << devProp.major << std::endl;
std::cout << " agent prop name " << devProp.name << std::endl;
/* Initial input,output for the host and create memory objects for the
* kernel*/
const char* input = "GdkknVnqkc";
size_t strlength = strlen(input);
std::cout << "input string:" << std::endl;
std::cout << input << std::endl;
char* output = reinterpret_cast<char*>(malloc(strlength + 1));
char* inputBuffer;
@@ -68,13 +63,7 @@ int main(int argc, char* argv[]) {
HIP_RC(hipFree(inputBuffer));
HIP_RC(hipFree(outputBuffer));
output[strlength] = '\0'; // Add the terminal character to the end of output.
std::cout << "\noutput string:" << std::endl;
std::cout << output << std::endl;
free(output);
std::cout << "Passed!\n";
return SUCCESS;
}
projects/rocprofiler/tests/featuretests/profiler/apps/vector_add_hip.cpp
-7
View File
@@ -60,11 +60,6 @@ int main() {
hipDeviceProp_t devProp;
HIP_RC(hipGetDeviceProperties(&devProp, 0));
std::cout << " System minor " << devProp.minor << std::endl;
std::cout << " System major " << devProp.major << std::endl;
std::cout << " agent prop name " << devProp.name << std::endl;
std::cout << "hip Device prop succeeded " << std::endl;
int i;
int errors;
@@ -102,8 +97,6 @@ int main() {
}
if (errors != 0) {
printf("FAILED: %d errors\n", errors);
} else {
printf("PASSED!\n");
}
HIP_RC(hipFree(deviceA));
projects/rocprofiler/tests/featuretests/profiler/profiler_gtest.cpp
-1
View File
@@ -821,7 +821,6 @@ void __attribute__((constructor)) globalsetting() {
std::string app_path = GetRunningPath(running_path);
std::stringstream gfx_path;
gfx_path << app_path << metrics_path;
std::cout << gfx_path.str() << std::endl;
setenv("ROCPROFILER_METRICS_PATH", gfx_path.str().c_str(), true);
}
projects/rocprofiler/tests/featuretests/utils/test_utils.cpp
-3
View File
@@ -41,9 +41,6 @@ std::string GetRunningPath(std::string string_to_erase) {
path.clear(); // reset path
path.append(real_path);
//std::cout << path << std::endl;
size_t pos = path.find(to_erase);
if (pos != std::string::npos) path.erase(pos, to_erase.length());
} else {
projects/rocprofiler/tests/unittests/amdsys.cpp
-265
View File
@@ -1,265 +0,0 @@
// TODO(aelwazir): To be checked
#include "hip/hip_runtime.h"
#include <cstdio>
#include <unistd.h>
#include <hip/hip_profile.h>
#include <iostream>
#define N 2560
//change here to run this app longer
#define num_iters 1
template<int n, int m>
__global__ void kernel(double* x) {
for (int idx = threadIdx.x + blockIdx.x * blockDim.x; idx < N; idx += gridDim.x * blockDim.x)
{
#pragma unroll
for (int i = 0; i < n; ++i)
x[idx] += i * m;
}
}
void cpuWork() {
// Do some CPU "work".
usleep(1000);
}
inline void hip_assert(hipError_t err, const char *file, int line)
{
if (err != hipSuccess)
{
fprintf(stderr,"HIP error: %s %s %d\n", hipGetErrorString(err), file, line);
exit(-1);
}
}
#define hipErrorCheck(f) { hip_assert((f), __FILE__, __LINE__); }
#define kernelErrorCheck() { hipErrorCheck(hipPeekAtLastError()); }
int main() {
double* x;
double* x_h;
size_t sz = N * sizeof(double);
std::cout << "running app....." << std::endl;
hipErrorCheck(hipHostMalloc(&x_h, sz));
memset(x_h, 0, sz);
hipErrorCheck(hipMallocManaged(&x, sz));
hipErrorCheck(hipMemset(x, 0, sz));
hipStream_t stream;
hipErrorCheck(hipStreamCreate(&stream));
hipFuncAttributes attr;
int blocks = 80;
int threads = 32;
int fact = 100;
for (int j = 0; j < num_iters; ++j) {
for (int n = 0; n < 25*fact; ++n) {
hipErrorCheck(hipMemcpyAsync(x, x_h, sz, hipMemcpyHostToDevice));
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,1>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,2>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,3>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,4>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,5>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,6>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,7>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,8>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,9>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,10>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,11>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,12>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,13>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,14>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,15>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,16>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,17>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,18>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,19>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,20>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,20>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,21>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,22>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,23>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,24>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,25>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,26>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,27>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,28>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,29>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,30>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,30>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,31>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,32>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,33>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,34>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,35>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,36>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,37>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,38>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,39>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<1,40>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipErrorCheck(hipMemcpyAsync(x_h, x, sz, hipMemcpyDeviceToHost));
hipErrorCheck(hipDeviceSynchronize());
}
hipErrorCheck(hipMemset(x, 0, sz));
cpuWork();
for (int n = 0; n < 200*fact; ++n) {
hipErrorCheck(hipFuncGetAttributes(&attr, reinterpret_cast<const void*>(kernel<10,1>)));
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<10,1>), dim3(blocks), dim3(threads), 0, stream, x);
kernelErrorCheck();
hipErrorCheck(hipStreamSynchronize(stream));
}
hipErrorCheck(hipMemset(x, 0, sz));
cpuWork();
for (int n = 0; n < 30*fact; ++n) {
for (int k = 0; k < 7; ++k) {
hipErrorCheck(hipFuncGetAttributes(&attr, reinterpret_cast<const void*>(kernel<8,1>)));
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<8,1>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
}
hipErrorCheck(hipDeviceSynchronize());
}
hipErrorCheck(hipMemset(x, 0, sz));
cpuWork();
for (int n = 0; n < 100*fact; ++n) {
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,1>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,2>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,3>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,4>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,5>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipErrorCheck(hipDeviceSynchronize());
}
hipErrorCheck(hipMemset(x, 0, sz));
cpuWork();
for (int n = 0; n < 100*fact; ++n) {
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,1>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,2>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,3>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,4>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<7,5>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipErrorCheck(hipDeviceSynchronize());
}
hipErrorCheck(hipMemset(x, 0, sz));
cpuWork();
for (int n = 0; n < 50*fact; ++n) {
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<6,1>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<6,2>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<6,3>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<6,4>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<6,5>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<6,6>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<6,7>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<6,8>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipErrorCheck(hipDeviceSynchronize());
}
hipErrorCheck(hipMemset(x, 0, sz));
cpuWork();
for (int n = 0; n < 50*fact; ++n) {
int val;
hipErrorCheck(hipDeviceGetAttribute(&val, hipDeviceAttributeMaxThreadsPerBlock, 0));
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<4000,1>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipErrorCheck(hipDeviceSynchronize());
}
hipErrorCheck(hipMemset(x, 0, sz));
cpuWork();
for (int n = 0; n < 50*fact; ++n) {
hipLaunchKernelGGL(HIP_KERNEL_NAME(kernel<5000,1>), dim3(blocks), dim3(threads), 0, 0, x);
kernelErrorCheck();
hipErrorCheck(hipDeviceSynchronize());
}
hipErrorCheck(hipMemset(x, 0, sz));
cpuWork();
hipErrorCheck(hipDeviceSynchronize());
}
hipErrorCheck(hipHostFree(x_h));
hipErrorCheck(hipFree(x));
hipErrorCheck(hipStreamDestroy(stream));
}