92 files changed, 244 insertions, 107 deletions

diff --git a/Copyright.txt b/Copyright.txt
index 7f51293..b703193 100644
--- a/Copyright.txt
+++ b/Copyright.txt
@@ -63,6 +63,7 @@ The following individuals and institutions are among the Contributors:
 * Ilya Lavrenov <ilya.lavrenov@itseez.com>
 * Insight Software Consortium <insightsoftwareconsortium.org>
 * Jan Woetzel
+* Jordan Williams <jordan@jwillikers.com>
 * Julien Schueller
 * Kelly Thompson <kgt@lanl.gov>
 * Konstantin Podsvirov <konstantin@podsvirov.pro>
diff --git a/Help/release/dev/cuda-ptx-separable-compilation.rst b/Help/release/dev/cuda-ptx-separable-compilation.rst
new file mode 100644
index 0000000..49c66c0
--- /dev/null
+++ b/Help/release/dev/cuda-ptx-separable-compilation.rst
@@ -0,0 +1,5 @@
+cuda-ptx-separable-compilation
+------------------------------
+
+* ``CUDA`` targets can now enable both :prop_tgt:`CUDA_SEPARABLE_COMPILATION` and
+  :prop_tgt:`CUDA_PTX_COMPILATION`.
diff --git a/Modules/CMakeCUDAInformation.cmake b/Modules/CMakeCUDAInformation.cmake
index e9cfed6..94b67dc 100644
--- a/Modules/CMakeCUDAInformation.cmake
+++ b/Modules/CMakeCUDAInformation.cmake
@@ -160,22 +160,9 @@ if(NOT DEFINED CMAKE_CUDA_ARCHIVE_FINISH)
   set(CMAKE_CUDA_ARCHIVE_FINISH "<CMAKE_RANLIB> <TARGET>")
 endif()
 
-#Specify how to compile when ptx has been requested
-if(NOT CMAKE_CUDA_COMPILE_PTX_COMPILATION)
-  set(CMAKE_CUDA_COMPILE_PTX_COMPILATION
-    "<CMAKE_CUDA_COMPILER> ${_CMAKE_CUDA_EXTRA_FLAGS} <DEFINES> <INCLUDES> <FLAGS> ${_CMAKE_COMPILE_AS_CUDA_FLAG} ${_CMAKE_CUDA_PTX_FLAG} <SOURCE> -o <OBJECT>")
-endif()
-
-#Specify how to compile when separable compilation has been requested
-if(NOT CMAKE_CUDA_COMPILE_SEPARABLE_COMPILATION)
-  set(CMAKE_CUDA_COMPILE_SEPARABLE_COMPILATION
-    "<CMAKE_CUDA_COMPILER> ${_CMAKE_CUDA_EXTRA_FLAGS} <DEFINES> <INCLUDES> <FLAGS> ${_CMAKE_COMPILE_AS_CUDA_FLAG} ${_CMAKE_CUDA_DEVICE_CODE} <SOURCE> -o <OBJECT>")
-endif()
-
-#Specify how to compile when whole compilation has been requested
-if(NOT CMAKE_CUDA_COMPILE_WHOLE_COMPILATION)
-  set(CMAKE_CUDA_COMPILE_WHOLE_COMPILATION
-    "<CMAKE_CUDA_COMPILER> ${_CMAKE_CUDA_EXTRA_FLAGS} <DEFINES> <INCLUDES> <FLAGS> ${_CMAKE_COMPILE_AS_CUDA_FLAG} -c <SOURCE> -o <OBJECT>")
+if(NOT CMAKE_CUDA_COMPILE_OBJECT)
+  set(CMAKE_CUDA_COMPILE_OBJECT
+    "<CMAKE_CUDA_COMPILER> ${_CMAKE_CUDA_EXTRA_FLAGS} <DEFINES> <INCLUDES> <FLAGS> ${_CMAKE_COMPILE_AS_CUDA_FLAG} <CUDA_COMPILE_MODE> <SOURCE> -o <OBJECT>")
 endif()
 
 # compile a cu file into an executable
diff --git a/Modules/Compiler/Clang-CUDA.cmake b/Modules/Compiler/Clang-CUDA.cmake
index 0223081..b105518 100644
--- a/Modules/Compiler/Clang-CUDA.cmake
+++ b/Modules/Compiler/Clang-CUDA.cmake
@@ -18,8 +18,9 @@ __compiler_clang_cxx_standards(CUDA)
 
 set(CMAKE_CUDA_COMPILER_HAS_DEVICE_LINK_PHASE TRUE)
 set(_CMAKE_COMPILE_AS_CUDA_FLAG "-x cuda")
+set(_CMAKE_CUDA_WHOLE_FLAG "-c")
+set(_CMAKE_CUDA_RDC_FLAG "-fgpu-rdc")
 set(_CMAKE_CUDA_PTX_FLAG "--cuda-device-only -S")
-set(_CMAKE_CUDA_DEVICE_CODE "-fgpu-rdc -c")
 
 # RulePlaceholderExpander expands crosscompile variables like sysroot and target only for CMAKE_<LANG>_COMPILER. Override the default.
 set(CMAKE_CUDA_LINK_EXECUTABLE "<CMAKE_CUDA_COMPILER> <LINK_FLAGS> <OBJECTS> -o <TARGET> <LINK_LIBRARIES>${__IMPLICIT_LINKS}")
diff --git a/Modules/Compiler/NVIDIA-CUDA.cmake b/Modules/Compiler/NVIDIA-CUDA.cmake
index c2fe42d..2f12b43 100644
--- a/Modules/Compiler/NVIDIA-CUDA.cmake
+++ b/Modules/Compiler/NVIDIA-CUDA.cmake
@@ -5,8 +5,9 @@ set(CMAKE_CUDA_VERBOSE_FLAG "-v")
 set(CMAKE_CUDA_VERBOSE_COMPILE_FLAG "-Xcompiler=-v")
 
 set(_CMAKE_COMPILE_AS_CUDA_FLAG "-x cu")
+set(_CMAKE_CUDA_WHOLE_FLAG "-c")
+set(_CMAKE_CUDA_RDC_FLAG "-rdc=true")
 set(_CMAKE_CUDA_PTX_FLAG "-ptx")
-set(_CMAKE_CUDA_DEVICE_CODE "-dc")
 
 if (CMAKE_CUDA_COMPILER_VERSION VERSION_GREATER_EQUAL 10.2.89)
   # The -forward-unknown-to-host-compiler flag was only
diff --git a/Modules/Platform/Generic-ELF.cmake b/Modules/Platform/Generic-ELF.cmake
new file mode 100644
index 0000000..943cb6b
--- /dev/null
+++ b/Modules/Platform/Generic-ELF.cmake
@@ -0,0 +1,7 @@
+# This is a platform definition file for platforms without
+# an operating system using the ELF executable format.
+# It is used when CMAKE_SYSTEM_NAME is set to "Generic-ELF"
+
+include(Platform/Generic)
+
+set(CMAKE_EXECUTABLE_SUFFIX .elf)
diff --git a/Modules/Platform/Windows-NVIDIA-CUDA.cmake b/Modules/Platform/Windows-NVIDIA-CUDA.cmake
index b83932e..6c1699b 100644
--- a/Modules/Platform/Windows-NVIDIA-CUDA.cmake
+++ b/Modules/Platform/Windows-NVIDIA-CUDA.cmake
@@ -1,11 +1,7 @@
 include(Platform/Windows-MSVC)
 
