lxl

CMakeLists.txt

+cmake_minimum_required(VERSION 3.20)
+
+project(01test LANGUAGES C CXX HIP)
+
+set(CMAKE_CXX_COMPILER g++)
+set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -std=c++17 -pthread")
+set(CMAKE_BUILD_TYPE release)
+
+if(NOT DEFINED ENV{ROCM_PATH})
+  set(ROCM_PATH /opt/dtk)
+else()
+  set(ROCM_PATH $ENV{ROCM_PATH})
+endif()
+add_definitions(-D__HIP_PLATFORM_AMD__)
+
+set(INCLUDE_PATH    /usr/local/include
+                    ${CMAKE_CURRENT_SOURCE_DIR}/include
+                    $ENV{ROCM_PATH}/include)
+
+include_directories(${INCLUDE_PATH})
+message(${ROCM_PATH}/lib)
+link_directories(${ROCM_PATH}/lib)
+
+set(LIBRARY      galaxyhip)
+link_libraries(${LIBRARY})
+
+set(SOURCE_FILES    ${CMAKE_CURRENT_SOURCE_DIR}/src/test.hip
+                    ${CMAKE_CURRENT_SOURCE_DIR}/src/main.cpp
+                    )
+
+
+add_executable(01test ${SOURCE_FILES})
+

include/common.h

+#include <iostream>
\ No newline at end of file

include/test.hpp

+#include <hip/hip_runtime.h>
+#define N 10000
+
+extern "C" float * compute(float *A, float *B, float *C);
\ No newline at end of file

src/main.cpp

+#include "test.hpp"
+
+#include "common.h"
+
+int main()
+{
+    float *A = (float *) malloc(N * sizeof(float)); 
+    float *B = (float *) malloc(N * sizeof(float)); 
+    float *C = (float *) malloc(N * sizeof(float)); 
+    for (int i = 0; i < N; i++) {
+        A[i] = 1; 
+        B[i] = 1; 
+        C[i] = 0; 
+    }
+    C = compute(A, B, C);
+    for (int i = 0; i < N; i++) { 
+        std::cout << C[i] << std::endl; 
+    }
+    free(A); 
+    free(B); 
+    free(C); 
+    return 0;
+}
\ No newline at end of file

src/test.hip

+#include "test.hpp"
+__global__ void add(float *d_A, float *d_B, float *d_C) { 
+    int tid = threadIdx.x + blockIdx.x * blockDim.x; 
+    if (tid < N) { 
+        d_C[tid] = d_A[tid] + d_B[tid]; 
+    } 
+}
+extern "C" float * compute(float *A, float *B, float *C) { 
+    // float *A = (float *) malloc(N * sizeof(float)); 
+    // float *B = (float *) malloc(N * sizeof(float)); 
+    // float *C = (float *) malloc(N * sizeof(float)); 
+    float *d_A = NULL; 
+    float *d_B = NULL; 
+    float *d_C = NULL; 
+    hipMalloc((void **) &d_A, N * sizeof(float)); 
+    hipMalloc((void **) &d_B, N * sizeof(float)); 
+    hipMalloc((void **) &d_C, N * sizeof(float)); 
+    // for (int i = 0; i < N; i++) {
+	//     A[i] = 1; 
+    //     B[i] = 1; 
+    //     C[i] = 0; 
+    // } 
+    hipMemcpy(d_A, A, sizeof(float) * N, hipMemcpyHostToDevice); 
+    hipMemcpy(d_B, B, sizeof(float) * N, hipMemcpyHostToDevice); 
+    hipMemcpy(d_C, C, sizeof(float) * N, hipMemcpyHostToDevice); 
+    dim3 blocksize(256, 1); 
+    dim3 gridsize(N / 256 + 1, 1); 
+    add<<<gridsize, blocksize >>> (d_A, d_B, d_C); 
+    hipMemcpy(C, d_C, sizeof(float) * N, hipMemcpyDeviceToHost); 
+    hipFree(d_A); 
+    hipFree(d_B); 
+    hipFree(d_C);
+    return C;
+}