From 0f6c153774b4890404445c32b14b0c90e56d0ba3 Mon Sep 17 00:00:00 2001 From: Chris Kitching Date: Wed, 18 Oct 2017 08:45:16 +0100 Subject: [PATCH] Migrate lit test to using FileCheck, so failures are readable It seems the test is already broken, but look how awesome the error message is now: /home/chris/HIP/tests/hipify-clang/axpy.cu:31:12: error: expected string not found in input // CHECK: hipLaunchKernel(HIP_KERNEL_NAME(axpy), dim3(1), dim3(kDataLen), 0, 0, a, device_x, device_y); ^ :31:2: note: scanning from here // ^ :33:2: note: possible intended match here hipLaunchKernelGGL(axpy, dim3(1), dim3(kDataLen), 0, 0, a, device_x, device_y); ^ [ROCm/clr commit: 74fd64d5c132081b56301e07b1220a608d8d8ac0] --- .../clr/hipamd/tests/hipify-clang/axpy.cu | 22 +++++++++++-------- 1 file changed, 13 insertions(+), 9 deletions(-) diff --git a/projects/clr/hipamd/tests/hipify-clang/axpy.cu b/projects/clr/hipamd/tests/hipify-clang/axpy.cu index 56a854fd0a..e0afa02a26 100644 --- a/projects/clr/hipamd/tests/hipify-clang/axpy.cu +++ b/projects/clr/hipamd/tests/hipify-clang/axpy.cu @@ -1,9 +1,9 @@ -// RUN: hipify "%s" -o=%t -- %cuda_args +// RUN: hipify "%s" -o=%t -- %cuda_args && cat %t | sed -Ee 's|//.+|// |g' | FileCheck %s #include __global__ void axpy(float a, float* x, float* y) { - // RUN: sh -c "test `grep -c -F 'y[hipThreadIdx_x] = a * x[hipThreadIdx_x];' %t` -eq 2" + // CHECK: y[hipThreadIdx_x] = a * x[hipThreadIdx_x]; y[threadIdx.x] = a * x[threadIdx.x]; } @@ -17,21 +17,25 @@ int main(int argc, char* argv[]) { // Copy input data to device. float* device_x; float* device_y; - // RUN: sh -c "test `grep -c -F 'hipMalloc(&device_x, kDataLen * sizeof(float));' %t` -eq 2" + + // CHECK: hipMalloc(&device_x, kDataLen * sizeof(float)); cudaMalloc(&device_x, kDataLen * sizeof(float)); - // RUN: sh -c "test `grep -c -F 'hipMalloc(&device_y, kDataLen * sizeof(float));' %t` -eq 2" + + // CHECK: hipMalloc(&device_y, kDataLen * sizeof(float)); cudaMalloc(&device_y, kDataLen * sizeof(float)); - // RUN: sh -c "test `grep -c -F 'hipMemcpy(device_x, host_x, kDataLen * sizeof(float), hipMemcpyHostToDevice);' %t` -eq 2" + + // CHECK: hipMemcpy(device_x, host_x, kDataLen * sizeof(float), hipMemcpyHostToDevice); cudaMemcpy(device_x, host_x, kDataLen * sizeof(float), cudaMemcpyHostToDevice); // Launch the kernel. - // RUN: sh -c "test `grep -c -F 'hipLaunchKernel(HIP_KERNEL_NAME(axpy), dim3(1), dim3(kDataLen), 0, 0, a, device_x, device_y);' %t` -eq 2" + // CHECK: hipLaunchKernel(HIP_KERNEL_NAME(axpy), dim3(1), dim3(kDataLen), 0, 0, a, device_x, device_y); axpy<<<1, kDataLen>>>(a, device_x, device_y); // Copy output data to host. - // RUN: sh -c "test `grep -c -F 'hipDeviceSynchronize();' %t` -eq 2" + // CHECK: hipDeviceSynchronize(); cudaDeviceSynchronize(); - // RUN: sh -c "test `grep -c -F 'hipMemcpy(host_y, device_y, kDataLen * sizeof(float), hipMemcpyDeviceToHost);' %t` -eq 2" + + // CHECK: hipMemcpy(host_y, device_y, kDataLen * sizeof(float), hipMemcpyDeviceToHost); cudaMemcpy(host_y, device_y, kDataLen * sizeof(float), cudaMemcpyDeviceToHost); // Print the results. @@ -39,7 +43,7 @@ int main(int argc, char* argv[]) { std::cout << "y[" << i << "] = " << host_y[i] << "\n"; } - // RUN: sh -c "test `grep -c -F 'hipDeviceReset();' %t` -eq 2" + // CHECK: hipDeviceReset(); cudaDeviceReset(); return 0; }