Merge pull request #262 from ChrisKitching/frontendaction

[HIPIFY] Mostly fix preprocessor-or-template induced issues

[ROCm/hip commit: aa05b3d84e]

projects/hip/tests/hipify-clang/allocators.cu

@@ -0,0 +1,18 @@
+// RUN: %run_test hipify "%s" "%t" %cuda_args
+
+#pragma once
+
+#include <cuda_runtime.h>
+
+
+/**
+ * Allocate GPU memory for `count` elements of type `T`.
+ */
+template<typename T>
+static T* gpuMalloc(size_t count) {
+    T* ret = nullptr;
+    // CHECK: hipMalloc(&ret, count * sizeof(T));
+    cudaMalloc(&ret, count * sizeof(T));
+    return ret;
+}
+

projects/hip/tests/hipify-clang/axpy.cu

@@ -2,6 +2,8 @@
 
 #include <iostream>
 
+// CHECK: #include <hip/hip_runtime.h>
+#include <cuda.h>
 
 #define TOKEN_PASTE(X, Y) X ## Y
 #define ARG_LIST_AS_MACRO a, device_x, device_y
@@ -33,8 +35,13 @@ int main(int argc, char* argv[]) {
   // CHECK: hipMalloc(&device_x, kDataLen * sizeof(float));
   cudaMalloc(&device_x, kDataLen * sizeof(float));
 
+#ifdef HERRING
   // CHECK: hipMalloc(&device_y, kDataLen * sizeof(float));
   cudaMalloc(&device_y, kDataLen * sizeof(float));
+#else
+  // CHECK: hipMalloc(&device_y, kDataLen * sizeof(double));
+  cudaMalloc(&device_y, kDataLen * sizeof(double));
+#endif
 
   // CHECK: hipMemcpy(device_x, host_x, kDataLen * sizeof(float), hipMemcpyHostToDevice);
   cudaMemcpy(device_x, host_x, kDataLen * sizeof(float), cudaMemcpyHostToDevice);

projects/hip/tests/hipify-clang/concurentKernels.cu

@@ -99,6 +99,7 @@ int main(int argc, char **argv)
     // use command-line specified CUDA device, otherwise use device with highest Gflops/s
     cuda_device = findCudaDevice(argc, (const char **)argv);
 
+    // CHECK: hipDeviceProp_t deviceProp;
     cudaDeviceProp deviceProp;
     // CHECK: checkCudaErrors(hipGetDevice(&cuda_device));
     checkCudaErrors(cudaGetDevice(&cuda_device));
@@ -135,6 +136,7 @@ int main(int argc, char **argv)
         checkCudaErrors(cudaStreamCreate(&(streams[i])));
     }
 
+    // CHECK: hipEvent_t start_event, stop_event;
     // create CUDA event handles
     cudaEvent_t start_event, stop_event;