-set(CMAKE_CUDA_COMPILE_PTX_COMPILATION
-  "<CMAKE_CUDA_COMPILER> ${_CMAKE_CUDA_EXTRA_FLAGS} <DEFINES> <INCLUDES> <FLAGS> ${_CMAKE_COMPILE_AS_CUDA_FLAG} -ptx <SOURCE> -o <OBJECT> -Xcompiler=-Fd<TARGET_COMPILE_PDB>,-FS")
-set(CMAKE_CUDA_COMPILE_SEPARABLE_COMPILATION
-  "<CMAKE_CUDA_COMPILER> ${_CMAKE_CUDA_EXTRA_FLAGS} <DEFINES> <INCLUDES> <FLAGS> ${_CMAKE_COMPILE_AS_CUDA_FLAG} -dc <SOURCE> -o <OBJECT> -Xcompiler=-Fd<TARGET_COMPILE_PDB>,-FS")
-set(CMAKE_CUDA_COMPILE_WHOLE_COMPILATION
-  "<CMAKE_CUDA_COMPILER> ${_CMAKE_CUDA_EXTRA_FLAGS} <DEFINES> <INCLUDES> <FLAGS> ${_CMAKE_COMPILE_AS_CUDA_FLAG} -c <SOURCE> -o <OBJECT> -Xcompiler=-Fd<TARGET_COMPILE_PDB>,-FS")
+set(CMAKE_CUDA_COMPILE_OBJECT
+  "<CMAKE_CUDA_COMPILER> ${_CMAKE_CUDA_EXTRA_FLAGS} <DEFINES> <INCLUDES> <FLAGS> ${_CMAKE_COMPILE_AS_CUDA_FLAG} <CUDA_COMPILE_MODE> <SOURCE> -o <OBJECT> -Xcompiler=-Fd<TARGET_COMPILE_PDB>,-FS")
 
 set(__IMPLICIT_LINKS)
 foreach(dir ${CMAKE_CUDA_HOST_IMPLICIT_LINK_DIRECTORIES})
diff --git a/Source/CMakeVersion.cmake b/Source/CMakeVersion.cmake
index 8a5ecc5..c39bea6 100644
--- a/Source/CMakeVersion.cmake
+++ b/Source/CMakeVersion.cmake
@@ -1,7 +1,7 @@
 # CMake version number components.
 set(CMake_VERSION_MAJOR 3)
 set(CMake_VERSION_MINOR 22)
-set(CMake_VERSION_PATCH 20211020)
+set(CMake_VERSION_PATCH 20211021)
 #set(CMake_VERSION_RC 0)
 set(CMake_VERSION_IS_DIRTY 0)
 
