From d6dd38f5210fedfd96ff088ef9b8a2424c79c4f0 Mon Sep 17 00:00:00 2001
From: Ellie Hermaszewska <ellieh@nvidia.com>
Date: Tue, 11 Apr 2023 23:28:58 +0800
Subject: Implement FileCheck tests for several test commands (#2747)

* Add missing expected.txt for test

* Diagnostics -> StdWriters in render test

* Allow specifying several test prefixes to run

`slang-test -- tests/foo tests/bar`

* Squash warnings in some tests

* Enable gfx debug layer in gfx test util

Makes this issue present consistently: https://github.com/shader-slang/slang/issues/2766

* Allow DebugDevice to return interfaces instantiated by the debugged object

* Check that we actaully have a shader cache for shader cache tests

* Implement FileCheck tests for several test commands

- SIMPLE, SIMPLE_EX
- SIMPLE_LINE
- REFLECTION, CPU_REFLECTION
- CROSS_COMPILE

It does not currently support the render tests or the COMPARE_COMPUTE commands

It is invoked by adding `(filecheck=MY_FILECHECK_PREFIX)` to the test command, for example

TEST:CROSS_COMPILE(filecheck=SPIRV): -target spirv-assembly

* Move LLVM FileCheck interface to slang-llvm

* Neaten slang-test tests

* Refine handling of expected output in slang-test

* Add example FileCheck buffer test

* Add cuda-kernel-export tests

Which were waiting on FileCheck

* Bump vs project files

* Make createLLVMFileCheck_V1 return a void* rather than specifically an IFileCheck

* Remove use of CharSlice from filecheck interface

* Bump slang-llvm version

---------

Co-authored-by: jsmall-nvidia <jsmall@nvidia.com>
---
 tests/autodiff/cuda-kernel-export.slang | 14 ++++++++++----
 1 file changed, 10 insertions(+), 4 deletions(-)

(limited to 'tests/autodiff/cuda-kernel-export.slang')

diff --git a/tests/autodiff/cuda-kernel-export.slang b/tests/autodiff/cuda-kernel-export.slang
index e16188abc..928133c94 100644
--- a/tests/autodiff/cuda-kernel-export.slang
+++ b/tests/autodiff/cuda-kernel-export.slang
@@ -1,7 +1,7 @@
-//DISABLE_TEST:SIMPLE: -target cuda -line-directive-mode none
+//TEST:SIMPLE(filecheck=CUDA): -target cuda -line-directive-mode none
+//TEST:SIMPLE(filecheck=TORCH): -target torch -line-directive-mode none
 
-// Verify that we can output a cuda device function with [CudaDeviceExport].
-// Disabled until we have FileCheck.
+// Verify that we can output a cuda device function with [CudaKernel].
 
 struct MySubType
 {
@@ -20,6 +20,7 @@ struct MyInput
     float normalVal;
 }
 
+// CUDA: __global__ void myKernel(TensorView inValues_[[#]], TensorView outValues_[[#]])
 [CudaKernel]
 void myKernel(TensorView<float> inValues, TensorView<float> outValues)
 {
@@ -28,6 +29,11 @@ void myKernel(TensorView<float> inValues, TensorView<float> outValues)
     outValues.store(cudaThreadIdx().x, sin(inValues.load(cudaThreadIdx().x)));
 }
 
+// TORCH:      {{^SLANG_PRELUDE_EXPORT$}}
+// TORCH-NEXT: void myKernel(TensorView {{[[:alnum:]_]+}}, TensorView {{[[:alnum:]_]+}});
+//
+// TORCH:      {{^SLANG_PRELUDE_EXPORT$}}
+// TORCH-NEXT: std::tuple<std::tuple<float, float>, std::tuple<std::tuple<std::tuple<torch::Tensor, torch::Tensor>>, std::tuple<std::tuple<torch::Tensor, torch::Tensor>>>> runCompute(std::tuple<torch::Tensor, float> input_[[#]])
 [TorchEntryPoint]
 public __extern_cpp MyType runCompute(MyInput input)
 {
@@ -44,4 +50,4 @@ public __extern_cpp MyType runCompute(MyInput input)
     rs.sub[1].array[0] = inValues;
     rs.sub[1].array[1] = outValues;
     return rs;
-}
\ No newline at end of file
+}
-- 
cgit v1.2.3