diff --git a/Source/cmMakefileTargetGenerator.cxx b/Source/cmMakefileTargetGenerator.cxx
index 8edadd3..9f2ae19 100644
--- a/Source/cmMakefileTargetGenerator.cxx
+++ b/Source/cmMakefileTargetGenerator.cxx
@@ -897,28 +897,31 @@ void cmMakefileTargetGenerator::WriteObjectRuleFiles(
 
   // Construct the compile rules.
   {
-    std::vector<std::string> compileCommands;
+    std::string cudaCompileMode;
     if (lang == "CUDA") {
-      std::string cmdVar;
       if (this->GeneratorTarget->GetPropertyAsBool(
             "CUDA_SEPARABLE_COMPILATION")) {
-        cmdVar = "CMAKE_CUDA_COMPILE_SEPARABLE_COMPILATION";
-      } else if (this->GeneratorTarget->GetPropertyAsBool(
-                   "CUDA_PTX_COMPILATION")) {
-        cmdVar = "CMAKE_CUDA_COMPILE_PTX_COMPILATION";
+        const std::string& rdcFlag =
+          this->Makefile->GetRequiredDefinition("_CMAKE_CUDA_RDC_FLAG");
+        cudaCompileMode = cmStrCat(cudaCompileMode, rdcFlag, " ");
+      }
+      if (this->GeneratorTarget->GetPropertyAsBool("CUDA_PTX_COMPILATION")) {
+        const std::string& ptxFlag =
+          this->Makefile->GetRequiredDefinition("_CMAKE_CUDA_PTX_FLAG");
+        cudaCompileMode = cmStrCat(cudaCompileMode, ptxFlag);
       } else {
-        cmdVar = "CMAKE_CUDA_COMPILE_WHOLE_COMPILATION";
+        const std::string& wholeFlag =
+          this->Makefile->GetRequiredDefinition("_CMAKE_CUDA_WHOLE_FLAG");
+        cudaCompileMode = cmStrCat(cudaCompileMode, wholeFlag);
       }
-      const std::string& compileRule =
-        this->Makefile->GetRequiredDefinition(cmdVar);
-      cmExpandList(compileRule, compileCommands);
-    } else {
-      const std::string cmdVar = "CMAKE_" + lang + "_COMPILE_OBJECT";
-      const std::string& compileRule =
-        this->Makefile->GetRequiredDefinition(cmdVar);
-      cmExpandList(compileRule, compileCommands);
+      vars.CudaCompileMode = cudaCompileMode.c_str();
     }
 
+    std::vector<std::string> compileCommands;
+    const std::string& compileRule = this->Makefile->GetRequiredDefinition(
+      "CMAKE_" + lang + "_COMPILE_OBJECT");
+    cmExpandList(compileRule, compileCommands);
+
     if (this->GeneratorTarget->GetPropertyAsBool("EXPORT_COMPILE_COMMANDS") &&
         lang_can_export_cmds && compileCommands.size() == 1) {
       std::string compileCommand = compileCommands[0];
diff --git a/Source/cmNinjaTargetGenerator.cxx b/Source/cmNinjaTargetGenerator.cxx
index 57fc020..2c45e1b 100644
--- a/Source/cmNinjaTargetGenerator.cxx
+++ b/Source/cmNinjaTargetGenerator.cxx
@@ -605,6 +605,7 @@ void cmNinjaTargetGenerator::WriteCompileRule(const std::string& lang,
   vars.TargetCompilePDB = "$TARGET_COMPILE_PDB";
   vars.ObjectDir = "$OBJECT_DIR";
   vars.ObjectFileDir = "$OBJECT_FILE_DIR";
+  vars.CudaCompileMode = "$CUDA_COMPILE_MODE";
   vars.ISPCHeader = "$ISPC_HEADER_FILE";
 
   cmMakefile* mf = this->GetMakefile();
@@ -815,27 +816,32 @@ void cmNinjaTargetGenerator::WriteCompileRule(const std::string& lang,
   vars.Flags = flags.c_str();
   vars.DependencyFile = rule.DepFile.c_str();
 
-  // Rule for compiling object file.
-  std::vector<std::string> compileCmds;
+  std::string cudaCompileMode;
   if (lang == "CUDA") {
-    std::string cmdVar;
     if (this->GeneratorTarget->GetPropertyAsBool(
           "CUDA_SEPARABLE_COMPILATION")) {
-      cmdVar = "CMAKE_CUDA_COMPILE_SEPARABLE_COMPILATION";
-    } else if (this->GeneratorTarget->GetPropertyAsBool(
-                 "CUDA_PTX_COMPILATION")) {
-      cmdVar = "CMAKE_CUDA_COMPILE_PTX_COMPILATION";
+      const std::string& rdcFlag =
+        this->Makefile->GetRequiredDefinition("_CMAKE_CUDA_RDC_FLAG");
+      cudaCompileMode = cmStrCat(cudaCompileMode, rdcFlag, " ");
+    }
+    if (this->GeneratorTarget->GetPropertyAsBool("CUDA_PTX_COMPILATION")) {
+      const std::string& ptxFlag =
+        this->Makefile->GetRequiredDefinition("_CMAKE_CUDA_PTX_FLAG");
+      cudaCompileMode = cmStrCat(cudaCompileMode, ptxFlag);
     } else {
-      cmdVar = "CMAKE_CUDA_COMPILE_WHOLE_COMPILATION";
+      const std::string& wholeFlag =
+        this->Makefile->GetRequiredDefinition("_CMAKE_CUDA_WHOLE_FLAG");
+      cudaCompileMode = cmStrCat(cudaCompileMode, wholeFlag);
     }
-    const std::string& compileCmd = mf->GetRequiredDefinition(cmdVar);
-    cmExpandList(compileCmd, compileCmds);
-  } else {
-    const std::string cmdVar = cmStrCat("CMAKE_", lang, "_COMPILE_OBJECT");
-    const std::string& compileCmd = mf->GetRequiredDefinition(cmdVar);
-    cmExpandList(compileCmd, compileCmds);
+    vars.CudaCompileMode = cudaCompileMode.c_str();
   }
 
+  // Rule for compiling object file.
+  std::vector<std::string> compileCmds;
+  const std::string cmdVar = cmStrCat("CMAKE_", lang, "_COMPILE_OBJECT");
+  const std::string& compileCmd = mf->GetRequiredDefinition(cmdVar);
+  cmExpandList(compileCmd, compileCmds);
+
   // See if we need to use a compiler launcher like ccache or distcc
   std::string compilerLauncher;
   if (!compileCmds.empty() &&
diff --git a/Source/cmRulePlaceholderExpander.cxx b/Source/cmRulePlaceholderExpander.cxx
index 7480aeb..4cee09d 100644
--- a/Source/cmRulePlaceholderExpander.cxx
+++ b/Source/cmRulePlaceholderExpander.cxx
@@ -85,6 +85,11 @@ std::string cmRulePlaceholderExpander::ExpandRuleVariable(
       return replaceValues.ObjectsQuoted;
     }
   }
+  if (replaceValues.CudaCompileMode) {
+    if (variable == "CUDA_COMPILE_MODE") {
+      return replaceValues.CudaCompileMode;
+    }
+  }
   if (replaceValues.AIXExports) {
     if (variable == "AIX_EXPORTS") {
       return replaceValues.AIXExports;
diff --git a/Source/cmRulePlaceholderExpander.h b/Source/cmRulePlaceholderExpander.h
index c22e0fa..852954f 100644
--- a/Source/cmRulePlaceholderExpander.h
+++ b/Source/cmRulePlaceholderExpander.h
@@ -65,6 +65,7 @@ public:
     const char* SwiftOutputFileMap = nullptr;
     const char* SwiftSources = nullptr;
     const char* ISPCHeader = nullptr;
+    const char* CudaCompileMode = nullptr;
     const char* Fatbinary = nullptr;
     const char* RegisterFile = nullptr;
     const char* Launcher = nullptr;
diff --git a/Source/cmVisualStudio10TargetGenerator.cxx b/Source/cmVisualStudio10TargetGenerator.cxx
index 1ea0461..4e93a4a 100644
--- a/Source/cmVisualStudio10TargetGenerator.cxx
+++ b/Source/cmVisualStudio10TargetGenerator.cxx
@@ -2337,7 +2337,8 @@ void cmVisualStudio10TargetGenerator::WriteAllSources(Elem& e0)
         }
       }
 
-      if (si.Kind == cmGeneratorTarget::SourceKindObjectSource) {
+      if (si.Kind == cmGeneratorTarget::SourceKindObjectSource ||
+          si.Kind == cmGeneratorTarget::SourceKindUnityBatched) {
         this->OutputSourceSpecificFlags(e2, si.Source);
       }
       if (si.Source->GetPropertyAsBool("SKIP_PRECOMPILE_HEADERS")) {
@@ -3210,18 +3211,17 @@ bool cmVisualStudio10TargetGenerator::ComputeCudaOptions(
   // the default to not have any extension
   cudaOptions.AddFlag("CompileOut", "$(IntDir)%(Filename).obj");
 
-  bool notPtx = true;
   if (this->GeneratorTarget->GetPropertyAsBool("CUDA_SEPARABLE_COMPILATION")) {
     cudaOptions.AddFlag("GenerateRelocatableDeviceCode", "true");
-  } else if (this->GeneratorTarget->GetPropertyAsBool(
-               "CUDA_PTX_COMPILATION")) {
+  }
+  bool notPtx = true;
+  if (this->GeneratorTarget->GetPropertyAsBool("CUDA_PTX_COMPILATION")) {
     cudaOptions.AddFlag("NvccCompilation", "ptx");
     // We drop the %(Extension) component as CMake expects all PTX files
     // to not have the source file extension at all
     cudaOptions.AddFlag("CompileOut", "$(IntDir)%(Filename).ptx");
     notPtx = false;
   }
-
   if (notPtx &&
       cmSystemTools::VersionCompareGreaterEq(
         "8.0", this->GlobalGenerator->GetPlatformToolsetCudaString())) {
diff --git a/Tests/CudaOnly/CMakeLists.txt b/Tests/CudaOnly/CMakeLists.txt
index a3fb409..65dfebb 100644
--- a/Tests/CudaOnly/CMakeLists.txt
+++ b/Tests/CudaOnly/CMakeLists.txt
@@ -16,6 +16,7 @@ add_cuda_test_macro(CudaOnly.WithDefs CudaOnlyWithDefs)
 add_cuda_test_macro(CudaOnly.CircularLinkLine CudaOnlyCircularLinkLine)
 add_cuda_test_macro(CudaOnly.ResolveDeviceSymbols CudaOnlyResolveDeviceSymbols)
 add_cuda_test_macro(CudaOnly.SeparateCompilation main/CudaOnlySeparateCompilation)
+add_cuda_test_macro(CudaOnly.SeparateCompilationPTX CudaOnlySeparateCompilationPTX)
 
 if(CMake_TEST_CUDA AND NOT CMake_TEST_CUDA STREQUAL "Clang")
   # Clang doesn't have flags for selecting the runtime.
diff --git a/Tests/CudaOnly/ExportPTX/CMakeLists.txt b/Tests/CudaOnly/ExportPTX/CMakeLists.txt
index e7e7bc4..f1667af 100644
--- a/Tests/CudaOnly/ExportPTX/CMakeLists.txt
+++ b/Tests/CudaOnly/ExportPTX/CMakeLists.txt
@@ -56,7 +56,7 @@ add_custom_command(
     "-DBIN_TO_C_COMMAND=${bin_to_c}"
     "-DOBJECTS=$<TARGET_OBJECTS:CudaPTX>"
     "-DOUTPUT=${output_file}"
-    -P ${CMAKE_CURRENT_SOURCE_DIR}/bin2c_wrapper.cmake
+    -P ${CMAKE_CURRENT_SOURCE_DIR}/../utils/bin2c_wrapper.cmake
   VERBATIM
   DEPENDS $<TARGET_OBJECTS:CudaPTX>
   COMMENT "Converting Object files to a C header"
diff --git a/Tests/CudaOnly/SeparateCompilationPTX/CMakeLists.txt b/Tests/CudaOnly/SeparateCompilationPTX/CMakeLists.txt
new file mode 100644
index 0000000..273f955
--- /dev/null
+++ b/Tests/CudaOnly/SeparateCompilationPTX/CMakeLists.txt
@@ -0,0 +1,51 @@
+cmake_minimum_required(VERSION 3.19)
+project (SeparateCompPTX CUDA)
+
+#Goal for this example:
+# How to generate PTX files with RDC enabled
+
+# PTX can be compiled only for a single virtual architecture at a time
+list(POP_FRONT CMAKE_CUDA_ARCHITECTURES temp)
+set(CMAKE_CUDA_ARCHITECTURES ${temp})
+string(APPEND CMAKE_CUDA_ARCHITECTURES "-virtual")
+
+add_library(CudaPTX OBJECT kernels.cu)
+set_property(TARGET CudaPTX PROPERTY CUDA_PTX_COMPILATION ON)
+set_property(TARGET CudaPTX PROPERTY CUDA_SEPARABLE_COMPILATION ON)
+
+
+set(output_file ${CMAKE_CURRENT_BINARY_DIR}/embedded_objs.h)
+
+find_package(CUDAToolkit REQUIRED)
+find_program(bin_to_c
+  NAMES bin2c
+  PATHS ${CUDAToolkit_BIN_DIR}
+  )
+if(NOT bin_to_c)
+  message(FATAL_ERROR
+    "bin2c not found:\n"
+    "  CUDAToolkit_BIN_DIR='${CUDAToolkit_BIN_DIR}'\n"
+    )
+endif()
+
+add_custom_command(
+  OUTPUT "${output_file}"
+  COMMAND ${CMAKE_COMMAND}
+    "-DBIN_TO_C_COMMAND=${bin_to_c}"
+    "-DOBJECTS=$<TARGET_OBJECTS:CudaPTX>"
+    "-DOUTPUT=${output_file}"
+    -P ${CMAKE_CURRENT_SOURCE_DIR}/../utils/bin2c_wrapper.cmake
+  VERBATIM
+  DEPENDS $<TARGET_OBJECTS:CudaPTX>
+  COMMENT "Converting Object files to a C header"
+  )
+
+add_executable(CudaOnlySeparateCompilationPTX main.cu ${output_file})
+target_compile_features(CudaOnlySeparateCompilationPTX PRIVATE cuda_std_11)
+target_include_directories(CudaOnlySeparateCompilationPTX PRIVATE
+                           ${CMAKE_CURRENT_BINARY_DIR} )
+target_link_libraries(CudaOnlySeparateCompilationPTX PRIVATE CUDA::cuda_driver)
+if(APPLE)
+  # Help the static cuda runtime find the driver (libcuda.dyllib) at runtime.
+  set_property(TARGET CudaOnlySeparateCompilationPTX PROPERTY BUILD_RPATH ${CMAKE_CUDA_IMPLICIT_LINK_DIRECTORIES})
+endif()
diff --git a/Tests/CudaOnly/SeparateCompilationPTX/kernels.cu b/Tests/CudaOnly/SeparateCompilationPTX/kernels.cu
new file mode 100644
index 0000000..f4a52d4
--- /dev/null
+++ b/Tests/CudaOnly/SeparateCompilationPTX/kernels.cu
@@ -0,0 +1,14 @@
+
+__global__ void kernelA(float* r, float* x, float* y, float* z, int size)
+{
+  for (int i = threadIdx.x; i < size; i += blockDim.x) {
+    r[i] = x[i] * y[i] + z[i];
+  }
+}
+
+__global__ void kernelB(float* r, float* x, float* y, float* z, int size)
+{
+  for (int i = threadIdx.x; i < size; i += blockDim.x) {
+    r[i] = x[i] * y[i] + z[i];
+  }
+}
diff --git a/Tests/CudaOnly/SeparateCompilationPTX/main.cu b/Tests/CudaOnly/SeparateCompilationPTX/main.cu
new file mode 100644
index 0000000..164cde5
--- /dev/null
+++ b/Tests/CudaOnly/SeparateCompilationPTX/main.cu
@@ -0,0 +1,30 @@
+#include <iostream>
+
+#include <cuda.h>
+
+#include "embedded_objs.h"
+
+int main()
+{
+  cuInit(0);
+  int count = 0;
+  cuDeviceGetCount(&count);
+  if (count == 0) {
+    std::cerr << "No CUDA devices found\n";
+    return 1;
+  }
+
+  CUdevice device;
+  cuDeviceGet(&device, 0);
+
+  CUcontext context;
+  cuCtxCreate(&context, 0, device);
+
+  CUmodule module;
+  cuModuleLoadData(&module, kernels);
+  if (module == nullptr) {
+    std::cerr << "Failed to load the embedded ptx" << std::endl;
+    return 1;
+  }
+  std::cout << module << std::endl;
+}
diff --git a/Tests/CudaOnly/ExportPTX/bin2c_wrapper.cmake b/Tests/CudaOnly/utils/bin2c_wrapper.cmake
index 0baf934..0baf934 100644
--- a/Tests/CudaOnly/ExportPTX/bin2c_wrapper.cmake
+++ b/Tests/CudaOnly/utils/bin2c_wrapper.cmake
diff --git a/Tests/RunCMake/CMakeLists.txt b/Tests/RunCMake/CMakeLists.txt
index c87c1b6..c566b42 100644
--- a/Tests/RunCMake/CMakeLists.txt
+++ b/Tests/RunCMake/CMakeLists.txt
@@ -347,7 +347,6 @@ endif()
 add_RunCMake_test(ScriptMode)
 add_RunCMake_test(Swift -DCMAKE_Swift_COMPILER=${CMAKE_Swift_COMPILER})
 add_RunCMake_test(TargetObjects)
-add_RunCMake_test(TargetSources)
 add_RunCMake_test(TargetProperties)
 add_RunCMake_test(ToolchainFile)
 add_RunCMake_test(find_dependency)
diff --git a/Tests/RunCMake/TargetSources/CMakeLists.txt b/Tests/RunCMake/TargetSources/CMakeLists.txt
deleted file mode 100644
index a06591c..0000000
--- a/Tests/RunCMake/TargetSources/CMakeLists.txt
+++ /dev/null
@@ -1,3 +0,0 @@
-cmake_minimum_required(VERSION 2.8.12)
-project(${RunCMake_TEST} CXX)
-include(${RunCMake_TEST}.cmake)
diff --git a/Tests/RunCMake/TargetSources/RelativePathInInterface-stdout.txt b/Tests/RunCMake/TargetSources/RelativePathInInterface-stdout.txt
deleted file mode 100644
index 4581d8a..0000000
--- a/Tests/RunCMake/TargetSources/RelativePathInInterface-stdout.txt
+++ /dev/null
@@ -1 +0,0 @@
--- iface: .*Tests/RunCMake/TargetSources/empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx-stdout.txt b/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx-stdout.txt
deleted file mode 100644
index 7f48082..0000000
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx-stdout.txt
+++ /dev/null
@@ -1 +0,0 @@
--- genexlib: \$<1:.*Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx/subdir_empty_1.cpp>;\$<1:.*Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx/../empty_1.cpp>;\$<1:empty_2.cpp>
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude-stdout.txt b/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude-stdout.txt
deleted file mode 100644
index aa4851f..0000000
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude-stdout.txt
+++ /dev/null
@@ -1 +0,0 @@
--- privatelib: .*Tests/RunCMake/TargetSources/RelativePathInSubdirInclude/subdir_empty_1.cpp;empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface-stdout.txt b/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface-stdout.txt
deleted file mode 100644
index 5990a05..0000000
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface-stdout.txt
+++ /dev/null
@@ -1 +0,0 @@
--- iface: .*Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/subdir_empty_1.cpp;.*Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/subdir_empty_2.cpp;.*Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/../empty_1.cpp;.*Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/../empty_2.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate-stdout.txt b/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate-stdout.txt
deleted file mode 100644
index fa5bcbf..0000000
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate-stdout.txt
+++ /dev/null
@@ -1 +0,0 @@
--- privatelib: .*Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/subdir_empty_1.cpp;.*Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/subdir_empty_2.cpp;.*Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/../empty_1.cpp;.*Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/../empty_2.cpp
diff --git a/Tests/RunCMake/TargetSources/RunCMakeTest.cmake b/Tests/RunCMake/TargetSources/RunCMakeTest.cmake
deleted file mode 100644
index b56ee44..0000000
--- a/Tests/RunCMake/TargetSources/RunCMakeTest.cmake
+++ /dev/null
@@ -1,22 +0,0 @@
-include(RunCMake)
-
-if(RunCMake_GENERATOR STREQUAL "Xcode")
-  run_cmake(ConfigNotAllowed)
-endif()
-
-run_cmake(OriginDebug)
-run_cmake(CMP0026-LOCATION)
-run_cmake(CMP0076-OLD)
-run_cmake(CMP0076-WARN)
-run_cmake(RelativePathInInterface)
-run_cmake(RelativePathInSubdirGenEx)
-run_cmake(RelativePathInSubdirInterface)
-run_cmake(RelativePathInSubdirPrivate)
-run_cmake(RelativePathInSubdirInclude)
-run_cmake(ExportBuild)
-run_cmake(AddCustomTargetPublicSources)
-run_cmake(AddCustomTargetPrivateSources)
-run_cmake(AddCustomTargetInterfaceSources)
-run_cmake(AddCustomTargetSources)
-run_cmake(AddCustomTargetCheckProperty)
-run_cmake(AddCustomTargetGenx)
diff --git a/Tests/RunCMake/UnityBuild/RunCMakeTest.cmake b/Tests/RunCMake/UnityBuild/RunCMakeTest.cmake
index c00f78b..8019f09 100644
--- a/Tests/RunCMake/UnityBuild/RunCMakeTest.cmake
+++ b/Tests/RunCMake/UnityBuild/RunCMakeTest.cmake
@@ -2,11 +2,9 @@ include(RunCMake)
 
 function(run_build name)
   set(RunCMake_TEST_BINARY_DIR ${RunCMake_BINARY_DIR}/${name}-build)
-  set(RunCMake_TEST_NO_CLEAN 1)
   run_cmake(${name})
+  set(RunCMake_TEST_NO_CLEAN 1)
   run_cmake_command(${name}-build ${CMAKE_COMMAND} --build . --config Debug)
-  unset(RunCMake_TEST_BINARY_DIR)
-  unset(RunCMake_TEST_NO_CLEAN)
 endfunction()
 
 run_cmake(unitybuild_c)
@@ -30,12 +28,10 @@ run_build(unitybuild_anon_ns_group_mode)
 
 function(run_test name)
   set(RunCMake_TEST_BINARY_DIR ${RunCMake_BINARY_DIR}/${name}-build)
-  set(RunCMake_TEST_NO_CLEAN 1)
   run_cmake(${name})
+  set(RunCMake_TEST_NO_CLEAN 1)
   run_cmake_command(${name}-build ${CMAKE_COMMAND} --build . --config Debug)
   run_cmake_command(${name}-test ${CMAKE_CTEST_COMMAND} -C Debug)
-  unset(RunCMake_TEST_BINARY_DIR)
-  unset(RunCMake_TEST_NO_CLEAN)
 endfunction()
 
 run_test(unitybuild_runtest)
diff --git a/Tests/RunCMake/VS10Project/RunCMakeTest.cmake b/Tests/RunCMake/VS10Project/RunCMakeTest.cmake
index d5ed136..f945b43 100644
--- a/Tests/RunCMake/VS10Project/RunCMakeTest.cmake
+++ b/Tests/RunCMake/VS10Project/RunCMakeTest.cmake
@@ -64,6 +64,21 @@ if (RunCMake_GENERATOR MATCHES "Visual Studio 1[0-4] 201[0-5]" OR
 else()
   run_cmake(UnityBuildNative)
   run_cmake(UnityBuildNativeGrouped)
+
+  function(run_UnityBuildPCH)
+    set(RunCMake_TEST_BINARY_DIR ${RunCMake_BINARY_DIR}/UnityBuildPCH-build)
+    run_cmake(UnityBuildPCH)
+    set(RunCMake_TEST_NO_CLEAN 1)
+    set(vcxproj "${RunCMake_TEST_BINARY_DIR}/UnityBuildPCH.vcxproj")
+    if(EXISTS "${vcxproj}")
+      file(STRINGS ${vcxproj} vcxproj_strings REGEX "ClCompile[^\n]*UnityBuildPCH\\.c")
+    endif()
+    if(vcxproj_strings MATCHES "Include=\"([^\"]+)\"")
+      set(src "${CMAKE_MATCH_1}")
+      run_cmake_command(UnityBuildPCH-build ${CMAKE_COMMAND} --build . --config Debug --target UnityBuildPCH -- -t:ClCompile -p:SelectedFiles=${src})
+    endif()
+  endfunction()
+  run_UnityBuildPCH()
 endif()
 
 run_cmake(VsDotnetTargetFramework)
diff --git a/Tests/RunCMake/VS10Project/UnityBuildPCH-build-check.cmake b/Tests/RunCMake/VS10Project/UnityBuildPCH-build-check.cmake
new file mode 100644
index 0000000..9043bb7
--- /dev/null
+++ b/Tests/RunCMake/VS10Project/UnityBuildPCH-build-check.cmake
@@ -0,0 +1,10 @@
+set(obj "${RunCMake_TEST_BINARY_DIR}/UnityBuildPCH.dir/Debug/UnityBuildPCH.obj")
+if(NOT EXISTS "${obj}")
+  set(RunCMake_TEST_FAILED "Expected object file does not exist:\n  ${obj}")
+  return()
+endif()
+set(lib "${RunCMake_TEST_BINARY_DIR}/Debug/UnityBuildPCH.lib")
+if(EXISTS "${lib}")
+  set(RunCMake_TEST_FAILED "Unexpected library file exists:\n  ${lib}")
+  return()
+endif()
diff --git a/Tests/RunCMake/VS10Project/UnityBuildPCH.c b/Tests/RunCMake/VS10Project/UnityBuildPCH.c
new file mode 100644
index 0000000..b96b068
--- /dev/null
+++ b/Tests/RunCMake/VS10Project/UnityBuildPCH.c
@@ -0,0 +1,4 @@
+int UnityBuildPCH(void)
+{
+  return 0;
+}
diff --git a/Tests/RunCMake/VS10Project/UnityBuildPCH.cmake b/Tests/RunCMake/VS10Project/UnityBuildPCH.cmake
new file mode 100644
index 0000000..875ffec
--- /dev/null
+++ b/Tests/RunCMake/VS10Project/UnityBuildPCH.cmake
@@ -0,0 +1,4 @@
+enable_language(C)
+add_library(UnityBuildPCH STATIC UnityBuildPCH.c)
+target_precompile_headers(UnityBuildPCH PRIVATE UnityBuildPCH.h)
+set_property(TARGET UnityBuildPCH PROPERTY UNITY_BUILD ON)
diff --git a/Tests/RunCMake/VS10Project/UnityBuildPCH.h b/Tests/RunCMake/VS10Project/UnityBuildPCH.h
new file mode 100644
index 0000000..fa882cb
--- /dev/null
+++ b/Tests/RunCMake/VS10Project/UnityBuildPCH.h
@@ -0,0 +1 @@
+/* empty file */
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetCheckProperty.cmake b/Tests/RunCMake/target_sources/AddCustomTargetCheckProperty.cmake
index 1787e87..1787e87 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetCheckProperty.cmake
+++ b/Tests/RunCMake/target_sources/AddCustomTargetCheckProperty.cmake
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetGenx.cmake b/Tests/RunCMake/target_sources/AddCustomTargetGenx.cmake
index 0078eab..0078eab 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetGenx.cmake
+++ b/Tests/RunCMake/target_sources/AddCustomTargetGenx.cmake
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetInterfaceSources-result.txt b/Tests/RunCMake/target_sources/AddCustomTargetInterfaceSources-result.txt
index d00491f..d00491f 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetInterfaceSources-result.txt
+++ b/Tests/RunCMake/target_sources/AddCustomTargetInterfaceSources-result.txt
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetInterfaceSources-stderr.txt b/Tests/RunCMake/target_sources/AddCustomTargetInterfaceSources-stderr.txt
index 9334bf6..9334bf6 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetInterfaceSources-stderr.txt
+++ b/Tests/RunCMake/target_sources/AddCustomTargetInterfaceSources-stderr.txt
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetInterfaceSources.cmake b/Tests/RunCMake/target_sources/AddCustomTargetInterfaceSources.cmake
index 42a8ca2..42a8ca2 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetInterfaceSources.cmake
+++ b/Tests/RunCMake/target_sources/AddCustomTargetInterfaceSources.cmake
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetPrivateSources.cmake b/Tests/RunCMake/target_sources/AddCustomTargetPrivateSources.cmake
index 11f0258..11f0258 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetPrivateSources.cmake
+++ b/Tests/RunCMake/target_sources/AddCustomTargetPrivateSources.cmake
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetPublicSources-result.txt b/Tests/RunCMake/target_sources/AddCustomTargetPublicSources-result.txt
index d00491f..d00491f 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetPublicSources-result.txt
+++ b/Tests/RunCMake/target_sources/AddCustomTargetPublicSources-result.txt
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetPublicSources-stderr.txt b/Tests/RunCMake/target_sources/AddCustomTargetPublicSources-stderr.txt
index afba4be..afba4be 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetPublicSources-stderr.txt
+++ b/Tests/RunCMake/target_sources/AddCustomTargetPublicSources-stderr.txt
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetPublicSources.cmake b/Tests/RunCMake/target_sources/AddCustomTargetPublicSources.cmake
index d9e82c0..d9e82c0 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetPublicSources.cmake
+++ b/Tests/RunCMake/target_sources/AddCustomTargetPublicSources.cmake
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetSources-result.txt b/Tests/RunCMake/target_sources/AddCustomTargetSources-result.txt
index d00491f..d00491f 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetSources-result.txt
+++ b/Tests/RunCMake/target_sources/AddCustomTargetSources-result.txt
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetSources-stderr.txt b/Tests/RunCMake/target_sources/AddCustomTargetSources-stderr.txt
index 4a153e9..4a153e9 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetSources-stderr.txt
+++ b/Tests/RunCMake/target_sources/AddCustomTargetSources-stderr.txt
diff --git a/Tests/RunCMake/TargetSources/AddCustomTargetSources.cmake b/Tests/RunCMake/target_sources/AddCustomTargetSources.cmake
index dd688d3..dd688d3 100644
--- a/Tests/RunCMake/TargetSources/AddCustomTargetSources.cmake
+++ b/Tests/RunCMake/target_sources/AddCustomTargetSources.cmake
diff --git a/Tests/RunCMake/TargetSources/CMP0026-LOCATION-result.txt b/Tests/RunCMake/target_sources/CMP0026-LOCATION-result.txt
index 573541a..573541a 100644
--- a/Tests/RunCMake/TargetSources/CMP0026-LOCATION-result.txt
+++ b/Tests/RunCMake/target_sources/CMP0026-LOCATION-result.txt
diff --git a/Tests/RunCMake/TargetSources/CMP0026-LOCATION-stderr.txt b/Tests/RunCMake/target_sources/CMP0026-LOCATION-stderr.txt
index d7ccedb..d7ccedb 100644
--- a/Tests/RunCMake/TargetSources/CMP0026-LOCATION-stderr.txt
+++ b/Tests/RunCMake/target_sources/CMP0026-LOCATION-stderr.txt
diff --git a/Tests/RunCMake/TargetSources/CMP0026-LOCATION.cmake b/Tests/RunCMake/target_sources/CMP0026-LOCATION.cmake
index 464df36..642856c 100644
--- a/Tests/RunCMake/TargetSources/CMP0026-LOCATION.cmake
+++ b/Tests/RunCMake/target_sources/CMP0026-LOCATION.cmake
@@ -1,5 +1,6 @@
 
 cmake_policy(SET CMP0026 OLD)
+enable_language(CXX)
 
 add_library(objlib OBJECT
     empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/CMP0076-OLD-result.txt b/Tests/RunCMake/target_sources/CMP0076-OLD-result.txt
index d00491f..d00491f 100644
--- a/Tests/RunCMake/TargetSources/CMP0076-OLD-result.txt
+++ b/Tests/RunCMake/target_sources/CMP0076-OLD-result.txt
diff --git a/Tests/RunCMake/TargetSources/CMP0076-OLD-stderr.txt b/Tests/RunCMake/target_sources/CMP0076-OLD-stderr.txt
index d47dd4d..d47dd4d 100644
--- a/Tests/RunCMake/TargetSources/CMP0076-OLD-stderr.txt
+++ b/Tests/RunCMake/target_sources/CMP0076-OLD-stderr.txt
diff --git a/Tests/RunCMake/TargetSources/CMP0076-OLD.cmake b/Tests/RunCMake/target_sources/CMP0076-OLD.cmake
index 4d8c268..4d8c268 100644
--- a/Tests/RunCMake/TargetSources/CMP0076-OLD.cmake
+++ b/Tests/RunCMake/target_sources/CMP0076-OLD.cmake
diff --git a/Tests/RunCMake/TargetSources/CMP0076-WARN-result.txt b/Tests/RunCMake/target_sources/CMP0076-WARN-result.txt
index d00491f..d00491f 100644
--- a/Tests/RunCMake/TargetSources/CMP0076-WARN-result.txt
+++ b/Tests/RunCMake/target_sources/CMP0076-WARN-result.txt
diff --git a/Tests/RunCMake/TargetSources/CMP0076-WARN-stderr.txt b/Tests/RunCMake/target_sources/CMP0076-WARN-stderr.txt
index bd888ee..bd888ee 100644
--- a/Tests/RunCMake/TargetSources/CMP0076-WARN-stderr.txt
+++ b/Tests/RunCMake/target_sources/CMP0076-WARN-stderr.txt
diff --git a/Tests/RunCMake/TargetSources/CMP0076-WARN.cmake b/Tests/RunCMake/target_sources/CMP0076-WARN.cmake
index 2e07331..2e07331 100644
--- a/Tests/RunCMake/TargetSources/CMP0076-WARN.cmake
+++ b/Tests/RunCMake/target_sources/CMP0076-WARN.cmake
diff --git a/Tests/RunCMake/TargetSources/CMP0076-WARN/CMakeLists.txt b/Tests/RunCMake/target_sources/CMP0076-WARN/CMakeLists.txt
index f9c7d6d..f9c7d6d 100644
--- a/Tests/RunCMake/TargetSources/CMP0076-WARN/CMakeLists.txt
+++ b/Tests/RunCMake/target_sources/CMP0076-WARN/CMakeLists.txt
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/subdir_empty_1.cpp b/Tests/RunCMake/target_sources/CMP0076-WARN/subdir_empty_1.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/subdir_empty_1.cpp
+++ b/Tests/RunCMake/target_sources/CMP0076-WARN/subdir_empty_1.cpp
diff --git a/Tests/RunCMake/target_sources/CMakeLists.txt b/Tests/RunCMake/target_sources/CMakeLists.txt
index 14ef56e..727f93a 100644
--- a/Tests/RunCMake/target_sources/CMakeLists.txt
+++ b/Tests/RunCMake/target_sources/CMakeLists.txt
@@ -1,5 +1,3 @@
 cmake_minimum_required(VERSION 3.11)
-
 project(${RunCMake_TEST} LANGUAGES NONE)
-
 include(${RunCMake_TEST}.cmake)
diff --git a/Tests/RunCMake/TargetSources/ConfigNotAllowed-result.txt b/Tests/RunCMake/target_sources/ConfigNotAllowed-result.txt
index d00491f..d00491f 100644
--- a/Tests/RunCMake/TargetSources/ConfigNotAllowed-result.txt
+++ b/Tests/RunCMake/target_sources/ConfigNotAllowed-result.txt
diff --git a/Tests/RunCMake/TargetSources/ConfigNotAllowed-stderr.txt b/Tests/RunCMake/target_sources/ConfigNotAllowed-stderr.txt
index c6b75fc..bc4afb7 100644
--- a/Tests/RunCMake/TargetSources/ConfigNotAllowed-stderr.txt
+++ b/Tests/RunCMake/target_sources/ConfigNotAllowed-stderr.txt
@@ -4,9 +4,9 @@ CMake Error in CMakeLists.txt:
 
   Config "Debug":
 
-    .*/Tests/RunCMake/TargetSources/empty_1.cpp
-    .*/Tests/RunCMake/TargetSources/empty_2.cpp
+    .*/Tests/RunCMake/target_sources/empty_1.cpp
+    .*/Tests/RunCMake/target_sources/empty_2.cpp
 
   Config "Release":
 
-    .*/Tests/RunCMake/TargetSources/empty_1.cpp
+    .*/Tests/RunCMake/target_sources/empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/ConfigNotAllowed.cmake b/Tests/RunCMake/target_sources/ConfigNotAllowed.cmake
index 02af379..02af379 100644
--- a/Tests/RunCMake/TargetSources/ConfigNotAllowed.cmake
+++ b/Tests/RunCMake/target_sources/ConfigNotAllowed.cmake
diff --git a/Tests/RunCMake/target_sources/empty_keyword_args.cmake b/Tests/RunCMake/target_sources/EmptyKeywordArgs.cmake
index 5cee451..5cee451 100644
--- a/Tests/RunCMake/target_sources/empty_keyword_args.cmake
+++ b/Tests/RunCMake/target_sources/EmptyKeywordArgs.cmake
diff --git a/Tests/RunCMake/TargetSources/ExportBuild-result.txt b/Tests/RunCMake/target_sources/ExportBuild-result.txt
index 573541a..573541a 100644
--- a/Tests/RunCMake/TargetSources/ExportBuild-result.txt
+++ b/Tests/RunCMake/target_sources/ExportBuild-result.txt
diff --git a/Tests/RunCMake/TargetSources/ExportBuild.cmake b/Tests/RunCMake/target_sources/ExportBuild.cmake
index b626aa6..b626aa6 100644
--- a/Tests/RunCMake/TargetSources/ExportBuild.cmake
+++ b/Tests/RunCMake/target_sources/ExportBuild.cmake
diff --git a/Tests/RunCMake/TargetSources/OriginDebug-result.txt b/Tests/RunCMake/target_sources/OriginDebug-result.txt
index 573541a..573541a 100644
--- a/Tests/RunCMake/TargetSources/OriginDebug-result.txt
+++ b/Tests/RunCMake/target_sources/OriginDebug-result.txt
diff --git a/Tests/RunCMake/TargetSources/OriginDebug-stderr.txt b/Tests/RunCMake/target_sources/OriginDebug-stderr.txt
index a40f463..502d5f1 100644
--- a/Tests/RunCMake/TargetSources/OriginDebug-stderr.txt
+++ b/Tests/RunCMake/target_sources/OriginDebug-stderr.txt
@@ -1,7 +1,7 @@
 CMake Debug Log at OriginDebug.cmake:13 \(add_library\):
   Used sources for target OriginDebug:
 
-   \* .*Tests/RunCMake/TargetSources/empty_2.cpp
+   \* .*Tests/RunCMake/target_sources/empty_2.cpp
 
 Call Stack \(most recent call first\):
   CMakeLists\.txt:[0-9]+ \(include\)
@@ -9,7 +9,7 @@ Call Stack \(most recent call first\):
 CMake Debug Log at OriginDebug.cmake:16 \(set_property\):
   Used sources for target OriginDebug:
 
-   \* .*Tests/RunCMake/TargetSources/empty_3.cpp
+   \* .*Tests/RunCMake/target_sources/empty_3.cpp
 
 Call Stack \(most recent call first\):
   CMakeLists\.txt:[0-9]+ \(include\)
@@ -17,7 +17,7 @@ Call Stack \(most recent call first\):
 CMake Debug Log at OriginDebug.cmake:20 \(target_sources\):
   Used sources for target OriginDebug:
 
-   \* .*Tests/RunCMake/TargetSources/empty_4.cpp
+   \* .*Tests/RunCMake/target_sources/empty_4.cpp
 
 Call Stack \(most recent call first\):
   CMakeLists\.txt:[0-9]+ \(include\)
@@ -25,7 +25,7 @@ Call Stack \(most recent call first\):
 CMake Debug Log at OriginDebug.cmake:14 \(target_link_libraries\):
   Used sources for target OriginDebug:
 
-   \* .*Tests/RunCMake/TargetSources/empty_1.cpp
+   \* .*Tests/RunCMake/target_sources/empty_1.cpp
 
 Call Stack \(most recent call first\):
   CMakeLists\.txt:[0-9]+ \(include\)
diff --git a/Tests/RunCMake/TargetSources/OriginDebug.cmake b/Tests/RunCMake/target_sources/OriginDebug.cmake
index d40a1d8..d40a1d8 100644
--- a/Tests/RunCMake/TargetSources/OriginDebug.cmake
+++ b/Tests/RunCMake/target_sources/OriginDebug.cmake
diff --git a/Tests/RunCMake/target_sources/RelativePathInInterface-stdout.txt b/Tests/RunCMake/target_sources/RelativePathInInterface-stdout.txt
new file mode 100644
index 0000000..19818b8
--- /dev/null
+++ b/Tests/RunCMake/target_sources/RelativePathInInterface-stdout.txt
@@ -0,0 +1 @@
+-- iface: .*Tests/RunCMake/target_sources/empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInInterface.cmake b/Tests/RunCMake/target_sources/RelativePathInInterface.cmake
index 0d3e9a4..25b22dd 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInInterface.cmake
+++ b/Tests/RunCMake/target_sources/RelativePathInInterface.cmake
@@ -1,4 +1,5 @@
 cmake_policy(SET CMP0076 NEW)
+enable_language(CXX)
 
 add_library(iface INTERFACE)
 target_sources(iface INTERFACE empty_1.cpp)
diff --git a/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx-stdout.txt b/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx-stdout.txt
new file mode 100644
index 0000000..a51a792
--- /dev/null
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx-stdout.txt
@@ -0,0 +1 @@
+-- genexlib: \$<1:.*Tests/RunCMake/target_sources/RelativePathInSubdirGenEx/subdir_empty_1.cpp>;\$<1:.*Tests/RunCMake/target_sources/RelativePathInSubdirGenEx/../empty_1.cpp>;\$<1:empty_2.cpp>
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx.cmake b/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx.cmake
index 1cdc2a7..9afcea5 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx.cmake
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx.cmake
@@ -1,4 +1,5 @@
 cmake_policy(SET CMP0076 NEW)
+enable_language(CXX)
 
 add_library(genexlib)
 add_subdirectory(RelativePathInSubdirGenEx)
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx/CMakeLists.txt b/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx/CMakeLists.txt
index 3bcf454..3bcf454 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx/CMakeLists.txt
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx/CMakeLists.txt
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/subdir_empty_1.cpp b/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx/subdir_empty_1.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/subdir_empty_1.cpp
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirGenEx/subdir_empty_1.cpp
diff --git a/Tests/RunCMake/target_sources/RelativePathInSubdirInclude-stdout.txt b/Tests/RunCMake/target_sources/RelativePathInSubdirInclude-stdout.txt
new file mode 100644
index 0000000..c42c88b
--- /dev/null
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInclude-stdout.txt
@@ -0,0 +1 @@
+-- privatelib: .*Tests/RunCMake/target_sources/RelativePathInSubdirInclude/subdir_empty_1.cpp;empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude.cmake b/Tests/RunCMake/target_sources/RelativePathInSubdirInclude.cmake
index 4acbeca..f5954c4 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude.cmake
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInclude.cmake
@@ -1,4 +1,5 @@
 cmake_policy(SET CMP0076 NEW)
+enable_language(CXX)
 
 add_library(privatelib)
 
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude/CMakeLists.txt b/Tests/RunCMake/target_sources/RelativePathInSubdirInclude/CMakeLists.txt
index 3dcb135..3dcb135 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude/CMakeLists.txt
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInclude/CMakeLists.txt
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude/subdir_empty_1.cpp b/Tests/RunCMake/target_sources/RelativePathInSubdirInclude/subdir_empty_1.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInclude/subdir_empty_1.cpp
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInclude/subdir_empty_1.cpp
diff --git a/Tests/RunCMake/target_sources/RelativePathInSubdirInterface-stdout.txt b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface-stdout.txt
new file mode 100644
index 0000000..ebbb29f
--- /dev/null
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface-stdout.txt
@@ -0,0 +1 @@
+-- iface: .*Tests/RunCMake/target_sources/RelativePathInSubdirInterface/subdir_empty_1.cpp;.*Tests/RunCMake/target_sources/RelativePathInSubdirInterface/subdir_empty_2.cpp;.*Tests/RunCMake/target_sources/RelativePathInSubdirInterface/../empty_1.cpp;.*Tests/RunCMake/target_sources/RelativePathInSubdirInterface/../empty_2.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface.cmake b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface.cmake
index 3652b4f..6a4e200 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface.cmake
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface.cmake
@@ -1,4 +1,5 @@
 cmake_policy(SET CMP0076 NEW)
+enable_language(CXX)
 
 add_library(iface INTERFACE)
 
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/CMakeLists.txt b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface/CMakeLists.txt
index 02e6966..02e6966 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/CMakeLists.txt
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface/CMakeLists.txt
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx/subdir_empty_1.cpp b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface/subdir_empty_1.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirGenEx/subdir_empty_1.cpp
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface/subdir_empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/subdir_empty_2.cpp b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface/subdir_empty_2.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/subdir_empty_2.cpp
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirInterface/subdir_empty_2.cpp
diff --git a/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate-stdout.txt b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate-stdout.txt
new file mode 100644
index 0000000..104f1de
--- /dev/null
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate-stdout.txt
@@ -0,0 +1 @@
+-- privatelib: .*Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/subdir_empty_1.cpp;.*Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/subdir_empty_2.cpp;.*Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/../empty_1.cpp;.*Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/../empty_2.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate.cmake b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate.cmake
index d0d3dc4..dd16e3f 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate.cmake
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate.cmake
@@ -1,4 +1,5 @@
 cmake_policy(SET CMP0076 NEW)
+enable_language(CXX)
 
 add_library(privatelib)
 
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/CMakeLists.txt b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/CMakeLists.txt
index 56ee853..56ee853 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirPrivate/CMakeLists.txt
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/CMakeLists.txt
diff --git a/Tests/RunCMake/TargetSources/CMP0076-WARN/subdir_empty_1.cpp b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/subdir_empty_1.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/CMP0076-WARN/subdir_empty_1.cpp
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/subdir_empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/subdir_empty_2.cpp b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/subdir_empty_2.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/RelativePathInSubdirInterface/subdir_empty_2.cpp
+++ b/Tests/RunCMake/target_sources/RelativePathInSubdirPrivate/subdir_empty_2.cpp
diff --git a/Tests/RunCMake/target_sources/RunCMakeTest.cmake b/Tests/RunCMake/target_sources/RunCMakeTest.cmake
index b67c598..9d64927 100644
--- a/Tests/RunCMake/target_sources/RunCMakeTest.cmake
+++ b/Tests/RunCMake/target_sources/RunCMakeTest.cmake
@@ -1,3 +1,23 @@
 include(RunCMake)
 
-run_cmake(empty_keyword_args)
+if(RunCMake_GENERATOR STREQUAL "Xcode")
+  run_cmake(ConfigNotAllowed)
+endif()
+
+run_cmake(EmptyKeywordArgs)
+run_cmake(OriginDebug)
+run_cmake(CMP0026-LOCATION)
+run_cmake(CMP0076-OLD)
+run_cmake(CMP0076-WARN)
+run_cmake(RelativePathInInterface)
+run_cmake(RelativePathInSubdirGenEx)
+run_cmake(RelativePathInSubdirInterface)
+run_cmake(RelativePathInSubdirPrivate)
+run_cmake(RelativePathInSubdirInclude)
+run_cmake(ExportBuild)
+run_cmake(AddCustomTargetPublicSources)
+run_cmake(AddCustomTargetPrivateSources)
+run_cmake(AddCustomTargetInterfaceSources)
+run_cmake(AddCustomTargetSources)
+run_cmake(AddCustomTargetCheckProperty)
+run_cmake(AddCustomTargetGenx)
diff --git a/Tests/RunCMake/TargetSources/empty_1.cpp b/Tests/RunCMake/target_sources/empty_1.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/empty_1.cpp
+++ b/Tests/RunCMake/target_sources/empty_1.cpp
diff --git a/Tests/RunCMake/TargetSources/empty_2.cpp b/Tests/RunCMake/target_sources/empty_2.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/empty_2.cpp
+++ b/Tests/RunCMake/target_sources/empty_2.cpp
diff --git a/Tests/RunCMake/TargetSources/empty_3.cpp b/Tests/RunCMake/target_sources/empty_3.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/empty_3.cpp
+++ b/Tests/RunCMake/target_sources/empty_3.cpp
diff --git a/Tests/RunCMake/TargetSources/empty_4.cpp b/Tests/RunCMake/target_sources/empty_4.cpp
index 11ec041..11ec041 100644
--- a/Tests/RunCMake/TargetSources/empty_4.cpp
+++ b/Tests/RunCMake/target_sources/empty_4.cpp
diff --git a/Tests/RunCMake/TargetSources/main.cpp b/Tests/RunCMake/target_sources/main.cpp
index 766b775..766b775 100644
--- a/Tests/RunCMake/TargetSources/main.cpp
+++ b/Tests/RunCMake/target_sources/main.cpp