Merge branch 'master' into merge-2.075

Conflicts:
	runtime/druntime

.circleci/config.yml

@@ -4,6 +4,7 @@ jobs:
     docker:
       - image: gcc
     environment:
+      - LLVM_VERSION: 5.0.0
       - HOST_LDC_VERSION: 1.3.0
     steps:
       - checkout
@@ -24,15 +25,22 @@ jobs:
             ninja --version
             gdb --version
             python -c "import lit; lit.main();" --version | head -n 1
-      - run:
-          name: Install LLVM nightly
-          command: |
-            add-apt-repository -y 'deb http://apt.llvm.org/unstable/ llvm-toolchain-5.0 main'
-            wget -O - http://apt.llvm.org/llvm-snapshot.gpg.key | apt-key add -
-            apt update
-            apt install -y llvm-5.0 llvm-5.0-dev libclang-common-5.0-dev libfuzzer-5.0-dev
       - restore_cache:
-          key: host-ldc-{{ .Environment.HOST_LDC_VERSION }}
+          keys:
+            - llvm-5.0.0
+            - host-ldc-{{ .Environment.HOST_LDC_VERSION }}
+      - run:
+          name: Install LLVM 5.0.0
+          command: |
+            if [[ ! -d llvm-$LLVM_VERSION ]]; then
+              wget -O llvm-$LLVM_VERSION.tar.xz http://releases.llvm.org/$LLVM_VERSION/clang+llvm-$LLVM_VERSION-linux-x86_64-ubuntu16.04.tar.xz
+              mkdir llvm-$LLVM_VERSION
+              tar -xpf llvm-$LLVM_VERSION.tar.xz --strip 1 -C llvm-$LLVM_VERSION
+            fi
+      - save_cache:
+          key: llvm-5.0.0
+          paths:
+            - llvm-5.0.0
       - run:
           name: Install LDC host compiler
           command: |
@@ -50,7 +58,7 @@ jobs:
             export HOST_LDMD=$PWD/ldc2-$HOST_LDC_VERSION-linux-x86_64/bin/ldmd2
             mkdir bootstrap
             cd bootstrap
-            cmake -G Ninja -DBUILD_SHARED_LIBS=OFF -DD_COMPILER=$HOST_LDMD ..
+            cmake -G Ninja -DLDC_WITH_LLD=OFF -DLLVM_ROOT_DIR=$PWD/../llvm-$LLVM_VERSION -DBUILD_SHARED_LIBS=OFF -DD_COMPILER=$HOST_LDMD ..
             ninja -j3
             bin/ldc2 -version
             cd ..
@@ -60,7 +68,7 @@ jobs:
             export HOST_LDMD=$PWD/bootstrap/bin/ldmd2
             mkdir build
             cd build
-            cmake -G Ninja -DLDC_INSTALL_LTOPLUGIN=ON -DLDC_INSTALL_LLVM_RUNTIME_LIBS=ON -DD_COMPILER=$HOST_LDMD ..
+            cmake -G Ninja -DLDC_WITH_LLD=OFF -DLLVM_ROOT_DIR=$PWD/../llvm-$LLVM_VERSION -DLDC_INSTALL_LTOPLUGIN=ON -DLDC_INSTALL_LLVM_RUNTIME_LIBS=ON -DD_COMPILER=$HOST_LDMD ..
             ninja -j3 all all-test-runners
             bin/ldc2 -version
             cd ..

.travis.yml

@@ -4,6 +4,9 @@ sudo: false
 
 matrix:
   include:
+    - os: linux
+      d: ldc
+      env: LLVM_VERSION=5.0.0 OPTS="-DLIB_SUFFIX=64 -DLDC_WITH_LLD=OFF"
     - os: linux
       d: ldc
       env: LLVM_VERSION=4.0.1 OPTS="-DLIB_SUFFIX=64"
@@ -18,7 +21,7 @@ matrix:
       env: LLVM_VERSION=3.7.1 OPTS="-DTEST_COVERAGE=ON"
     - os: osx
       d: ldc-beta
-      env: LLVM_VERSION=6.0.0-1 OPTS="-DBUILD_SHARED_LIBS=OFF" LLVM_SPIRV_AVAILABLE=ON
+      env: LLVM_VERSION=6.0.0-2 OPTS="-DBUILD_SHARED_LIBS=OFF" LLVM_SPIRV_AVAILABLE=ON
     - os: osx
       d: ldc
       env: LLVM_VERSION=4.0.0 OPTS="-DBUILD_SHARED_LIBS=ON"
@@ -27,7 +30,8 @@ matrix:
 
 cache:
   directories:
-    - llvm-spirv-6.0.0-1
+    - llvm-spirv-6.0.0-2
+    - llvm-5.0.0
     - llvm-4.0.1
     - llvm-4.0.0
     - llvm-3.9.1
@@ -44,7 +48,9 @@ addons:
 before_install:
   -
     if [ "${TRAVIS_OS_NAME}" = "linux" ]; then
-      if [ "${LLVM_VERSION}" = "4.0.1" ]; then
+      if [ "${LLVM_VERSION}" = "5.0.0" ]; then
+        export LLVM_ARCH="linux-x86_64-ubuntu14.04";
+      elif [ "${LLVM_VERSION}" = "4.0.1" ]; then
         export LLVM_ARCH="x86_64-linux-gnu-debian8";
       else
         export LLVM_ARCH="x86_64-linux-gnu-ubuntu-14.04";
@@ -54,7 +60,7 @@ before_install:
     fi;
     if [ -z "$(ls -A llvm-$LLVM_VERSION)" ]; then
       if [ -n "${LLVM_SPIRV_AVAILABLE}" ]; then
-        wget -O llvm-spirv-$LLVM_VERSION.tar.bz2 https://github.com/thewilsonator/llvm/releases/download/pre-intrinsics/LLVM-6.0.0git-7375e8c-Darwin.tar.bz2;
+        wget -O llvm-spirv-$LLVM_VERSION.tar.bz2 https://github.com/thewilsonator/llvm/releases/download/pre-intrinsics/LLVM-6.0.0svn-Darwin.tar.bz2;
         mkdir llvm-spirv-$LLVM_VERSION;
         tar -xjf llvm-spirv-$LLVM_VERSION.tar.bz2 --strip 1 -C llvm-spirv-$LLVM_VERSION;
       else

CMakeLists.txt

@@ -127,23 +127,27 @@ if(NOT MSVC_IDE)
 endif()
 
 if(MSVC)
-    if(${D_COMPILER_ID} STREQUAL "DigitalMars")
     if(CMAKE_SIZEOF_VOID_P EQUAL 8)
-            message(STATUS "Let DMD output 64bit object files")
+        message(STATUS "Let D host compiler output 64bit object files")
         append("-m64" DDMD_DFLAGS)
     else()
-            message(STATUS "Let DMD output 32bit COFF object files")
+        message(STATUS "Let D host compiler output 32bit COFF object files")
+        if(${D_COMPILER_ID} STREQUAL "DigitalMars")
             append("-m32mscoff" DDMD_DFLAGS)
+        else()
+            append("-m32" DDMD_DFLAGS)
+        endif()
     endif()
 
-        if(MSVC_VERSION GREATER 1800) # VS 2015+
+    if(${D_COMPILER_ID} STREQUAL "DigitalMars" AND (MSVC_VERSION GREATER 1800)) # VS 2015+
         append("-Llegacy_stdio_definitions.lib" DDMD_DFLAGS)
     endif()
-    endif()
 
     # Link against the static MSVC runtime; CMake's C(++) flags apparently default to the dynamic one.
     # Host DMD/LDMD already defaults to linking against the static MSVC runtime.
-    # NOTE: Requires LLVM to be built with CMake variable LLVM_USE_CRT_{RELEASE,DEBUG,...}=MT[d].
+    if(${LLVM_CXXFLAGS} MATCHES "(^| )/MDd?( |$)")
+        message(FATAL_ERROR "LLVM must be built with CMake option LLVM_USE_CRT_<CMAKE_BUILD_TYPE>=MT[d]")
+    endif()
     foreach(flag_var
             CMAKE_C_FLAGS CMAKE_C_FLAGS_DEBUG CMAKE_C_FLAGS_RELEASE CMAKE_C_FLAGS_MINSIZEREL CMAKE_C_FLAGS_RELWITHDEBINFO
             CMAKE_CXX_FLAGS CMAKE_CXX_FLAGS_DEBUG CMAKE_CXX_FLAGS_RELEASE CMAKE_CXX_FLAGS_MINSIZEREL CMAKE_CXX_FLAGS_RELWITHDEBINFO)
@@ -449,10 +453,9 @@ endif()
 #
 # LLD integration (requires LLVM >= 3.9 with LLD headers & libs)
 #
-set(LDC_WITH_LLD OFF)
+if(NOT DEFINED LDC_WITH_LLD)
-if(LDC_LLVM_VER GREATER 308)
+    if(LDC_LLVM_VER GREATER 308)
         # check for LLD header
-    unset(LDC_WITH_LLD)
         if(NOT MSVC)
             set(CMAKE_REQUIRED_FLAGS -std=c++11)
         endif()
@@ -460,10 +463,13 @@ if(LDC_LLVM_VER GREATER 308)
         CHECK_INCLUDE_FILE_CXX(lld/Driver/Driver.h LDC_WITH_LLD)
         unset(CMAKE_REQUIRED_FLAGS)
         unset(CMAKE_REQUIRED_INCLUDES)
-    if(LDC_WITH_LLD)
+    else()
+        set(LDC_WITH_LLD OFF)
+    endif()
+endif()
+if(LDC_WITH_LLD)
     message(STATUS "Building LDC with LLD support")
     append("-DLDC_WITH_LLD" LDC_CXXFLAGS)
-    endif()
 endif()
 
 #
@@ -614,6 +620,12 @@ else()
     if(MSVC)
         # Issue 1297 – set LDC's stack to 8 MiB like on Linux and Mac (default: 1 MiB).
         list(APPEND LDC_TRANSLATED_LINKER_FLAGS "-L/STACK:8388608")
+        # VS 2017+: Use undocumented /NOOPTTLS MS linker switch to keep on emitting
+        # a .tls section. Required for older host druntime versions, otherwise the
+        # GC TLS ranges are garbage starting with VS 2017 Update 15.3.
+        if(MSVC_VERSION GREATER 1900) # VS 2017+
+            list(APPEND LDC_TRANSLATED_LINKER_FLAGS "-L/NOOPTTLS")
+        endif()
     endif()
 endif()
 
@@ -764,13 +776,22 @@ function(copy_compilerrt_lib llvm_lib_name ldc_lib_name fixup_dylib)
     endif()
 endfunction()
 if (LDC_INSTALL_LLVM_RUNTIME_LIBS)
+    # Locate LLVM sanitizer runtime libraries, and copy them to our lib folder
+    # Note: libFuzzer is part of compiler-rt version >= 6.0, but was part of LLVM =< 5.0
+
     if(APPLE)
         copy_compilerrt_lib("darwin/libclang_rt.asan_osx_dynamic.dylib" "libldc_rt.asan_osx_dynamic.dylib" TRUE)
+        if(NOT (LDC_LLVM_VER LESS 600))
+            copy_compilerrt_lib("darwin/libclang_rt.fuzzer_osx.a" "libldc_rt.fuzzer_osx.a" FALSE)
+        endif()
     elseif(UNIX)
         copy_compilerrt_lib("linux/libclang_rt.asan-x86_64.a" "libldc_rt.asan-x86_64.a" FALSE)
+        if(NOT (LDC_LLVM_VER LESS 600))
+            copy_compilerrt_lib("linux/libclang_rt.fuzzer-x86_64.a" "libldc_rt.fuzzer-x86_64.a" FALSE)
+        endif()
     endif()
 
-    # Locate libFuzzer runtime library, and copy it to our lib folder
+    if(LDC_LLVM_VER LESS 600)
         set(LLVM_LIBFUZZER_PATH ${LLVM_LIBRARY_DIRS}/libFuzzer.a)
         if(EXISTS ${LLVM_LIBFUZZER_PATH})
             message(STATUS "Copying libFuzzer library: ${LLVM_LIBFUZZER_PATH} --> libFuzzer.a")
@@ -780,6 +801,7 @@ if (LDC_INSTALL_LLVM_RUNTIME_LIBS)
         else()
             message(STATUS "Not found: ${LLVM_LIBFUZZER_PATH}")
         endif()
+    endif()
 endif()
 
 #

README.md

@@ -1,6 +1,7 @@
 LDC – the LLVM-based D Compiler
 ===============================
 
+[![Build Status](https://circleci.com/gh/ldc-developers/ldc/tree/master.svg?style=svg)][6]
 [![Build Status](https://semaphoreci.com/api/v1/ldc-developers/ldc/branches/master/shields_badge.svg)][4]
 [![Build Status](https://travis-ci.org/ldc-developers/ldc.png?branch=master)][1]
 [![Build Status](https://ci.appveyor.com/api/projects/status/2cfhvg79782n4nth/branch/master?svg=true)][5]
@@ -95,3 +96,4 @@ Feedback of any kind is very much appreciated!
 [3]: https://www.bountysource.com/trackers/283332-ldc?utm_source=283332&utm_medium=shield&utm_campaign=TRACKER_BADGE "Bountysource"
 [4]: https://semaphoreci.com/ldc-developers/ldc "Semaphore CI Build Status"
 [5]: https://ci.appveyor.com/project/kinke/ldc/history "AppVeyor CI Build Status"
+[6]: https://circleci.com/gh/ldc-developers/ldc/tree/master "Circle CI Build Status"

appveyor.yml

@@ -14,7 +14,7 @@ skip_tags: true
 environment:
   matrix:
     - APPVEYOR_JOB_ARCH:   x64
-      APPVEYOR_BUILD_WORKER_IMAGE: Previous Visual Studio 2017
+      APPVEYOR_BUILD_WORKER_IMAGE: Visual Studio 2017
       D_COMPILER:          ldc
     - APPVEYOR_JOB_ARCH:   x86
       APPVEYOR_BUILD_WORKER_IMAGE: Visual Studio 2015

driver/archiver.cpp

@@ -201,10 +201,17 @@ int performWriteOperation(object::Archive *OldArchive,
       writeArchive(ArchiveName, NewMembers, Symtab, Kind, Deterministic, Thin,
                    std::move(OldArchiveBuf));
 
+#if LDC_LLVM_VER >= 600
+  if (Result) {
+    fail("error writing '" + ArchiveName + "': " + Result.message());
+    return 1;
+  }
+#else
   if (Result.second) {
     fail(Result.second, Result.first);
     return 1;
   }
+#endif
 
   return 0;
 }

driver/cl_options.cpp

@@ -519,6 +519,11 @@ cl::list<std::string>
                              " list. Use 'ocl-xy0' for OpenCL x.y, and "
                              "'cuda-xy0' for CUDA CC x.y"),
                      cl::value_desc("targets"));
+cl::opt<std::string>
+    dcomputeFilePrefix("mdcompute-file-prefix",
+                       cl::desc("Prefix to prepend to the generated kernel files."),
+                       cl::init("kernels"),
+                       cl::value_desc("prefix"));
 #endif
 
 static cl::extrahelp footer(

driver/cl_options.h

@@ -124,6 +124,7 @@ extern cl::opt<std::string> saveOptimizationRecord;
 #endif
 #if LDC_LLVM_SUPPORTED_TARGET_SPIRV || LDC_LLVM_SUPPORTED_TARGET_NVPTX
 extern cl::list<std::string> dcomputeTargets;
+extern cl::opt<std::string> dcomputeFilePrefix;
 #endif
 }
 #endif

driver/configfile.cpp

@@ -201,3 +201,19 @@ bool ConfigFile::read(const char *explicitConfFile, const char *section) {
 
   return readConfig(pathcstr, section, binpath.c_str());
 }
+
+void ConfigFile::extendCommandLine(llvm::SmallVectorImpl<const char *> &args) {
+  // insert 'switches' before all user switches
+  args.insert(args.begin() + 1, switches.begin(), switches.end());
+
+  // append 'post-switches', but before a first potential '-run'
+  size_t runIndex = 0;
+  for (size_t i = 1; i < args.size(); ++i) {
+    if (strcmp(args[i], "-run") == 0 || strcmp(args[i], "--run") == 0) {
+      runIndex = i;
+      break;
+    }
+  }
+  args.insert(runIndex == 0 ? args.end() : args.begin() + runIndex,
+              postSwitches.begin(), postSwitches.end());
+}

driver/configfile.d

@@ -12,6 +12,7 @@
 //===----------------------------------------------------------------------===//
 module driver.configfile;
 
+import ddmd.root.array;
 import driver.config;
 import core.stdc.stdio;
 import core.stdc.string;
@@ -29,16 +30,13 @@ string prepareBinDir(const(char)* binDir)
 }
 
 
-ArraySetting findSwitches(Setting s)
+ArraySetting findArraySetting(GroupSetting section, string name)
 {
-    auto grp = cast(GroupSetting)s;
+    if (!section) return null;
-    if (!grp) return null;
+    foreach (c; section.children)
-    foreach (c; grp.children)
     {
-        if (c.name == "switches")
+        if (c.type == Setting.Type.array && c.name == name)
-        {
+            return cast(ArraySetting) c;
-            return cast(ArraySetting)c;
-        }
     }
     return null;
 }
@@ -88,74 +86,78 @@ unittest
 }
 
 
-struct ConfigFile
+extern(C++) struct ConfigFile
 {
-public:
-
-    alias s_iterator = const(char)**;
-
 private:
 
     // representation
 
     const(char)* pathcstr;
-    s_iterator switches_b;
+    Array!(const(char)*) switches;
-    s_iterator switches_e;
+    Array!(const(char)*) postSwitches;
 
-    extern(C++)
+    bool readConfig(const(char)* cfPath, const(char)* sectionName, const(char)* binDir)
-    bool readConfig(const(char)* cfPath, const(char)* section, const(char)* binDir)
     {
+        switches.setDim(0);
+        postSwitches.setDim(0);
+
         immutable dBinDir = prepareBinDir(binDir);
-        const dSec = section[0 .. strlen(section)];
+        const dSec = sectionName[0 .. strlen(sectionName)];
 
         try
         {
-            auto settingSections = parseConfigFile(cfPath);
+            GroupSetting section, defaultSection;
-
+            foreach (s; parseConfigFile(cfPath))
-            bool sectionFound;
-            ArraySetting secSwitches;
-            ArraySetting defSwitches;
-
-            foreach (s; settingSections)
             {
+                if (s.type != Setting.Type.group)
+                    continue;
                 if (s.name == dSec)
-                {
+                    section = cast(GroupSetting) s;
-                    sectionFound = true;
-                    secSwitches = findSwitches(s);
-                }
                 else if (s.name == "default")
-                {
+                    defaultSection = cast(GroupSetting) s;
-                    sectionFound = true;
-                    defSwitches = findSwitches(s);
-                }
             }
 
-            if (!sectionFound)
+            if (!section && !defaultSection)
             {
                 const dCfPath = cfPath[0 .. strlen(cfPath)];
-                if (section)
+                if (sectionName)
                     throw new Exception("Could not look up section '" ~ cast(string) dSec
                                         ~ "' nor the 'default' section in " ~ cast(string) dCfPath);
                 else
                     throw new Exception("Could not look up 'default' section in " ~ cast(string) dCfPath);
             }
 
-            auto switches = secSwitches ? secSwitches : defSwitches;
+            ArraySetting findArray(string name)
-            if (!switches)
+            {
+                auto r = findArraySetting(section, name);
+                if (!r)
+                    r = findArraySetting(defaultSection, name);
+                return r;
+            }
+
+            auto switches = findArray("switches");
+            auto postSwitches = findArray("post-switches");
+            if (!switches && !postSwitches)
             {
                 const dCfPath = cfPath[0 .. strlen(cfPath)];
                 throw new Exception("Could not look up switches in " ~ cast(string) dCfPath);
             }
 
-            auto finalSwitches = new const(char)*[switches.vals.length];
+            void applyArray(ref Array!(const(char)*) output, ArraySetting input)
-            foreach (i, sw; switches.vals)
+            {
+                if (!input)
+                    return;
+
+                output.reserve(input.vals.length);
+                foreach (sw; input.vals)
                 {
                     const finalSwitch = sw.replace("%%ldcbinarypath%%", dBinDir) ~ '\0';
-                finalSwitches[i] = finalSwitch.ptr;
+                    output.push(finalSwitch.ptr);
+                }
             }
 
-            switches_b = finalSwitches.ptr;
+            applyArray(this.switches, switches);
-            switches_e = finalSwitches.ptr + finalSwitches.length;
+            applyArray(this.postSwitches, postSwitches);
 
             return true;
         }

driver/configfile.h

@@ -14,20 +14,19 @@
 #ifndef LDC_DRIVER_CONFIGFILE_H
 #define LDC_DRIVER_CONFIGFILE_H
 
+#include "llvm/ADT/SmallVector.h"
 #include <string>
-#include <vector>
+
+#include "array.h"
 
 class ConfigFile {
 public:
-  using s_iterator = const char **;
-
   bool read(const char *explicitConfFile, const char *section);
 
-  s_iterator switches_begin() { return switches_b; }
-  s_iterator switches_end() { return switches_e; }
-
   std::string path() { return std::string(pathcstr); }
 
+  void extendCommandLine(llvm::SmallVectorImpl<const char *> &args);
+
 private:
   bool locate(std::string &pathstr);
 
@@ -35,8 +34,8 @@ private:
   bool readConfig(const char *cfPath, const char *section, const char *binDir);
 
   const char *pathcstr = nullptr;
-  s_iterator switches_b = nullptr;
+  Array<const char *> switches;
-  s_iterator switches_e = nullptr;
+  Array<const char *> postSwitches;
 };
 
 #endif // LDC_DRIVER_CONFIGFILE_H

driver/dcomputecodegenerator.cpp

@@ -17,47 +17,67 @@
 #include <string>
 #include <algorithm>
 
+#if !(LDC_LLVM_SUPPORTED_TARGET_SPIRV || LDC_LLVM_SUPPORTED_TARGET_NVPTX)
+
+DComputeCodeGenManager::DComputeCodeGenManager(llvm::LLVMContext &c) : ctx(c) {}
+void DComputeCodeGenManager::emit(Module *) {}
+void DComputeCodeGenManager::writeModules() {}
+DComputeCodeGenManager::~DComputeCodeGenManager() {}
+
+#else
+
 DComputeTarget *
 DComputeCodeGenManager::createComputeTarget(const std::string &s) {
-  int v;
+#if LDC_LLVM_SUPPORTED_TARGET_SPIRV
 #define OCL_VALID_VER_INIT 100, 110, 120, 200, 210, 220
   const std::array<int, 6> valid_ocl_versions = {{OCL_VALID_VER_INIT}};
-#define CUDA_VALID_VER_INIT 100, 110, 120, 130, 200, 210, 300, 350, 370,\
- 500, 520, 600, 610, 620
-  const std::array<int, 14> vaild_cuda_versions = {{CUDA_VALID_VER_INIT}};
 
   if (s.substr(0, 4) == "ocl-") {
-    v = atoi(s.c_str() + 4);
+    const int v = atoi(s.c_str() + 4);
     if (std::find(valid_ocl_versions.begin(), valid_ocl_versions.end(), v) !=
         valid_ocl_versions.end()) {
       return createOCLTarget(ctx, v);
     }
-  } else if (s.substr(0, 5) == "cuda-") {
+  }
-    v = atoi(s.c_str() + 5);
+#endif
 
-    if (std::find(vaild_cuda_versions.begin(), vaild_cuda_versions.end(), v) !=
+#if LDC_LLVM_SUPPORTED_TARGET_NVPTX
-        vaild_cuda_versions.end()) {
+#define CUDA_VALID_VER_INIT 100, 110, 120, 130, 200, 210, 300, 350, 370,\
+ 500, 520, 600, 610, 620
+  const std::array<int, 14> valid_cuda_versions = {{CUDA_VALID_VER_INIT}};
+
+  if (s.substr(0, 5) == "cuda-") {
+    const int v = atoi(s.c_str() + 5);
+    if (std::find(valid_cuda_versions.begin(), valid_cuda_versions.end(), v) !=
+        valid_cuda_versions.end()) {
       return createCUDATarget(ctx, v);
     }
   }
+#endif
+
 #define XSTR(x) #x
 #define STR(x) XSTR((x))
 
   error(Loc(),
         "unrecognised or invalid DCompute targets: the format is ocl-xy0 "
-        "for OpenCl x.y and cuda-xy0 for CUDA CC x.y. Valid versions "
+        "for OpenCl x.y and cuda-xy0 for CUDA CC x.y."
-        "for OpenCl are " STR(OCL_VALID_VER_INIT) ". Valid versions for CUDA "
+#if LDC_LLVM_SUPPORTED_TARGET_SPIRV
-        "are " STR(CUDA_VALID_VER_INIT));
+        " Valid versions for OpenCl are " STR(OCL_VALID_VER_INIT) "."
+#endif
+#if LDC_LLVM_SUPPORTED_TARGET_NVPTX
+        " Valid versions for CUDA are " STR(CUDA_VALID_VER_INIT)
+#endif
+  );
   fatal();
   return nullptr;
 }
 
 DComputeCodeGenManager::DComputeCodeGenManager(llvm::LLVMContext &c) : ctx(c) {
-#if LDC_LLVM_SUPPORTED_TARGET_SPIRV || LDC_LLVM_SUPPORTED_TARGET_NVPTX
   for (auto &option : opts::dcomputeTargets) {
     targets.push_back(createComputeTarget(option));
   }
-#endif
+  oldGIR = gIR;
+  oldGTargetMachine = gTargetMachine;
 }
 
 void DComputeCodeGenManager::emit(Module *m) {
@@ -72,3 +92,10 @@ void DComputeCodeGenManager::writeModules() {
     target->writeModule();
   }
 }
+
+DComputeCodeGenManager::~DComputeCodeGenManager() {
+  gIR = oldGIR;
+  gTargetMachine = oldGTargetMachine;
+}
+
+#endif // LDC_LLVM_SUPPORTED_TARGET_SPIRV || LDC_LLVM_SUPPORTED_TARGET_NVPTX

driver/dcomputecodegenerator.h

@@ -13,6 +13,10 @@
 #include "gen/dcompute/target.h"
 #include "llvm/ADT/SmallVector.h"
 
+namespace llvm {
+  class TargetMachine;
+}
+
 // gets run on modules marked @compute
 // All @compute D modules are emitted into one LLVM module once per target.
 class DComputeCodeGenManager {
@@ -20,12 +24,14 @@ class DComputeCodeGenManager {
   llvm::LLVMContext &ctx;
   llvm::SmallVector<DComputeTarget *, 2> targets;
   DComputeTarget *createComputeTarget(const std::string &s);
-
+  IRState *oldGIR = nullptr;
+  llvm::TargetMachine *oldGTargetMachine = nullptr;
 public:
   void emit(Module *m);
   void writeModules();
 
   DComputeCodeGenManager(llvm::LLVMContext &c);
+  ~DComputeCodeGenManager();
 };
 
 #endif

driver/linker-gcc.cpp

@@ -180,58 +180,38 @@ void ArgsBuilder::addLTOLinkFlags() {
 
 //////////////////////////////////////////////////////////////////////////////
 
-// Returns true on success.
-bool addDarwinASanLinkFlags(std::vector<std::string> &args) {
-  std::string searchPaths[] = {
-    exe_path::prependLibDir("libldc_rt.asan_osx_dynamic.dylib"),
-    exe_path::prependLibDir("libclang_rt.asan_osx_dynamic.dylib"),
-  };
-
-  for (const auto &filepath : searchPaths) {
-    if (llvm::sys::fs::exists(filepath)) {
-      args.push_back(filepath);
-
-      // Add @executable_path to rpath to support having the dylib copied with
-      // the executable.
-      args.push_back("-rpath");
-      args.push_back("@executable_path");
-
-      // Add the path to the resource dir to rpath to support using the dylib
-      // from the default location without copying.
-      args.push_back("-rpath");
-      args.push_back(llvm::sys::path::parent_path(filepath));
-
-      return true;
-    }
-  }
-
-  // We did not find the library.
-  return false;
-}
-
 // Returns the arch name as used in the compiler_rt libs.
 // FIXME: implement correctly for non-x86 platforms (e.g. ARM)
 llvm::StringRef getCompilerRTArchName() {
   return global.params.targetTriple->getArchName();
 }
 
-bool addUnixlikeASanLinkFlags(std::vector<std::string> &args) {
+// Returns the libname as full path and with arch suffix and extension.
+// For example, with name="libldc_rt.fuzzer", the returned string is
+// "libldc_rt.fuzzer_osx.a" on Darwin.
+std::string getFullCompilerRTLibPath(llvm::StringRef name,
+                                     bool sharedLibrary = false) {
+  if (global.params.targetTriple->isOSDarwin()) {
+    return exe_path::prependLibDir(
+        name + (sharedLibrary ? "_osx_dynamic.dylib" : "_osx.a"));
+  } else {
+    return exe_path::prependLibDir(name + "-" + getCompilerRTArchName() +
+                                   (sharedLibrary ? ".so" : ".a"));
+  }
+}
+
+void ArgsBuilder::addASanLinkFlags() {
   // Examples: "libclang_rt.asan-x86_64.a" or "libclang_rt.asan-arm.a" and
   // "libclang_rt.asan-x86_64.so"
 
-  auto arch = getCompilerRTArchName();
+  // TODO: let user choose to link with shared lib.
-
+  // In case of shared ASan, I think we also need to statically link with
-  // TODO: let user choose to link with shared lib. In case of shared ASan, I
+  // libclang_rt.asan-preinit-<arch>.a on Linux. On Darwin, the only option is
-  // think we also need to statically link with
+  // to use the shared library.
-  // libclang_rt.asan-preinit-<arch>.a
+  bool linkSharedASan = global.params.targetTriple->isOSDarwin();
-  bool linkSharedASan = false;
-  const char *extension = linkSharedASan ? ".so" : ".a";
-
   std::string searchPaths[] = {
-      exe_path::prependLibDir("libldc_rt.asan-" + llvm::Twine(arch) +
+      getFullCompilerRTLibPath("libldc_rt.asan", linkSharedASan),
-                              extension),
+      getFullCompilerRTLibPath("libclang_rt.asan", linkSharedASan),
-      exe_path::prependLibDir("libclang_rt.asan-" + llvm::Twine(arch) +
-                              extension),
   };
 
   for (const auto &filepath : searchPaths) {
@@ -239,39 +219,39 @@ bool addUnixlikeASanLinkFlags(std::vector<std::string> &args) {
       args.push_back(filepath);
 
       if (linkSharedASan) {
-        // TODO: add -rpath
+        // Add @executable_path to rpath to support having the shared lib copied
+        // with the executable.
+        args.push_back("-rpath");
+        args.push_back("@executable_path");
+
+        // Add the path to the resource dir to rpath to support using the shared
+        // lib from the default location without copying.
+        args.push_back("-rpath");
+        args.push_back(llvm::sys::path::parent_path(filepath));
       }
 
-      return true;
+      return;
     }
   }
 
-  // We did not find the library.
+  // When we reach here, we did not find the ASan library.
-  return false;
-}
-
-void ArgsBuilder::addASanLinkFlags() {
-  bool success = false;
-  if (global.params.targetTriple->isOSDarwin()) {
-    success = addDarwinASanLinkFlags(args);
-  } else {
-    success = addUnixlikeASanLinkFlags(args);
-  }
-
-  if (!success) {
   // Fallback, requires Clang. The asan library contains a versioned symbol
   // name and a linker error will happen when the LDC-LLVM and Clang-LLVM
   // versions don't match.
   args.push_back("-fsanitize=address");
-  }
 }
 
 // Adds all required link flags for -fsanitize=fuzzer when libFuzzer library is
 // found.
 void ArgsBuilder::addFuzzLinkFlags() {
   std::string searchPaths[] = {
+#if LDC_LLVM_VER >= 600
+    getFullCompilerRTLibPath("libldc_rt.fuzzer"),
+    getFullCompilerRTLibPath("libclang_rt.fuzzer"),
+#else
     exe_path::prependLibDir("libFuzzer.a"),
     exe_path::prependLibDir("libLLVMFuzzer.a"),
+#endif
   };
 
   for (const auto &filepath : searchPaths) {

driver/main.cpp

@@ -354,9 +354,7 @@ void parseCommandLine(int argc, char **argv, Strings &sourceFiles,
   // just ignore errors for now, they are still printed
   cfg_file.read(explicitConfFile, cfg_triple.c_str());
 
-  // insert switches from config file before all explicit ones
+  cfg_file.extendCommandLine(allArguments);
-  allArguments.insert(allArguments.begin() + 1, cfg_file.switches_begin(),
-                      cfg_file.switches_end());
 
   // finalize by expanding response files specified in config file
   expandResponseFiles(allocator, allArguments);
@@ -398,7 +396,7 @@ void parseCommandLine(int argc, char **argv, Strings &sourceFiles,
     fprintf(global.stdmsg, "binary    %s\n", exe_path::getExePath().c_str());
     fprintf(global.stdmsg, "version   %s (DMD %s, LLVM %s)\n",
             global.ldc_version, global.version, global.llvm_version);
-    const std::string &path = cfg_file.path();
+    const std::string path = cfg_file.path();
     if (!path.empty()) {
       fprintf(global.stdmsg, "config    %s (%s)\n", path.c_str(),
               cfg_triple.c_str());

gen/asm-x86.h

@@ -599,6 +599,7 @@ static const char *alternateMnemonics[N_AltMn] = {
 #define N Opr_NoType
 
 // D=dest, N=notype
+// clang-format off
 static AsmOpInfo asmOpInfo[N_AsmOpInfo] = {
     /* Op_Invalid   */ {},
     /* Op_Adjust    */ {{0, 0, 0}, 0, Clb_EAX /*just AX*/},
@@ -830,6 +831,7 @@ static AsmOpInfo asmOpInfo[N_AsmOpInfo] = {
     /// immediate does not contribute to size
     /// * Op_cmovCC    */  { D|rw,  mrw,  0,    1 } // ->dstsrc
 };
+// clang-format on
 
 #undef mri
 #undef mr
@@ -2062,14 +2064,14 @@ struct AsmProcessor {
         }
         regInfo[i].gccName = std::string(buf, p - buf);
         if ((i <= Reg_ST || i > Reg_ST7) && i != Reg_EFLAGS) {
-          regInfo[i].ident =
+          regInfo[i].ident = Identifier::idPool(regInfo[i].name.data(),
-              Identifier::idPool(regInfo[i].name.data(), regInfo[i].name.size());
+                                                regInfo[i].name.size());
         }
       }
 
       for (int i = 0; i < N_PtrNames; i++) {
-        ptrTypeIdentTable[i] = Identifier::idPool(ptrTypeNameTable[i],
+        ptrTypeIdentTable[i] = Identifier::idPool(
-                                                  std::strlen(ptrTypeNameTable[i]));
+            ptrTypeNameTable[i], std::strlen(ptrTypeNameTable[i]));
       }
 
       Handled = createExpression(Loc(), TOKvoid, sizeof(Expression));
@@ -3398,48 +3400,38 @@ struct AsmProcessor {
   }
 
   Expression *parseEqualExp() {
-    Expression *exp = parseRelExp();
+    const auto exp = parseRelExp();
-    while (1) {
+    const auto tok = token->value;
-      switch (token->value) {
+    if (tok != TOKequal && tok != TOKnotequal) {
-      case TOKequal:
-      case TOKnotequal: {
-        TOK tok = token->value;
-        nextToken();
-        Expression *exp2 = parseRelExp();
-        if (isIntExp(exp) && isIntExp(exp2)) {
-          exp = intOp(tok, exp, exp2);
-        } else {
-          stmt->error("bad integral operand");
-        }
-      }
-      default:
       return exp;
     }
+
+    nextToken();
+    const auto exp2 = parseRelExp();
+    if (isIntExp(exp) && isIntExp(exp2)) {
+      return intOp(tok, exp, exp2);
     }
+
+    stmt->error("bad integral operand");
+    // TODO: Return ErrorExp?
     return exp;
   }
 
   Expression *parseRelExp() {
-    Expression *exp = parseShiftExp();
+    const auto exp = parseShiftExp();
-    while (1) {
+    const auto tok = token->value;
-      switch (token->value) {
+    if (tok != TOKgt && tok != TOKge && tok != TOKlt && tok != TOKle) {
-      case TOKgt:
-      case TOKge:
-      case TOKlt:
-      case TOKle: {
-        TOK tok = token->value;
-        nextToken();
-        Expression *exp2 = parseShiftExp();
-        if (isIntExp(exp) && isIntExp(exp2)) {
-          exp = intOp(tok, exp, exp2);
-        } else {
-          stmt->error("bad integral operand");
-        }
-      }
-      default:
       return exp;
     }
+
+    nextToken();
+    const auto exp2 = parseShiftExp();
+    if (isIntExp(exp) && isIntExp(exp2)) {
+      return intOp(tok, exp, exp2);
     }
+
+    stmt->error("bad integral operand");
+    // TODO: Return ErrorExp?
     return exp;
   }
 
@@ -3975,13 +3967,6 @@ struct AsmProcessor {
   }
 };
 
-#if D_GCC_VER < 40
-// struct rtx was modified for c++; this macro from rtl.h needs to
-// be modified accordingly.
-#undef XEXP
-#define XEXP(RTX, N) (RTL_CHECK2(RTX, N, 'e', 'u').rt_rtx)
-#endif
-
 // FIXME
 #define HOST_WIDE_INT long
 bool getFrameRelativeValue(LLValue *decl, HOST_WIDE_INT *result) {

gen/dcompute/target.cpp

@@ -7,19 +7,21 @@
 //
 //===----------------------------------------------------------------------===//
 
+#if LDC_LLVM_SUPPORTED_TARGET_SPIRV || LDC_LLVM_SUPPORTED_TARGET_NVPTX
+
 #include "ddmd/dsymbol.h"
 #include "ddmd/mars.h"
 #include "ddmd/module.h"
 #include "ddmd/scope.h"
 #include "driver/linker.h"
 #include "driver/toobj.h"
+#include "driver/cl_options.h"
 #include "gen/dcompute/target.h"
 #include "gen/llvmhelpers.h"
 #include "gen/runtime.h"
 #include <string>
 
 void DComputeTarget::doCodeGen(Module *m) {
-
   // process module members
   for (unsigned k = 0; k < m->members->dim; k++) {
     Dsymbol *dsym = (*m->members)[k];
@@ -45,7 +47,7 @@ void DComputeTarget::writeModule() {
 
   std::string filename;
   llvm::raw_string_ostream os(filename);
-  os << "kernels_" << short_name << tversion << '_'
+  os << opts::dcomputeFilePrefix << '_' << short_name << tversion << '_'
      << (global.params.is64bit ? 64 : 32) << '.' << binSuffix;
 
   const char *path = FileName::combine(global.params.objdir, os.str().c_str());
@@ -56,3 +58,5 @@ void DComputeTarget::writeModule() {
   delete _ir;
   _ir = nullptr;
 }
+
+#endif

gen/dcompute/target.h

@@ -9,10 +9,12 @@
 
 #ifndef LDC_GEN_DCOMPUTE_TARGET_H
 #define LDC_GEN_DCOMPUTE_TARGET_H
+
 #include "gen/irstate.h"
 #include "llvm/IR/LLVMContext.h"
 #include "llvm/IR/Function.h"
 #include <array>
+
 namespace llvm {
 class Module;
 class Function;
@@ -51,7 +53,12 @@ public:
   virtual void addKernelMetadata(FuncDeclaration *df, llvm::Function *llf) = 0;
 };
 
+#if LDC_LLVM_SUPPORTED_TARGET_NVPTX
 DComputeTarget *createCUDATarget(llvm::LLVMContext &c, int sm);
+#endif
+
+#if LDC_LLVM_SUPPORTED_TARGET_SPIRV
 DComputeTarget *createOCLTarget(llvm::LLVMContext &c, int oclver);
+#endif
 
 #endif

gen/dcompute/targetCUDA.cpp

@@ -7,6 +7,8 @@
 //
 //===----------------------------------------------------------------------===//
 
+#if LDC_LLVM_SUPPORTED_TARGET_NVPTX
+
 #include "gen/dcompute/target.h"
 #include "gen/dcompute/druntime.h"
 #include "gen/metadata.h"
@@ -77,3 +79,5 @@ public:
 DComputeTarget *createCUDATarget(llvm::LLVMContext &c, int sm) {
   return new TargetCUDA(c, sm);
 };
+
+#endif // LDC_LLVM_SUPPORTED_TARGET_NVPTX

gen/dcompute/targetOCL.cpp

@@ -7,6 +7,8 @@
 // See the LICENSE file for details.
 //===----------------------------------------------------------------------===//
 
+#if LDC_LLVM_SUPPORTED_TARGET_SPIRV
+
 #include "gen/dcompute/target.h"
 #include "gen/dcompute/druntime.h"
 #include "ddmd/template.h"
@@ -198,3 +200,5 @@ public:
 DComputeTarget *createOCLTarget(llvm::LLVMContext &c, int oclver) {
   return new TargetOCL(c, oclver);
 }
+
+#endif // LDC_LLVM_SUPPORTED_TARGET_SPIRV

gen/dibuilder.cpp

@@ -113,12 +113,26 @@ ldc::DIScope ldc::DIBuilder::GetCurrentScope() {
   return fn->diLexicalBlocks.top();
 }
 
-void ldc::DIBuilder::Declare(const Loc &loc, llvm::Value *var,
+// Sets the memory address for a debuginfo variable.
+void ldc::DIBuilder::Declare(const Loc &loc, llvm::Value *storage,
                              ldc::DILocalVariable divar,
                              ldc::DIExpression diexpr) {
   unsigned charnum = (loc.linnum ? loc.charnum : 0);
   auto debugLoc = llvm::DebugLoc::get(loc.linnum, charnum, GetCurrentScope());
-  DBuilder.insertDeclare(var, divar, diexpr, debugLoc, IR->scopebb());
+  DBuilder.insertDeclare(storage, divar, diexpr, debugLoc, IR->scopebb());
+}
+
+// Sets the (current) value for a debuginfo variable.
+void ldc::DIBuilder::SetValue(const Loc &loc, llvm::Value *value,
+                              ldc::DILocalVariable divar,
+                              ldc::DIExpression diexpr) {
+  unsigned charnum = (loc.linnum ? loc.charnum : 0);
+  auto debugLoc = llvm::DebugLoc::get(loc.linnum, charnum, GetCurrentScope());
+  DBuilder.insertDbgValueIntrinsic(value,
+#if LDC_LLVM_VER < 600
+                                   0,
+#endif
+                                   divar, diexpr, debugLoc, IR->scopebb());
 }
 
 ldc::DIFile ldc::DIBuilder::CreateFile(Loc &loc) {
@@ -1020,7 +1034,7 @@ void ldc::DIBuilder::EmitValue(llvm::Value *val, VarDeclaration *vd) {
 
 void ldc::DIBuilder::EmitLocalVariable(llvm::Value *ll, VarDeclaration *vd,
                                        Type *type, bool isThisPtr,
-                                       bool forceAsLocal,
+                                       bool forceAsLocal, bool isRefRVal,
                                        llvm::ArrayRef<int64_t> addr) {
   if (!mustEmitFullDebugInfo())
     return;
@@ -1040,11 +1054,27 @@ void ldc::DIBuilder::EmitLocalVariable(llvm::Value *ll, VarDeclaration *vd,
   if (static_cast<llvm::MDNode *>(TD) == nullptr)
     return; // unsupported
 
-  if (vd->isRef() || vd->isOut()) {
+  const bool isRefOrOut = vd->isRef() || vd->isOut(); // incl. special-ref vars
+
+  // For MSVC x64 targets, declare params rewritten by ExplicitByvalRewrite as
+  // DI references, as if they were ref parameters.
+  const bool isPassedExplicitlyByval =
+      isTargetMSVCx64 && !isRefOrOut && isaArgument(ll) && addr.empty();
+
+  bool useDbgValueIntrinsic = false;
+  if (isRefOrOut || isPassedExplicitlyByval) {
+    // With the exception of special-ref loop variables, the reference/pointer
+    // itself is constant. So we don't have to attach the debug information to a
+    // memory location and can use llvm.dbg.value to set the constant pointer
+    // for the DI reference.
+    useDbgValueIntrinsic =
+        isPassedExplicitlyByval || (!isSpecialRefVar(vd) && isRefRVal);
 #if LDC_LLVM_VER >= 308
-    auto T = DtoType(type);
+    // Note: createReferenceType expects the size to be the size of a pointer,
-    TD = DBuilder.createReferenceType(llvm::dwarf::DW_TAG_reference_type, TD,
+    // not the size of the type the reference refers to.
-                                      getTypeAllocSize(T) * 8, // size (bits)
+    TD = DBuilder.createReferenceType(
+        llvm::dwarf::DW_TAG_reference_type, TD,
+        gDataLayout->getPointerSizeInBits(), // size (bits)
         DtoAlignment(type) * 8);             // align (bits)
 #else
     TD = DBuilder.createReferenceType(llvm::dwarf::DW_TAG_reference_type, TD);
@@ -1122,10 +1152,15 @@ void ldc::DIBuilder::EmitLocalVariable(llvm::Value *ll, VarDeclaration *vd,
 #endif
   variableMap[vd] = debugVariable;
 
-  // declare
+  if (useDbgValueIntrinsic) {
-  Declare(vd->loc, ll, debugVariable, addr.empty()
+    SetValue(vd->loc, ll, debugVariable,
-                                          ? DBuilder.createExpression()
+             addr.empty() ? DBuilder.createExpression()
                           : DBuilder.createExpression(addr));
+  } else {
+    Declare(vd->loc, ll, debugVariable,
+            addr.empty() ? DBuilder.createExpression()
+                         : DBuilder.createExpression(addr));
+  }
 }
 
 void ldc::DIBuilder::EmitGlobalVariable(llvm::GlobalVariable *llVar,

gen/dibuilder.h

@@ -115,15 +115,22 @@ public:
 
   /// \brief Emits all things necessary for making debug info for a local
   /// variable vd.
-  /// \param ll       LL lvalue of the variable.
+  /// \param ll       LL value which, in combination with `addr`, yields the
+  /// storage/lvalue of the variable. For special-ref loop variables, specify
+  /// the storage/lvalue of the reference/pointer.
   /// \param vd       Variable declaration to emit debug info for.
   /// \param type     Type of variable if different from vd->type
   /// \param isThisPtr Variable is hidden this pointer
   /// \param forceAsLocal Emit as local even if the variable is a parameter
-  /// \param addr     An array of complex address operations.
+  /// \param isRefRVal Only relevant for ref/out parameters: indicates whether
+  /// ll & addr specify the reference's rvalue, i.e., the lvalue of the original
+  /// variable, instead of the reference's lvalue.
+  /// \param addr     An array of complex address operations encoding a DWARF
+  /// expression.
   void
   EmitLocalVariable(llvm::Value *ll, VarDeclaration *vd, Type *type = nullptr,
                     bool isThisPtr = false, bool forceAsLocal = false,
+                    bool isRefRVal = false,
                     llvm::ArrayRef<int64_t> addr = llvm::ArrayRef<int64_t>());
 
   /// \brief Emits all things necessary for making debug info for a global
@@ -139,7 +146,9 @@ private:
   llvm::LLVMContext &getContext();
   Module *getDefinedModule(Dsymbol *s);
   DIScope GetCurrentScope();
-  void Declare(const Loc &loc, llvm::Value *var, ldc::DILocalVariable divar,
+  void Declare(const Loc &loc, llvm::Value *storage, ldc::DILocalVariable divar,
+               ldc::DIExpression diexpr);
+  void SetValue(const Loc &loc, llvm::Value *value, ldc::DILocalVariable divar,
                 ldc::DIExpression diexpr);
   void AddFields(AggregateDeclaration *sd, ldc::DIFile file,
                  llvm::SmallVector<llvm::Metadata *, 16> &elems);

gen/functions.cpp

@@ -768,8 +768,14 @@ void defineParameters(IrFuncTy &irFty, VarDeclarations &parameters) {
       ++llArgIdx;
     }
 
-    if (global.params.symdebug)
+    // The debuginfos for captured params are handled later by
-      gIR->DBuilder.EmitLocalVariable(irparam->value, vd, paramType);
+    // DtoCreateNestedContext().
+    if (global.params.symdebug && vd->nestedrefs.dim == 0) {
+      // Reference (ref/out) parameters have no storage themselves as they are
+      // constant pointers, so pass the reference rvalue to EmitLocalVariable().
+      gIR->DBuilder.EmitLocalVariable(irparam->value, vd, paramType, false,
+                                      false, /*isRefRVal=*/true);
+    }
   }
 }
 

gen/nested.cpp

@@ -130,10 +130,13 @@ DValue *DtoNestedVariable(Loc &loc, Type *astype, VarDeclaration *vd,
     Logger::cout() << "of type: " << *val->getType() << '\n';
   }
   const bool isRefOrOut = vd->isRef() || vd->isOut();
-  if (!isSpecialRefVar(vd) && (byref || isRefOrOut)) {
+  if (isSpecialRefVar(vd)) {
+    // Handled appropriately by makeVarDValue() and EmitLocalVariable(), pass
+    // storage of pointer (reference lvalue).
+  } else if (byref || isRefOrOut) {
     val = DtoAlignedLoad(val);
-    // ref/out variables get a reference-debuginfo-type in
+    // ref/out variables get a reference-debuginfo-type in EmitLocalVariable();
-    // DIBuilder::EmitLocalVariable()
+    // pass the GEP as reference lvalue in that case.
     if (!isRefOrOut)
       gIR->DBuilder.OpDeref(dwarfAddrOps);
     IF_LOG {
@@ -143,11 +146,13 @@ DValue *DtoNestedVariable(Loc &loc, Type *astype, VarDeclaration *vd,
   }
 
   if (!skipDIDeclaration && global.params.symdebug) {
+#if LDC_LLVM_VER < 500
     // Because we are passing a GEP instead of an alloca to
     // llvm.dbg.declare, we have to make the address dereference explicit.
     gIR->DBuilder.OpDeref(dwarfAddrOps);
-    gIR->DBuilder.EmitLocalVariable(gep, vd, nullptr, false, true,
+#endif
-                                    dwarfAddrOps);
+    gIR->DBuilder.EmitLocalVariable(gep, vd, nullptr, false,
+                                    /*forceAsLocal=*/true, false, dwarfAddrOps);
   }
 
   return makeVarDValue(astype, vd, val);
@@ -472,6 +477,7 @@ void DtoCreateNestedContext(FuncGenState &funcGen) {
 
       IrLocal *irLocal = getIrLocal(vd);
       LLValue *gep = DtoGEPi(frame, 0, irLocal->nestedIndex, vd->toChars());
+      LLSmallVector<int64_t, 2> dwarfAddrOps;
       if (vd->isParameter()) {
         IF_LOG Logger::println("nested param: %s", vd->toChars());
         LOG_SCOPE
@@ -482,6 +488,7 @@ void DtoCreateNestedContext(FuncGenState &funcGen) {
         if (vd->isRef() || vd->isOut()) {
           Logger::println("Captured by reference, copying pointer to nested frame");
           DtoAlignedStore(parm->value, gep);
+          // pass GEP as reference lvalue to EmitLocalVariable()
         } else {
           Logger::println("Copying to nested frame");
           // The parameter value is an alloca'd stack slot.
@@ -498,11 +505,13 @@ void DtoCreateNestedContext(FuncGenState &funcGen) {
       }
 
       if (global.params.symdebug) {
-        LLSmallVector<int64_t, 2> addr;
+#if LDC_LLVM_VER < 500
         // Because we are passing a GEP instead of an alloca to
         // llvm.dbg.declare, we have to make the address dereference explicit.
-        gIR->DBuilder.OpDeref(addr);
+        gIR->DBuilder.OpDeref(dwarfAddrOps);
-        gIR->DBuilder.EmitLocalVariable(gep, vd, nullptr, false, false, addr);
+#endif
+        gIR->DBuilder.EmitLocalVariable(gep, vd, nullptr, false, false, false,
+                                        dwarfAddrOps);
       }
     }
   }

ldc2.conf.in

@@ -6,9 +6,12 @@ default:
 {
     // default switches injected before all explicit command-line switches
     switches = [
-        "-I@RUNTIME_DIR@/src",
+        "-I@RUNTIME_DIR@/src",@SHARED_LIBS_RPATH@
-        "-L-L@PROJECT_BINARY_DIR@/../lib@LIB_SUFFIX@", @MULTILIB_ADDITIONAL_PATH@@SHARED_LIBS_RPATH@
         "-defaultlib=druntime-ldc",
         "-debuglib=druntime-ldc-debug"@ADDITIONAL_DEFAULT_LDC_SWITCHES@
     ];
+    // default switches appended after all explicit command-line switches
+    post-switches = [
+        "-L-L@PROJECT_BINARY_DIR@/../lib@LIB_SUFFIX@",@MULTILIB_ADDITIONAL_PATH@
+    ];
 };

ldc2_install.conf.in

@@ -8,8 +8,11 @@ default:
     switches = [
         "-I@INCLUDE_INSTALL_DIR@/ldc",
         "-I@INCLUDE_INSTALL_DIR@",
-        "-L-L@CMAKE_INSTALL_LIBDIR@", @MULTILIB_ADDITIONAL_INSTALL_PATH@
         "-defaultlib=phobos2-ldc,druntime-ldc",
         "-debuglib=phobos2-ldc-debug,druntime-ldc-debug"@ADDITIONAL_DEFAULT_LDC_SWITCHES@
     ];
+    // default switches appended after all explicit command-line switches
+    post-switches = [
+        "-L-L@CMAKE_INSTALL_LIBDIR@",@MULTILIB_ADDITIONAL_INSTALL_PATH@
+    ];
 };

ldc2_phobos.conf.in

@@ -8,9 +8,12 @@ default:
     switches = [
         "-I@RUNTIME_DIR@/src",
         "-I@PROFILERT_DIR@/d",
-        "-I@PHOBOS2_DIR@",
+        "-I@PHOBOS2_DIR@",@SHARED_LIBS_RPATH@
-        "-L-L@CMAKE_BINARY_DIR@/lib@LIB_SUFFIX@", @MULTILIB_ADDITIONAL_PATH@@SHARED_LIBS_RPATH@
         "-defaultlib=phobos2-ldc,druntime-ldc",
         "-debuglib=phobos2-ldc-debug,druntime-ldc-debug"@ADDITIONAL_DEFAULT_LDC_SWITCHES@
     ];
+    // default switches appended after all explicit command-line switches
+    post-switches = [
+        "-L-L@CMAKE_BINARY_DIR@/lib@LIB_SUFFIX@",@MULTILIB_ADDITIONAL_PATH@
+    ];
 };

runtime/CMakeLists.txt

@@ -1,7 +1,11 @@
-project(runtime)
+project(runtime C)
 
 cmake_minimum_required(VERSION 2.8.9)
 
+set(CMAKE_MODULE_PATH ${CMAKE_MODULE_PATH} ${CMAKE_CURRENT_SOURCE_DIR})
+
+include(CheckIncludeFile)
+
 # Verify required variables if this CMake project is NOT embedded in the LDC CMake project.
 if(NOT LDC_EXE)
     if(NOT LDC_EXE_FULL)
@@ -37,14 +41,19 @@ set(D_FLAGS               -w                                  CACHE STRING "Runt
 set(D_FLAGS_DEBUG         -g;-link-debuglib                   CACHE STRING "Runtime D compiler flags (debug libraries), separated by ';'")
 set(D_FLAGS_RELEASE       -O3;-release                        CACHE STRING "Runtime D compiler flags (release libraries), separated by ';'")
 set(COMPILE_ALL_D_FILES_AT_ONCE ON                            CACHE BOOL   "Compile all D files for a lib in a single command line instead of separately")
+set(RT_ARCHIVE_WITH_LDC   AUTO                                CACHE STRING "Whether to archive the static runtime libs via LDC instead of CMake archiver")
 set(RT_CFLAGS             ""                                  CACHE STRING "Runtime extra C compiler flags, separated by ' '")
 set(LD_FLAGS              ""                                  CACHE STRING "Runtime extra C linker flags, separated by ' '")
 set(C_SYSTEM_LIBS         AUTO                                CACHE STRING "C system libraries for linking shared libraries and test runners, separated by ';'")
-set(TARGET_SYSTEM         AUTO                                CACHE STRING "Targeted platform for cross-compilation (e.g., 'Linux;UNIX', 'Darwin;APPLE;UNIX', 'Windows;MSVC')")
+set(TARGET_SYSTEM         AUTO                                CACHE STRING "Target OS/toolchain for cross-compilation (e.g., 'Linux;UNIX', 'Darwin;APPLE;UNIX', 'Windows;MSVC')")
+set(LDC_TARGET_PRESET     ""                                  CACHE STRING "Use a preset target configuration for cross-compilation, by passing format OS-arch (e.g., 'Linux-arm', 'Mac-x64', 'Windows-aarch64', 'Android-x86')")
 
 set(CMAKE_INSTALL_LIBDIR ${CMAKE_INSTALL_PREFIX}/lib${LIB_SUFFIX})
 
-# Auto-detect TARGET_SYSTEM
+# Use LDC_TARGET_PRESET to configure target runtime
+include(PresetRuntimeConfiguration)
+
+# Auto-detect TARGET_SYSTEM from host
 if("${TARGET_SYSTEM}" STREQUAL "AUTO")
     set(TARGET_SYSTEM ${CMAKE_SYSTEM_NAME})
     # Additionally add MSVC, APPLE and/or UNIX
@@ -75,8 +84,19 @@ else()
     set(MULTILIB_SUFFIX 64)
 endif()
 
+if(${RT_ARCHIVE_WITH_LDC} STREQUAL "AUTO")
+    # LLVM's MSVC archiver (v4.0.1) apparently has issues with objects with debuginfos,
+    # leading to MS linker warnings when linking against the static debug runtime libs.
+    if("${TARGET_SYSTEM}" MATCHES "MSVC")
+        set(RT_ARCHIVE_WITH_LDC OFF)
+    else()
+        set(RT_ARCHIVE_WITH_LDC ON)
+    endif()
+endif()
+
 set(SHARED_LIBS_SUPPORTED OFF)
-if("${TARGET_SYSTEM}" MATCHES "Linux|FreeBSD|APPLE")
+if("${TARGET_SYSTEM}" MATCHES "Linux|FreeBSD|APPLE"
+   AND (NOT "${TARGET_SYSTEM}" MATCHES "Android"))
     set(SHARED_LIBS_SUPPORTED ON)
 endif()
 
@@ -102,7 +122,9 @@ endif()
 
 # Auto-detect C system libraries
 if("${C_SYSTEM_LIBS}" STREQUAL "AUTO")
-    if("${TARGET_SYSTEM}" MATCHES "Linux")
+    if("${TARGET_SYSTEM}" MATCHES "Android")
+        set(C_SYSTEM_LIBS m c)
+    elseif("${TARGET_SYSTEM}" MATCHES "Linux")
         set(C_SYSTEM_LIBS m pthread rt dl)
     else()
         set(C_SYSTEM_LIBS m pthread)
@@ -293,7 +315,7 @@ foreach(target ${LLVM_TARGETS_TO_BUILD})
 endforeach()
 
 # Always build zlib and other C parts of the runtime in release mode, regardless
-# of what the user chose for LDC itself.
+# of what the user chose for LDC itself. Also add other C_FLAGS here.
 # 1) Set up CMAKE_C_FLAGS_RELEASE
 if("${TARGET_SYSTEM}" MATCHES "MSVC")
     # Omit all references to the default C runtime libs.
@@ -305,6 +327,11 @@ if("${TARGET_SYSTEM}" MATCHES "MSVC")
     # warning C4996: zlib uses 'deprecated' POSIX names
     append("/wd4131 /wd4206 /wd4996" CMAKE_C_FLAGS_RELEASE)
 endif()
+CHECK_INCLUDE_FILE(unistd.h HAVE_UNISTD_H)
+if (HAVE_UNISTD_H)
+    # Needed for zlib
+    append("-DHAVE_UNISTD_H" CMAKE_C_FLAGS_RELEASE)
+endif()
 # 2) Set all other CMAKE_C_FLAGS variants to CMAKE_C_FLAGS_RELEASE
 set(variables
     CMAKE_C_FLAGS_DEBUG
@@ -417,13 +444,24 @@ macro(compile_phobos2 d_flags lib_suffix path_suffix all_at_once outlist_o outli
     )
 endmacro()
 
+# Define an optional 'D' CMake linker language for the static runtime libs,
+# using LDC as archiver. LDC handles (cross-)archiving internally via LLVM
+# and supports LTO objects.
+set(CMAKE_D_CREATE_STATIC_LIBRARY "${LDC_EXE_FULL} -lib -of=<TARGET> <OBJECTS>")
+foreach(f ${D_FLAGS})
+    append("\"${f}\"" CMAKE_D_CREATE_STATIC_LIBRARY)
+endforeach()
+
 # Sets the common properties for all library targets.
-function(set_common_library_properties target)
+function(set_common_library_properties target is_shared)
     set_target_properties(${target} PROPERTIES
         VERSION ${DMDFE_VERSION}
         SOVERSION ${DMDFE_PATCH_VERSION}
         LINKER_LANGUAGE C
     )
+    if(RT_ARCHIVE_WITH_LDC AND NOT is_shared)
+        set_target_properties(${target} PROPERTIES LINKER_LANGUAGE D)
+    endif()
 
     # ldc2 defaults to position-independent code on Linux to match the implicit
     # linker default on Ubuntu 16.10 and above. As we might be building on an
@@ -458,7 +496,7 @@ macro(build_runtime_libs druntime_o druntime_bc phobos2_o phobos2_bc c_flags ld_
         COMPILE_FLAGS               "${c_flags}"
         LINK_FLAGS                  "${ld_flags}"
     )
-    set_common_library_properties(druntime-ldc${target_suffix})
+    set_common_library_properties(druntime-ldc${target_suffix} "${is_shared}")
 
     # When building a shared library, we need to link in all the default
     # libraries otherwise implicitly added by LDC to make it loadable from
@@ -481,7 +519,7 @@ macro(build_runtime_libs druntime_o druntime_bc phobos2_o phobos2_bc c_flags ld_
             COMPILE_FLAGS               "${c_flags}"
             LINK_FLAGS                  "${ld_flags}"
         )
-        set_common_library_properties(phobos2-ldc${target_suffix})
+        set_common_library_properties(phobos2-ldc${target_suffix} "${is_shared}")
 
         if("${is_shared}" STREQUAL "ON")
             target_link_libraries(phobos2-ldc${target_suffix}
@@ -595,7 +633,7 @@ macro(build_all_runtime_variants d_flags c_flags ld_flags path_suffix outlist_ta
         # static profile-rt
         build_profile_runtime("${d_flags};${D_FLAGS};${D_FLAGS_RELEASE}" "${c_flags}" "${ld_flags}" "" "${path_suffix}" ${outlist_targets})
         get_target_suffix("" "${path_suffix}" target_suffix)
-        set_common_library_properties(ldc-profile-rt${target_suffix})
+        set_common_library_properties(ldc-profile-rt${target_suffix} OFF)
     endif()
 endmacro()
 

runtime/PresetRuntimeConfiguration.cmake

@@ -0,0 +1,97 @@
+# - Preset cross-compilation configurations for C/ASM and D compilation
+# and linking, for supported targets
+#
+# This module sets compiler flags for a few C and assembly files in
+# DRuntime and Phobos, and linker flags to link the standard library as
+# a shared library and build the test runners, for various
+# cross-compilation targets that the LDC developers have tried out.
+#
+# It is enabled by setting LDC_TARGET_PRESET to a supported platform,
+# after which the appropriate TARGET_SYSTEM is set and a target triple
+# is appended to D_FLAGS.
+#
+# You can pass in custom RT_CFLAGS and LD_FLAGS of your choosing, but
+# if they're left unconfigured, they will also be set to sensible
+# defaults.
+
+if(NOT LDC_TARGET_PRESET STREQUAL "")
+    if(LDC_TARGET_PRESET MATCHES "Android")
+        set(ANDROID_API "21")
+    endif()
+    # This initial RT_CFLAGS/LD_FLAGS configuration for Android is a
+    # convenience for natively compiling, because CMake cannot detect
+    # Android as a separate platform from Linux.
+    if(RT_CFLAGS STREQUAL "" AND LDC_TARGET_PRESET MATCHES "Android")
+        set(RT_CFLAGS_UNCONFIGURED True)
+        set(RT_CFLAGS "-ffunction-sections -funwind-tables -fstack-protector-strong -Wno-invalid-command-line-argument -Wno-unused-command-line-argument -no-canonical-prefixes -g -DNDEBUG -DANDROID  -D__ANDROID_API__=${ANDROID_API} -Wa,--noexecstack -Wformat -Werror=format-security -fpie")
+
+        if(LDC_TARGET_PRESET MATCHES "arm")
+            append("-target armv7-none-linux-androideabi${ANDROID_API} -march=armv7-a -mfloat-abi=softfp -mfpu=vfpv3-d16 -mthumb -Os" RT_CFLAGS)
+        elseif(LDC_TARGET_PRESET MATCHES "aarch64")
+            append("-target aarch64-none-linux-android -O2" RT_CFLAGS)
+        endif()
+    endif()
+
+    if(LD_FLAGS STREQUAL "" AND LDC_TARGET_PRESET MATCHES "Android")
+        set(LD_FLAGS_UNCONFIGURED True)
+        set(LD_FLAGS "-Wl,--gc-sections -Wl,-z,nocopyreloc -no-canonical-prefixes -Wl,--no-undefined -Wl,-z,noexecstack -Wl,-z,relro -Wl,-z,now -Wl,--warn-shared-textrel -Wl,--fatal-warnings -fpie -pie")
+
+        if(LDC_TARGET_PRESET MATCHES "arm")
+            append("-target armv7-none-linux-androideabi${ANDROID_API} -Wl,--fix-cortex-a8" LD_FLAGS)
+        elseif(LDC_TARGET_PRESET MATCHES "aarch64")
+            append("-target aarch64-none-linux-android" LD_FLAGS)
+        endif()
+    endif()
+
+    if(LDC_TARGET_PRESET MATCHES "Windows")
+        set(TARGET_SYSTEM "Windows;MSVC")
+        if(LDC_TARGET_PRESET MATCHES "x64")
+            # stub example, fill in with the rest
+            list(APPEND D_FLAGS "-mtriple=x86_64-pc-windows-msvc")
+        endif()
+    elseif(LDC_TARGET_PRESET MATCHES "Android")
+        set(TARGET_SYSTEM "Android;Linux;UNIX")
+
+        # Check if we're using the NDK by looking for the toolchains
+        # directory in CC
+        if(CMAKE_C_COMPILER MATCHES "toolchains")
+            # Extract the NDK path and platform from CC
+            string(REGEX REPLACE ".toolchains.+" "" NDK_PATH ${CMAKE_C_COMPILER})
+            string(REGEX REPLACE ".+/prebuilt/([^/]+)/.+" "\\1" NDK_HOST_PLATFORM ${CMAKE_C_COMPILER})
+
+            if(LDC_TARGET_PRESET MATCHES "arm")
+                set(TARGET_ARCH "arm")
+                set(LLVM_TARGET_TRIPLE "armv7-none-linux-android")
+                set(TOOLCHAIN_TARGET_TRIPLE "arm-linux-androideabi")
+            elseif(LDC_TARGET_PRESET MATCHES "aarch64")
+                set(TARGET_ARCH "arm64")
+                set(LLVM_TARGET_TRIPLE "aarch64-none-linux-android")
+                set(TOOLCHAIN_TARGET_TRIPLE "aarch64-linux-android")
+            else()
+                message(FATAL_ERROR "Android platform ${LDC_TARGET_PRESET} is not supported.")
+            endif()
+            list(APPEND D_FLAGS "-mtriple=${LLVM_TARGET_TRIPLE}")
+            set(TOOLCHAIN_VERSION "4.9")
+
+            if(RT_CFLAGS_UNCONFIGURED)
+                append("-gcc-toolchain ${NDK_PATH}/toolchains/${TOOLCHAIN_TARGET_TRIPLE}-${TOOLCHAIN_VERSION}/prebuilt/${NDK_HOST_PLATFORM} --sysroot ${NDK_PATH}/sysroot -isystem ${NDK_PATH}/sysroot/usr/include/${TOOLCHAIN_TARGET_TRIPLE}" RT_CFLAGS)
+
+                if(LDC_TARGET_PRESET MATCHES "arm")
+                    append("-fno-integrated-as" RT_CFLAGS)
+                endif()
+            endif()
+
+            if(LD_FLAGS_UNCONFIGURED)
+                set(LD_BFD "-fuse-ld=bfd")
+                # work around Windows bug, android-ndk/ndk#75
+                if(NDK_HOST_PLATFORM MATCHES "windows")
+                    set(LD_BFD "${LD_BFD}.exe")
+                endif()
+
+                append("--sysroot=${NDK_PATH}/platforms/android-${ANDROID_API}/arch-${TARGET_ARCH} -gcc-toolchain ${NDK_PATH}/toolchains/${TOOLCHAIN_TARGET_TRIPLE}-${TOOLCHAIN_VERSION}/prebuilt/${NDK_HOST_PLATFORM} ${LD_BFD}" LD_FLAGS)
+            endif()
+        endif()
+    else()
+        message(FATAL_ERROR "LDC_TARGET_PRESET ${LDC_TARGET_PRESET} is not supported yet, pull requests to add common flags are welcome.")
+    endif()
+endif()

runtime/druntime

@@ -1 +1 @@
-Subproject commit 4e6daad2916745dcb426d4211d87fb76536c2c4f
+Subproject commit b96cff50fcdd944b0648d77170e8cd88b9743e24

runtime/ldc-build-runtime.d.in

@@ -12,6 +12,8 @@ struct Config {
     string ldcSourceDir;
     bool ninja;
     bool buildTestrunners;
+    string targetPreset;
+    string[] targetSystem;
     string[] dFlags;
     string[] cFlags;
     string[] linkerFlags;
@@ -117,10 +119,10 @@ void prepareLdcSource() {
     removeVersionSuffix("git-");
     removeVersionSuffix("-dirty");
 
-    const localArchiveFile = "ldc-src.tar.gz";
-    if (!localArchiveFile.exists) {
     import std.format : format;
-        const url = "https://github.com/ldc-developers/ldc/releases/download/v%1$s/ldc-%1$s-src.tar.gz".format(ldcVersion);
+    const localArchiveFile = "ldc-src.zip";
+    if (!localArchiveFile.exists) {
+        const url = "https://github.com/ldc-developers/ldc/releases/download/v%1$s/ldc-%1$s-src.zip".format(ldcVersion);
         writefln("Downloading LDC source archive: %s", url);
         import std.net.curl : download;
         download(url, localArchiveFile);
@@ -133,28 +135,36 @@ void prepareLdcSource() {
         }
     }
 
-    if (!ldcSrc.exists)
+    extractZipArchive(localArchiveFile, ".");
-        mkdir(ldcSrc);
+    rename("ldc-%1$s-src".format(ldcVersion), ldcSrc);
-
-    import std.array : split;
-    exec("tar -xzf ldc-src.tar.gz --strip 1 -C ldc-src".split);
 }
 
 void runCMake() {
-    import std.array : byPair, join;
+    import std.array : empty, byPair, join;
+    import std.regex : matchFirst;
 
     const wd = WorkingDirScope(config.buildDir);
 
+    if(config.dFlags.empty)
+        config.dFlags ~= "-w";
+
+    if(config.targetSystem.empty)
+        config.targetSystem ~= "AUTO";
+
     string[] args = [
         "cmake",
         "-DLDC_EXE_FULL=" ~ config.ldcExecutable,
         "-DD_VERSION=@D_VERSION@",
         "-DDMDFE_MINOR_VERSION=@DMDFE_MINOR_VERSION@",
         "-DDMDFE_PATCH_VERSION=@DMDFE_PATCH_VERSION@",
+        "-DLDC_TARGET_PRESET=" ~ config.targetPreset,
+        "-DTARGET_SYSTEM=" ~ config.targetSystem.join(";"),
         "-DD_FLAGS=" ~ config.dFlags.join(";"),
         "-DRT_CFLAGS=" ~ config.cFlags.join(" "),
         "-DLD_FLAGS=" ~ config.linkerFlags.join(" "),
     ];
+    if(config.targetPreset.matchFirst("^Android"))
+        args ~= ["-DCMAKE_SYSTEM_NAME=Linux", "-DCMAKE_C_COMPILER_WORKS=True"];
     foreach (pair; config.cmakeVars.byPair)
         args ~= "-D" ~ pair[0] ~ '=' ~ pair[1];
     if (config.ninja)
@@ -200,6 +210,23 @@ void exec(string[] command ...) {
     }
 }
 
+void extractZipArchive(string archivePath, string destination) {
+    import std.string : endsWith;
+    import std.zip;
+
+    auto archive = new ZipArchive(std.file.read(archivePath));
+    foreach (name, am; archive.directory) {
+        const destPath = buildNormalizedPath(destination, name);
+
+        const isDir = name.endsWith("/");
+        const destDir = isDir ? destPath : destPath.dirName;
+        mkdirRecurse(destDir);
+
+        if (!isDir)
+            std.file.write(destPath, archive.expand(am));
+    }
+}
+
 void parseCommandLine(string[] args) {
     import std.getopt : arraySep, getopt, defaultGetoptPrinter;
 
@@ -212,6 +239,8 @@ void parseCommandLine(string[] args) {
         "ldcSrcDir",   "Path to LDC source directory (if not specified: downloads & extracts source archive into '<buildDir>/ldc-src')", &config.ldcSourceDir,
         "ninja",       "Use Ninja as CMake build system", &config.ninja,
         "testrunners", "Build the testrunner executables too", &config.buildTestrunners,
+        "targetPreset","Target configuration preset by LDC devs, e.g. Android-arm", &config.targetPreset,
+        "targetSystem","Target OS/toolchain (separated by ';'), e.g. Windows;MSVC", &config.targetSystem,
         "dFlags",      "LDC flags for the D modules (separated by ';')", &config.dFlags,
         "cFlags",      "C/ASM compiler flags for the handful of C/ASM files (separated by ';')", &config.cFlags,
         "linkerFlags", "C linker flags for shared libraries and testrunner executables (separated by ';')", &config.linkerFlags,
@@ -237,7 +266,7 @@ void parseCommandLine(string[] args) {
             "  * CMake\n" ~
             "  * either Make or Ninja (recommended, enable with '--ninja')\n" ~
             "  * C toolchain (compiler and linker)\n" ~
-            "  * tar (can be worked around via '--ldcSrcDir=path/to/src')\n" ~
+            "--targetPreset currently supports Android-arm or Android-aarch64.\n" ~
             "All arguments are optional.\n" ~
             "CMake variables (see runtime/CMakeLists.txt in LDC source) can be specified via arguments like 'VAR=value'.\n",
             helpInformation.options

runtime/profile-rt/profile-rt-50/InstrProfData.inc

@@ -153,7 +153,17 @@ INSTR_PROF_RAW_HEADER(uint64_t, ValueKindLast, IPVK_Last)
 VALUE_PROF_FUNC_PARAM(uint64_t, TargetValue, Type::getInt64Ty(Ctx)) \
                       INSTR_PROF_COMMA
 VALUE_PROF_FUNC_PARAM(void *, Data, Type::getInt8PtrTy(Ctx)) INSTR_PROF_COMMA
+#ifndef VALUE_RANGE_PROF
 VALUE_PROF_FUNC_PARAM(uint32_t, CounterIndex, Type::getInt32Ty(Ctx))
+#else /* VALUE_RANGE_PROF */
+VALUE_PROF_FUNC_PARAM(uint32_t, CounterIndex, Type::getInt32Ty(Ctx)) \
+                      INSTR_PROF_COMMA
+VALUE_PROF_FUNC_PARAM(uint64_t, PreciseRangeStart, Type::getInt64Ty(Ctx)) \
+                      INSTR_PROF_COMMA
+VALUE_PROF_FUNC_PARAM(uint64_t, PreciseRangeLast, Type::getInt64Ty(Ctx)) \
+                      INSTR_PROF_COMMA
+VALUE_PROF_FUNC_PARAM(uint64_t, LargeValue, Type::getInt64Ty(Ctx))
+#endif /*VALUE_RANGE_PROF */
 #undef VALUE_PROF_FUNC_PARAM
 #undef INSTR_PROF_COMMA
 /* VALUE_PROF_FUNC_PARAM end */
@@ -174,13 +184,15 @@ VALUE_PROF_FUNC_PARAM(uint32_t, CounterIndex, Type::getInt32Ty(Ctx))
  * name hash and the function address.
  */
 VALUE_PROF_KIND(IPVK_IndirectCallTarget, 0)
+/* For memory intrinsic functions size profiling. */
+VALUE_PROF_KIND(IPVK_MemOPSize, 1)
 /* These two kinds must be the last to be
  * declared. This is to make sure the string
  * array created with the template can be
  * indexed with the kind value.
  */
 VALUE_PROF_KIND(IPVK_First, IPVK_IndirectCallTarget)
-VALUE_PROF_KIND(IPVK_Last, IPVK_IndirectCallTarget)
+VALUE_PROF_KIND(IPVK_Last, IPVK_MemOPSize)
 
 #undef VALUE_PROF_KIND
 /* VALUE_PROF_KIND end */
@@ -234,6 +246,31 @@ COVMAP_HEADER(uint32_t, Int32Ty, Version, \
 /* COVMAP_HEADER end.  */
 
 
+#ifdef INSTR_PROF_SECT_ENTRY
+#define INSTR_PROF_DATA_DEFINED
+INSTR_PROF_SECT_ENTRY(IPSK_data, \
+                      INSTR_PROF_QUOTE(INSTR_PROF_DATA_COMMON), \
+                      INSTR_PROF_QUOTE(INSTR_PROF_DATA_COFF), "__DATA,")
+INSTR_PROF_SECT_ENTRY(IPSK_cnts, \
+                      INSTR_PROF_QUOTE(INSTR_PROF_CNTS_COMMON), \
+                      INSTR_PROF_QUOTE(INSTR_PROF_CNTS_COFF), "__DATA,")
+INSTR_PROF_SECT_ENTRY(IPSK_name, \
+                      INSTR_PROF_QUOTE(INSTR_PROF_NAME_COMMON), \
+                      INSTR_PROF_QUOTE(INSTR_PROF_NAME_COFF), "__DATA,")
+INSTR_PROF_SECT_ENTRY(IPSK_vals, \
+                      INSTR_PROF_QUOTE(INSTR_PROF_VALS_COMMON), \
+                      INSTR_PROF_QUOTE(INSTR_PROF_VALS_COFF), "__DATA,")
+INSTR_PROF_SECT_ENTRY(IPSK_vnodes, \
+                      INSTR_PROF_QUOTE(INSTR_PROF_VNODES_COMMON), \
+                      INSTR_PROF_QUOTE(INSTR_PROF_VNODES_COFF), "__DATA,")
+INSTR_PROF_SECT_ENTRY(IPSK_covmap, \
+                      INSTR_PROF_QUOTE(INSTR_PROF_COVMAP_COMMON), \
+                      INSTR_PROF_QUOTE(INSTR_PROF_COVMAP_COFF), "__LLVM_COV,")
+
+#undef INSTR_PROF_SECT_ENTRY
+#endif
+
+
 #ifdef INSTR_PROF_VALUE_PROF_DATA
 #define INSTR_PROF_DATA_DEFINED
 
@@ -610,17 +647,47 @@ serializeValueProfDataFrom(ValueProfRecordClosure *Closure,
  * specified via command line. */
 #define INSTR_PROF_PROFILE_NAME_VAR __llvm_profile_filename
 
+/* section name strings common to all targets other
+   than WIN32 */
+#define INSTR_PROF_DATA_COMMON __llvm_prf_data
+#define INSTR_PROF_NAME_COMMON __llvm_prf_names
+#define INSTR_PROF_CNTS_COMMON __llvm_prf_cnts
+#define INSTR_PROF_VALS_COMMON __llvm_prf_vals
+#define INSTR_PROF_VNODES_COMMON __llvm_prf_vnds
+#define INSTR_PROF_COVMAP_COMMON __llvm_covmap
+/* Win32 */
+#define INSTR_PROF_DATA_COFF .lprfd
+#define INSTR_PROF_NAME_COFF .lprfn
+#define INSTR_PROF_CNTS_COFF .lprfc
+#define INSTR_PROF_VALS_COFF .lprfv
+#define INSTR_PROF_VNODES_COFF .lprfnd
+#define INSTR_PROF_COVMAP_COFF .lcovmap
+
+#ifdef _WIN32
 /* Runtime section names and name strings.  */
-#define INSTR_PROF_DATA_SECT_NAME __llvm_prf_data
+#define INSTR_PROF_DATA_SECT_NAME INSTR_PROF_DATA_COFF
-#define INSTR_PROF_NAME_SECT_NAME __llvm_prf_names
+#define INSTR_PROF_NAME_SECT_NAME INSTR_PROF_NAME_COFF
-#define INSTR_PROF_CNTS_SECT_NAME __llvm_prf_cnts
+#define INSTR_PROF_CNTS_SECT_NAME INSTR_PROF_CNTS_COFF
 /* Array of pointers. Each pointer points to a list
  * of value nodes associated with one value site.
  */
-#define INSTR_PROF_VALS_SECT_NAME __llvm_prf_vals
+#define INSTR_PROF_VALS_SECT_NAME INSTR_PROF_VALS_COFF
 /* Value profile nodes section. */
-#define INSTR_PROF_VNODES_SECT_NAME __llvm_prf_vnds
+#define INSTR_PROF_VNODES_SECT_NAME INSTR_PROF_VNODES_COFF
-#define INSTR_PROF_COVMAP_SECT_NAME __llvm_covmap
+#define INSTR_PROF_COVMAP_SECT_NAME INSTR_PROF_COVMAP_COFF
+#else
+/* Runtime section names and name strings.  */
+#define INSTR_PROF_DATA_SECT_NAME INSTR_PROF_DATA_COMMON
+#define INSTR_PROF_NAME_SECT_NAME INSTR_PROF_NAME_COMMON
+#define INSTR_PROF_CNTS_SECT_NAME INSTR_PROF_CNTS_COMMON
+/* Array of pointers. Each pointer points to a list
+ * of value nodes associated with one value site.
+ */
+#define INSTR_PROF_VALS_SECT_NAME INSTR_PROF_VALS_COMMON
+/* Value profile nodes section. */
+#define INSTR_PROF_VNODES_SECT_NAME INSTR_PROF_VNODES_COMMON
+#define INSTR_PROF_COVMAP_SECT_NAME INSTR_PROF_COVMAP_COMMON
+#endif
 
 #define INSTR_PROF_DATA_SECT_NAME_STR                                          \
   INSTR_PROF_QUOTE(INSTR_PROF_DATA_SECT_NAME)
@@ -649,6 +716,9 @@ serializeValueProfDataFrom(ValueProfRecordClosure *Closure,
 #define INSTR_PROF_VALUE_PROF_FUNC __llvm_profile_instrument_target
 #define INSTR_PROF_VALUE_PROF_FUNC_STR \
         INSTR_PROF_QUOTE(INSTR_PROF_VALUE_PROF_FUNC)
+#define INSTR_PROF_VALUE_RANGE_PROF_FUNC __llvm_profile_instrument_range
+#define INSTR_PROF_VALUE_RANGE_PROF_FUNC_STR \
+        INSTR_PROF_QUOTE(INSTR_PROF_VALUE_RANGE_PROF_FUNC)
 
 /* InstrProfile per-function control data alignment.  */
 #define INSTR_PROF_DATA_ALIGNMENT 8

runtime/profile-rt/profile-rt-50/InstrProfiling.c

@@ -19,8 +19,6 @@
 
 COMPILER_RT_WEAK uint64_t INSTR_PROF_RAW_VERSION_VAR = INSTR_PROF_RAW_VERSION;
 
-COMPILER_RT_WEAK char INSTR_PROF_PROFILE_NAME_VAR[1] = {0};
-
 COMPILER_RT_VISIBILITY uint64_t __llvm_profile_get_magic(void) {
   return sizeof(void *) == sizeof(uint64_t) ? (INSTR_PROF_RAW_MAGIC_64)
                                             : (INSTR_PROF_RAW_MAGIC_32);

runtime/profile-rt/profile-rt-50/InstrProfilingBuffer.c

@@ -45,15 +45,24 @@ uint64_t __llvm_profile_get_size_for_buffer_internal(
          (CountersEnd - CountersBegin) * sizeof(uint64_t) + NamesSize + Padding;
 }
 
+COMPILER_RT_VISIBILITY
+void initBufferWriter(ProfDataWriter *BufferWriter, char *Buffer) {
+  BufferWriter->Write = lprofBufferWriter;
+  BufferWriter->WriterCtx = Buffer;
+}
+
 COMPILER_RT_VISIBILITY int __llvm_profile_write_buffer(char *Buffer) {
-  return lprofWriteData(lprofBufferWriter, Buffer, 0);
+  ProfDataWriter BufferWriter;
+  initBufferWriter(&BufferWriter, Buffer);
+  return lprofWriteData(&BufferWriter, 0, 0);
 }
 
 COMPILER_RT_VISIBILITY int __llvm_profile_write_buffer_internal(
     char *Buffer, const __llvm_profile_data *DataBegin,
     const __llvm_profile_data *DataEnd, const uint64_t *CountersBegin,
     const uint64_t *CountersEnd, const char *NamesBegin, const char *NamesEnd) {
-  return lprofWriteDataImpl(lprofBufferWriter, Buffer, DataBegin, DataEnd,
+  ProfDataWriter BufferWriter;
-                            CountersBegin, CountersEnd, 0, NamesBegin,
+  initBufferWriter(&BufferWriter, Buffer);
-                            NamesEnd);
+  return lprofWriteDataImpl(&BufferWriter, DataBegin, DataEnd, CountersBegin,
+                            CountersEnd, 0, NamesBegin, NamesEnd, 0);
 }

runtime/profile-rt/profile-rt-50/InstrProfilingFile.c

@@ -91,24 +91,39 @@ static const char *getCurFilename(char *FilenameBuf);
 static unsigned doMerging() { return lprofCurFilename.MergePoolSize; }
 
 /* Return 1 if there is an error, otherwise return  0.  */
-static uint32_t fileWriter(ProfDataIOVec *IOVecs, uint32_t NumIOVecs,
+static uint32_t fileWriter(ProfDataWriter *This, ProfDataIOVec *IOVecs,
-                           void **WriterCtx) {
+                           uint32_t NumIOVecs) {
   uint32_t I;
-  FILE *File = (FILE *)*WriterCtx;
+  FILE *File = (FILE *)This->WriterCtx;
   for (I = 0; I < NumIOVecs; I++) {
+    if (IOVecs[I].Data) {
       if (fwrite(IOVecs[I].Data, IOVecs[I].ElmSize, IOVecs[I].NumElm, File) !=
           IOVecs[I].NumElm)
         return 1;
+    } else {
+      if (fseek(File, IOVecs[I].ElmSize * IOVecs[I].NumElm, SEEK_CUR) == -1)
+        return 1;
+    }
   }
   return 0;
 }
 
+static void initFileWriter(ProfDataWriter *This, FILE *File) {
+  This->Write = fileWriter;
+  This->WriterCtx = File;
+}
+
 COMPILER_RT_VISIBILITY ProfBufferIO *
 lprofCreateBufferIOInternal(void *File, uint32_t BufferSz) {
   FreeHook = &free;
   DynamicBufferIOBuffer = (uint8_t *)calloc(BufferSz, 1);
   VPBufferSize = BufferSz;
-  return lprofCreateBufferIO(fileWriter, File);
+  ProfDataWriter *fileWriter =
+      (ProfDataWriter *)calloc(sizeof(ProfDataWriter), 1);
+  initFileWriter(fileWriter, File);
+  ProfBufferIO *IO = lprofCreateBufferIO(fileWriter);
+  IO->OwnFileWriter = 1;
+  return IO;
 }
 
 static void setupIOBuffer() {
@@ -122,9 +137,10 @@ static void setupIOBuffer() {
 
 /* Read profile data in \c ProfileFile and merge with in-memory
    profile counters. Returns -1 if there is fatal error, otheriwse
-   0 is returned.
+   0 is returned. Returning 0 does not mean merge is actually
+   performed. If merge is actually done, *MergeDone is set to 1.
 */
-static int doProfileMerging(FILE *ProfileFile) {
+static int doProfileMerging(FILE *ProfileFile, int *MergeDone) {
   uint64_t ProfileFileSize;
   char *ProfileBuffer;
 
@@ -169,9 +185,21 @@ static int doProfileMerging(FILE *ProfileFile) {
   __llvm_profile_merge_from_buffer(ProfileBuffer, ProfileFileSize);
   (void)munmap(ProfileBuffer, ProfileFileSize);
 
+  *MergeDone = 1;
+
   return 0;
 }
 
+/* Create the directory holding the file, if needed. */
+static void createProfileDir(const char *Filename) {
+  size_t Length = strlen(Filename);
+  if (lprofFindFirstDirSeparator(Filename)) {
+    char *Copy = (char *)COMPILER_RT_ALLOCA(Length + 1);
+    strncpy(Copy, Filename, Length + 1);
+    __llvm_profile_recursive_mkdir(Copy);
+  }
+}
+
 /* Open the profile data for merging. It opens the file in r+b mode with
  * file locking.  If the file has content which is compatible with the
  * current process, it also reads in the profile data in the file and merge
@@ -180,23 +208,23 @@ static int doProfileMerging(FILE *ProfileFile) {
  * dumper. With profile merging enabled, each executable as well as any of
  * its instrumented shared libraries dump profile data into their own data file.
 */
-static FILE *openFileForMerging(const char *ProfileFileName) {
+static FILE *openFileForMerging(const char *ProfileFileName, int *MergeDone) {
   FILE *ProfileFile;
   int rc;
 
+  createProfileDir(ProfileFileName);
   ProfileFile = lprofOpenFileEx(ProfileFileName);
   if (!ProfileFile)
     return NULL;
 
-  rc = doProfileMerging(ProfileFile);
+  rc = doProfileMerging(ProfileFile, MergeDone);
-  if (rc || COMPILER_RT_FTRUNCATE(ProfileFile, 0L) ||
+  if (rc || (!*MergeDone && COMPILER_RT_FTRUNCATE(ProfileFile, 0L)) ||
       fseek(ProfileFile, 0L, SEEK_SET) == -1) {
     PROF_ERR("Profile Merging of file %s failed: %s\n", ProfileFileName,
              strerror(errno));
     fclose(ProfileFile);
     return NULL;
   }
-  fseek(ProfileFile, 0L, SEEK_SET);
   return ProfileFile;
 }
 
@@ -205,17 +233,20 @@ static int writeFile(const char *OutputName) {
   int RetVal;
   FILE *OutputFile;
 
+  int MergeDone = 0;
   if (!doMerging())
     OutputFile = fopen(OutputName, "ab");
   else
-    OutputFile = openFileForMerging(OutputName);
+    OutputFile = openFileForMerging(OutputName, &MergeDone);
 
   if (!OutputFile)
     return -1;
 
   FreeHook = &free;
   setupIOBuffer();
-  RetVal = lprofWriteData(fileWriter, OutputFile, lprofGetVPDataReader());
+  ProfDataWriter fileWriter;
+  initFileWriter(&fileWriter, OutputFile);
+  RetVal = lprofWriteData(&fileWriter, lprofGetVPDataReader(), MergeDone);
 
   fclose(OutputFile);
   return RetVal;
@@ -233,18 +264,13 @@ static void truncateCurrentFile(void) {
   if (!Filename)
     return;
 
-  /* Create the directory holding the file, if needed. */
-  if (lprofFindFirstDirSeparator(Filename)) {
-    char *Copy = (char *)COMPILER_RT_ALLOCA(Length + 1);
-    strncpy(Copy, Filename, Length + 1);
-    __llvm_profile_recursive_mkdir(Copy);
-  }
-
   /* By pass file truncation to allow online raw profile
    * merging. */
   if (lprofCurFilename.MergePoolSize)
     return;
 
+  createProfileDir(Filename);
+
   /* Truncate the file.  Later we'll reopen and append. */
   File = fopen(Filename, "w");
   if (!File)
@@ -524,6 +550,7 @@ int __llvm_profile_write_file(void) {
   int rc, Length;
   const char *Filename;
   char *FilenameBuf;
+  int PDeathSig = 0;
 
   if (lprofProfileDumped()) {
     PROF_NOTE("Profile data not written to file: %s.\n", 
@@ -550,10 +577,18 @@ int __llvm_profile_write_file(void) {
     return -1;
   }
 
+  // Temporarily suspend getting SIGKILL when the parent exits.
+  PDeathSig = lprofSuspendSigKill();
+
   /* Write profile data to the file. */
   rc = writeFile(Filename);
   if (rc)
     PROF_ERR("Failed to write file \"%s\": %s\n", Filename, strerror(errno));
+
+  // Restore SIGKILL.
+  if (PDeathSig == 1)
+    lprofRestoreSigKill();
+
   return rc;
 }
 

runtime/profile-rt/profile-rt-50/InstrProfilingInternal.h

@@ -48,17 +48,21 @@ typedef struct ProfDataIOVec {
   size_t NumElm;
 } ProfDataIOVec;
 
-typedef uint32_t (*WriterCallback)(ProfDataIOVec *, uint32_t NumIOVecs,
+struct ProfDataWriter;
-                                   void **WriterCtx);
+typedef uint32_t (*WriterCallback)(struct ProfDataWriter *This, ProfDataIOVec *,
+                                   uint32_t NumIOVecs);
+
+typedef struct ProfDataWriter {
+  WriterCallback Write;
+  void *WriterCtx;
+} ProfDataWriter;
 
 /*!
  * The data structure for buffered IO of profile data.
  */
 typedef struct ProfBufferIO {
-  /* File handle.  */
+  ProfDataWriter *FileWriter;
-  void *File;
+  uint32_t OwnFileWriter;
-  /* Low level IO callback. */
-  WriterCallback FileWriter;
   /* The start of the buffer. */
   uint8_t *BufferStart;
   /* Total size of the buffer. */
@@ -73,7 +77,7 @@ ProfBufferIO *lprofCreateBufferIOInternal(void *File, uint32_t BufferSz);
 /*!
  * This is the interface to create a handle for buffered IO.
  */
-ProfBufferIO *lprofCreateBufferIO(WriterCallback FileWriter, void *File);
+ProfBufferIO *lprofCreateBufferIO(ProfDataWriter *FileWriter);
 
 /*!
  * The interface to destroy the bufferIO handle and reclaim
@@ -96,8 +100,9 @@ int lprofBufferIOFlush(ProfBufferIO *BufferIO);
 /* The low level interface to write data into a buffer. It is used as the
  * callback by other high level writer methods such as buffered IO writer
  * and profile data writer.  */
-uint32_t lprofBufferWriter(ProfDataIOVec *IOVecs, uint32_t NumIOVecs,
+uint32_t lprofBufferWriter(ProfDataWriter *This, ProfDataIOVec *IOVecs,
-                           void **WriterCtx);
+                           uint32_t NumIOVecs);
+void initBufferWriter(ProfDataWriter *BufferWriter, char *Buffer);
 
 struct ValueProfData;
 struct ValueProfRecord;
@@ -133,15 +138,17 @@ typedef struct VPDataReaderType {
                                         uint32_t N);
 } VPDataReaderType;
 
-int lprofWriteData(WriterCallback Writer, void *WriterCtx,
+/* Write profile data to destinitation. If SkipNameDataWrite is set to 1,
-                   VPDataReaderType *VPDataReader);
+   the name data is already in destintation, we just skip over it. */
-int lprofWriteDataImpl(WriterCallback Writer, void *WriterCtx,
+int lprofWriteData(ProfDataWriter *Writer, VPDataReaderType *VPDataReader,
+                   int SkipNameDataWrite);
+int lprofWriteDataImpl(ProfDataWriter *Writer,
                        const __llvm_profile_data *DataBegin,
                        const __llvm_profile_data *DataEnd,
                        const uint64_t *CountersBegin,
                        const uint64_t *CountersEnd,
                        VPDataReaderType *VPDataReader, const char *NamesBegin,
-                       const char *NamesEnd);
+                       const char *NamesEnd, int SkipNameDataWrite);
 
 /* Merge value profile data pointed to by SrcValueProfData into
  * in-memory profile counters pointed by to DstData.  */

runtime/profile-rt/profile-rt-50/InstrProfilingNameVar.c

@@ -1,11 +1,11 @@
-//===- InstrProfilingNameVar.c - profile name variable setup --------------===//
+/*===- InstrProfilingNameVar.c - profile name variable setup  -------------===*\
-//
+|*
-//                     The LLVM Compiler Infrastructure
+|*                     The LLVM Compiler Infrastructure
-//
+|*
-// This file is distributed under the University of Illinois Open Source
+|* This file is distributed under the University of Illinois Open Source
-// License. See LICENSE.TXT for details.
+|* License. See LICENSE.TXT for details.
-//
+|*
-//===----------------------------------------------------------------------===//
+\*===----------------------------------------------------------------------===*/
 
 #include "InstrProfiling.h"
 

runtime/profile-rt/profile-rt-50/InstrProfilingUtil.c

@@ -29,6 +29,11 @@
 #include <stdlib.h>
 #include <string.h>
 
+#if defined(__linux__)
+#include <signal.h>
+#include <sys/prctl.h>
+#endif
+
 COMPILER_RT_VISIBILITY
 void __llvm_profile_recursive_mkdir(char *path) {
   int i;
@@ -219,3 +224,21 @@ COMPILER_RT_VISIBILITY const char *lprofFindLastDirSeparator(const char *Path) {
 #endif
   return Sep;
 }
+
+COMPILER_RT_VISIBILITY int lprofSuspendSigKill() {
+#if defined(__linux__)
+  int PDeachSig = 0;
+  /* Temporarily suspend getting SIGKILL upon exit of the parent process. */
+  if (prctl(PR_GET_PDEATHSIG, &PDeachSig) == 0 && PDeachSig == SIGKILL)
+    prctl(PR_SET_PDEATHSIG, 0);
+  return (PDeachSig == SIGKILL);
+#else
+  return 0;
+#endif
+}
+
+COMPILER_RT_VISIBILITY void lprofRestoreSigKill() {
+#if defined(__linux__)
+  prctl(PR_SET_PDEATHSIG, SIGKILL);
+#endif
+}

runtime/profile-rt/profile-rt-50/InstrProfilingUtil.h

@@ -51,4 +51,12 @@ int lprofGetHostName(char *Name, int Len);
 unsigned lprofBoolCmpXchg(void **Ptr, void *OldV, void *NewV);
 void *lprofPtrFetchAdd(void **Mem, long ByteIncr);
 
+/* Temporarily suspend SIGKILL. Return value of 1 means a restore is needed.
+ * Other return values mean no restore is needed.
+ */
+int lprofSuspendSigKill();
+
+/* Restore previously suspended SIGKILL. */
+void lprofRestoreSigKill();
+
 #endif /* PROFILE_INSTRPROFILINGUTIL_H */

runtime/profile-rt/profile-rt-50/InstrProfilingValue.c

@@ -219,6 +219,35 @@ __llvm_profile_instrument_target(uint64_t TargetValue, void *Data,
   }
 }
 
+/*
+ * The target values are partitioned into multiple regions/ranges. There is one
+ * contiguous region which is precise -- every value in the range is tracked
+ * individually. A value outside the precise region will be collapsed into one
+ * value depending on the region it falls in.
+ *
+ * There are three regions:
+ * 1. (-inf, PreciseRangeStart) and (PreciseRangeLast, LargeRangeValue) belong
+ * to one region -- all values here should be mapped to one value of
+ * "PreciseRangeLast + 1".
+ * 2. [PreciseRangeStart, PreciseRangeLast]
+ * 3. Large values: [LargeValue, +inf) maps to one value of LargeValue.
+ *
+ * The range for large values is optional. The default value of INT64_MIN
+ * indicates it is not specified.
+ */
+COMPILER_RT_VISIBILITY void __llvm_profile_instrument_range(
+    uint64_t TargetValue, void *Data, uint32_t CounterIndex,
+    int64_t PreciseRangeStart, int64_t PreciseRangeLast, int64_t LargeValue) {
+
+  if (LargeValue != INT64_MIN && (int64_t)TargetValue >= LargeValue)
+    TargetValue = LargeValue;
+  else if ((int64_t)TargetValue < PreciseRangeStart ||
+           (int64_t)TargetValue > PreciseRangeLast)
+    TargetValue = PreciseRangeLast + 1;
+
+  __llvm_profile_instrument_target(TargetValue, Data, CounterIndex);
+}
+
 /*
  * A wrapper struct that represents value profile runtime data.
  * Like InstrProfRecord class which is used by profiling host tools,

runtime/profile-rt/profile-rt-50/InstrProfilingWriter.c

@@ -31,41 +31,44 @@ COMPILER_RT_VISIBILITY uint32_t VPBufferSize = 0;
 /* The buffer writer is reponsponsible in keeping writer state
  * across the call.
  */
-COMPILER_RT_VISIBILITY uint32_t lprofBufferWriter(ProfDataIOVec *IOVecs,
+COMPILER_RT_VISIBILITY uint32_t lprofBufferWriter(ProfDataWriter *This,
-                                                  uint32_t NumIOVecs,
+                                                  ProfDataIOVec *IOVecs,
-                                                  void **WriterCtx) {
+                                                  uint32_t NumIOVecs) {
   uint32_t I;
-  char **Buffer = (char **)WriterCtx;
+  char **Buffer = (char **)&This->WriterCtx;
   for (I = 0; I < NumIOVecs; I++) {
     size_t Length = IOVecs[I].ElmSize * IOVecs[I].NumElm;
+    if (IOVecs[I].Data)
       memcpy(*Buffer, IOVecs[I].Data, Length);
     *Buffer += Length;
   }
   return 0;
 }
 
-static void llvmInitBufferIO(ProfBufferIO *BufferIO, WriterCallback FileWriter,
+static void llvmInitBufferIO(ProfBufferIO *BufferIO, ProfDataWriter *FileWriter,
-                             void *File, uint8_t *Buffer, uint32_t BufferSz) {
+                             uint8_t *Buffer, uint32_t BufferSz) {
-  BufferIO->File = File;
   BufferIO->FileWriter = FileWriter;
+  BufferIO->OwnFileWriter = 0;
   BufferIO->BufferStart = Buffer;
   BufferIO->BufferSz = BufferSz;
   BufferIO->CurOffset = 0;
 }
 
 COMPILER_RT_VISIBILITY ProfBufferIO *
-lprofCreateBufferIO(WriterCallback FileWriter, void *File) {
+lprofCreateBufferIO(ProfDataWriter *FileWriter) {
   uint8_t *Buffer = DynamicBufferIOBuffer;
   uint32_t BufferSize = VPBufferSize;
   if (!Buffer) {
     Buffer = &BufferIOBuffer[0];
     BufferSize = sizeof(BufferIOBuffer);
   }
-  llvmInitBufferIO(&TheBufferIO, FileWriter, File, Buffer, BufferSize);
+  llvmInitBufferIO(&TheBufferIO, FileWriter, Buffer, BufferSize);
   return &TheBufferIO;
 }
 
 COMPILER_RT_VISIBILITY void lprofDeleteBufferIO(ProfBufferIO *BufferIO) {
+  if (BufferIO->OwnFileWriter)
+    FreeHook(BufferIO->FileWriter);
   if (DynamicBufferIOBuffer) {
     FreeHook(DynamicBufferIOBuffer);
     DynamicBufferIOBuffer = 0;
@@ -83,13 +86,16 @@ lprofBufferIOWrite(ProfBufferIO *BufferIO, const uint8_t *Data, uint32_t Size) {
   /* Special case, bypass the buffer completely. */
   ProfDataIOVec IO[] = {{Data, sizeof(uint8_t), Size}};
   if (Size > BufferIO->BufferSz) {
-    if (BufferIO->FileWriter(IO, 1, &BufferIO->File))
+    if (BufferIO->FileWriter->Write(BufferIO->FileWriter, IO, 1))
       return -1;
   } else {
     /* Write the data to buffer */
     uint8_t *Buffer = BufferIO->BufferStart + BufferIO->CurOffset;
-    lprofBufferWriter(IO, 1, (void **)&Buffer);
+    ProfDataWriter BufferWriter;
-    BufferIO->CurOffset = Buffer - BufferIO->BufferStart;
+    initBufferWriter(&BufferWriter, (char *)Buffer);
+    lprofBufferWriter(&BufferWriter, IO, 1);
+    BufferIO->CurOffset =
+        (uint8_t *)BufferWriter.WriterCtx - BufferIO->BufferStart;
   }
   return 0;
 }
@@ -98,7 +104,7 @@ COMPILER_RT_VISIBILITY int lprofBufferIOFlush(ProfBufferIO *BufferIO) {
   if (BufferIO->CurOffset) {
     ProfDataIOVec IO[] = {
         {BufferIO->BufferStart, sizeof(uint8_t), BufferIO->CurOffset}};
-    if (BufferIO->FileWriter(IO, 1, &BufferIO->File))
+    if (BufferIO->FileWriter->Write(BufferIO->FileWriter, IO, 1))
       return -1;
     BufferIO->CurOffset = 0;
   }
@@ -201,7 +207,7 @@ static int writeOneValueProfData(ProfBufferIO *BufferIO,
   return 0;
 }
 
-static int writeValueProfData(WriterCallback Writer, void *WriterCtx,
+static int writeValueProfData(ProfDataWriter *Writer,
                               VPDataReaderType *VPDataReader,
                               const __llvm_profile_data *DataBegin,
                               const __llvm_profile_data *DataEnd) {
@@ -211,7 +217,7 @@ static int writeValueProfData(WriterCallback Writer, void *WriterCtx,
   if (!VPDataReader)
     return 0;
 
-  BufferIO = lprofCreateBufferIO(Writer, WriterCtx);
+  BufferIO = lprofCreateBufferIO(Writer);
 
   for (DI = DataBegin; DI < DataEnd; DI++) {
     if (writeOneValueProfData(BufferIO, VPDataReader, DI))
@@ -225,9 +231,9 @@ static int writeValueProfData(WriterCallback Writer, void *WriterCtx,
   return 0;
 }
 
-COMPILER_RT_VISIBILITY int lprofWriteData(WriterCallback Writer,
+COMPILER_RT_VISIBILITY int lprofWriteData(ProfDataWriter *Writer,
-                                          void *WriterCtx,
+                                          VPDataReaderType *VPDataReader,
-                                          VPDataReaderType *VPDataReader) {
+                                          int SkipNameDataWrite) {
   /* Match logic in __llvm_profile_write_buffer(). */
   const __llvm_profile_data *DataBegin = __llvm_profile_begin_data();
   const __llvm_profile_data *DataEnd = __llvm_profile_end_data();
@@ -235,18 +241,17 @@ COMPILER_RT_VISIBILITY int lprofWriteData(WriterCallback Writer,
   const uint64_t *CountersEnd = __llvm_profile_end_counters();
   const char *NamesBegin = __llvm_profile_begin_names();
   const char *NamesEnd = __llvm_profile_end_names();
-  return lprofWriteDataImpl(Writer, WriterCtx, DataBegin, DataEnd,
+  return lprofWriteDataImpl(Writer, DataBegin, DataEnd, CountersBegin,
-                            CountersBegin, CountersEnd, VPDataReader,
+                            CountersEnd, VPDataReader, NamesBegin, NamesEnd,
-                            NamesBegin, NamesEnd);
+                            SkipNameDataWrite);
 }
 
 COMPILER_RT_VISIBILITY int
-lprofWriteDataImpl(WriterCallback Writer, void *WriterCtx,
+lprofWriteDataImpl(ProfDataWriter *Writer, const __llvm_profile_data *DataBegin,
-                   const __llvm_profile_data *DataBegin,
                    const __llvm_profile_data *DataEnd,
                    const uint64_t *CountersBegin, const uint64_t *CountersEnd,
                    VPDataReaderType *VPDataReader, const char *NamesBegin,
-                   const char *NamesEnd) {
+                   const char *NamesEnd, int SkipNameDataWrite) {
 
   /* Calculate size of sections. */
   const uint64_t DataSize = __llvm_profile_get_data_size(DataBegin, DataEnd);
@@ -268,14 +273,14 @@ lprofWriteDataImpl(WriterCallback Writer, void *WriterCtx,
 #include "InstrProfData.inc"
 
   /* Write the data. */
-  ProfDataIOVec IOVec[] = {{&Header, sizeof(__llvm_profile_header), 1},
+  ProfDataIOVec IOVec[] = {
+      {&Header, sizeof(__llvm_profile_header), 1},
       {DataBegin, sizeof(__llvm_profile_data), DataSize},
       {CountersBegin, sizeof(uint64_t), CountersSize},
-                           {NamesBegin, sizeof(uint8_t), NamesSize},
+      {SkipNameDataWrite ? NULL : NamesBegin, sizeof(uint8_t), NamesSize},
       {Zeroes, sizeof(uint8_t), Padding}};
-  if (Writer(IOVec, sizeof(IOVec) / sizeof(*IOVec), &WriterCtx))
+  if (Writer->Write(Writer, IOVec, sizeof(IOVec) / sizeof(*IOVec)))
     return -1;
 
-  return writeValueProfData(Writer, WriterCtx, VPDataReader, DataBegin,
+  return writeValueProfData(Writer, VPDataReader, DataBegin, DataEnd);
-                            DataEnd);
 }

runtime/profile-rt/profile-rt-60/InstrProfilingFile.c

@@ -519,8 +519,10 @@ void __llvm_profile_initialize_file(void) {
 
   EnvFilenamePat = getFilenamePatFromEnv();
   if (EnvFilenamePat) {
-    SelectedPat = EnvFilenamePat;
+    /* Pass CopyFilenamePat = 1, to ensure that the filename would be valid 
-    PNS = PNS_environment;
+       at the  moment when __llvm_profile_write_file() gets executed. */
+    parseAndSetFilename(EnvFilenamePat, PNS_environment, 1);
+    return;
   } else if (hasCommandLineOverrider) {
     SelectedPat = INSTR_PROF_PROFILE_NAME_VAR;
     PNS = PNS_command_line;

runtime/profile-rt/profile-rt-60/InstrProfilingNameVar.c

@@ -1,11 +1,11 @@
-//===- InstrProfilingNameVar.c - profile name variable setup --------------===//
+/*===- InstrProfilingNameVar.c - profile name variable setup  -------------===*\
-//
+|*
-//                     The LLVM Compiler Infrastructure
+|*                     The LLVM Compiler Infrastructure
-//
+|*
-// This file is distributed under the University of Illinois Open Source
+|* This file is distributed under the University of Illinois Open Source
-// License. See LICENSE.TXT for details.
+|* License. See LICENSE.TXT for details.
-//
+|*
-//===----------------------------------------------------------------------===//
+\*===----------------------------------------------------------------------===*/
 
 #include "InstrProfiling.h"
 

tests/codegen/dcompute_cl_addrspaces.d

@@ -1,6 +1,6 @@
 // REQUIRES: target_SPIRV
-// RUN: %ldc -c -mdcompute-targets=ocl-220 -m64 -output-ll -output-o %s && FileCheck %s --check-prefix=LL < kernels_ocl220_64.ll \
+// RUN: %ldc -c -mdcompute-targets=ocl-220 -m64 -mdcompute-file-prefix=addrspace -output-ll -output-o %s && FileCheck %s --check-prefix=LL < addrspace_ocl220_64.ll \
-// RUN: && %llvm-spirv -to-text kernels_ocl220_64.spv && FileCheck %s --check-prefix=SPT < kernels_ocl220_64.spt
+// RUN: && %llvm-spirv -to-text addrspace_ocl220_64.spv && FileCheck %s --check-prefix=SPT < addrspace_ocl220_64.spt
 @compute(CompileFor.deviceOnly) module dcompute_cl_addrspaces;
 import ldc.dcompute;
 

tests/codegen/dcompute_cu_addrspaces.d

@@ -1,6 +1,6 @@
 // REQUIRES: atleast_llvm309
 // REQUIRES: target_NVPTX
-// RUN: %ldc -c -mdcompute-targets=cuda-350 -m64 -output-ll -output-o %s && FileCheck %s --check-prefix=LL < kernels_cuda350_64.ll && FileCheck %s --check-prefix=PTX < kernels_cuda350_64.ptx
+// RUN: %ldc -c -mdcompute-targets=cuda-350 -m64 -mdcompute-file-prefix=addrspace -output-ll -output-o %s && FileCheck %s --check-prefix=LL < addrspace_cuda350_64.ll && FileCheck %s --check-prefix=PTX < addrspace_cuda350_64.ptx
 @compute(CompileFor.deviceOnly) module dcompute_cu_addrspaces;
 import ldc.dcompute;
 

tests/codegen/dcompute_host_and_device.d

@@ -0,0 +1,16 @@
+// Check that we can generate code for both the host and device in one compiler invocation
+// REQUIRES: atleast_llvm309
+// REQUIRES: target_NVPTX
+// RUN: %ldc -mdcompute-targets=cuda-350 -mdcompute-file-prefix=host_and_device -Iinputs %s %S/inputs/kernel.d
+
+import inputs.kernel : foo;
+
+int tlGlobal;
+__gshared int gGlobal;
+
+void main(string[] args)
+{
+    tlGlobal = 0;
+    gGlobal  = 0;
+    string s = foo.mangleof;
+}

tests/codegen/inputs/kernel.d

@@ -0,0 +1,5 @@
+@compute(CompileFor.deviceOnly)
+module inputs.kernel;
+
+import ldc.dcompute;
+@kernel void foo() {}

tests/debuginfo/args_cdb.d

@@ -13,17 +13,21 @@ import core.simd;
 // enable case sensitive symbol lookup
 // CDB: .symopt-1
 
+struct Small { size_t val; }
+struct Large { size_t a, b, c, d, e, f, g, h; }
+
 int byValue(ubyte ub, ushort us, uint ui, ulong ul,
             float f, double d, real r,
             cdouble c, int delegate() dg, int function() fun,
             int[] slice, float[int] aa, ubyte[16] fa,
             float4 f4, double4 d4,
-            Interface ifc, TypeInfo_Class ti, typeof(null) np)
+            Small small, Large large,
+            TypeInfo_Class ti, typeof(null) np)
 {
-// CDB: bp `args_cdb.d:25`
+// CDB: bp `args_cdb.d:27`
 // CDB: g
-    // arguments implicitly passed by reference on x64 and not shown if unused
+    // arguments implicitly passed by reference aren't shown if unused
-    float cim = c.im + fa[7] + dg() + ifc.offset;
+    float cim = c.im + fa[7] + dg() + small.val + large.a;
     return 1;
 // CHECK: !args_cdb.byValue
 // CDB: dv /t
@@ -42,12 +46,13 @@ int byValue(ubyte ub, ushort us, uint ui, ulong ul,
 // CHECK: <function> * fun = {{0x[0-9a-f`]*}}
 // x86: struct int[] slice =
 // CHECK: unsigned char * aa = {{0x[0-9a-f`]*}}
-// "Internal implementation error for fa" with cdb on x64, ok in VS
+// x64: unsigned char (*)[16] fa
 // x86: unsigned char [16] fa
 // x86: float [4] f4 = float [4]
 // x86: double [4] d4 = double [4]
-// x64: Interface * ifc
+// CHECK: Small small
-// x86: Interface ifc
+// x64: Large * large
+// x86: Large large
 // CHECK: struct TypeInfo_Class * ti = {{0x[0-9a-f`]*}}
 // CHECK: void * np = {{0x[0`]*}}
 
@@ -64,31 +69,35 @@ int byValue(ubyte ub, ushort us, uint ui, ulong ul,
 
 // CDB: ?? dg
 // CHECK: int delegate()
-// CHECK-NEXT: context
+// CHECK-NEXT: context :
-// CHECK-NEXT: funcptr
+// CHECK-NEXT: funcptr :
 // CHECK-SAME: args_cdb.main.__lambda
 
 // CDB: ?? slice
 // CHECK: struct int[]
 // CHECK-NEXT: length : 2
-// CHECK-NEXT: ptr
+// CHECK-NEXT: ptr :
+// CHECK-SAME: 0n10
 
-// CDB: ?? fa[1]
+// CDB: ?? (*fa)[1]
-// "Internal implementation error for fa" with cdb on x64, ok in VS
+// x64: unsigned char 0x0e
-// no-x86: unsigned char 0x0e (displays 0xf6)
+// no-x86: would be fa[1], but displays garbage anyway
 
 // CDB: ?? f4[1]
 // CHECK: float 16
 
-// CDB: ?? d4[2]
+// CDB: ?? d4[1]
 // CHECK: double 17
 
-// CDB: ?? ifc
+// CDB: ?? small
-// CHECK: Interface
+// CHECK: Small
-// CHECK-NEXT: classinfo
+// x64-NEXT: val : 0x12
-// CHECK-NEXT: vtbl
+// no-x86-NEXT: val : 0x12 (displays garbage)
-// x64-NEXT: offset : 0x12
+
-// no-x86-NEXT: offset : 0x12 (displays 0)
+// CDB: ?? large
+// CHECK: Large
+// x64-NEXT: a : 0x13
+// no-x86-NEXT: a : 0x13 (displays garbage)
 
 // CDB: ?? ti
 // CHECK: TypeInfo_Class
@@ -100,9 +109,10 @@ int byPtr(ubyte* ub, ushort* us, uint* ui, ulong* ul,
           cdouble* c, int delegate()* dg, int function()* fun,
           int[]* slice, float[int]* aa, ubyte[16]* fa,
           float4* f4, double4* d4,
-          Interface* ifc, TypeInfo_Class* ti, typeof(null)* np)
+          Small* small, Large* large,
+          TypeInfo_Class* ti, typeof(null)* np)
 {
-// CDB: bp `args_cdb.d:106`
+// CDB: bp `args_cdb.d:115`
 // CDB: g
     return 3;
 // CHECK: !args_cdb.byPtr
@@ -127,8 +137,8 @@ int byPtr(ubyte* ub, ushort* us, uint* ui, ulong* ul,
 // CHECK-NEXT: im : 9
 // CDB: ?? *dg
 // CHECK: int delegate()
-// CHECK-NEXT: context
+// CHECK-NEXT: context :
-// CHECK-NEXT: funcptr
+// CHECK-NEXT: funcptr :
 // CHECK-SAME: args_cdb.main.__lambda
 // CDB: ?? *fun
 // CHECK: <function> *
@@ -145,13 +155,16 @@ int byPtr(ubyte* ub, ushort* us, uint* ui, ulong* ul,
 // CHECK: float 16
 // CDB: ?? (*d4)[2]
 // CHECK: double 17
-// CDB: ?? *ifc
+// CDB: ?? *small
-// CHECK: struct Interface
+// CHECK: struct Small
-// CHECK: offset : 0x12
+// CHECK-NEXT: val : 0x12
+// CDB: ?? *large
+// CHECK: struct Large
+// CHECK-NEXT: a : 0x13
+// CHECK-NEXT: b :
 // CDB: ?? *ti
 // CHECK: struct TypeInfo_Class
 // CHECK-NEXT: m_init : byte[]
-// shows bad member values
 // CDB: ?? *np
 // CHECK: void * {{0x[0`]*}}
 }
@@ -161,9 +174,10 @@ int byRef(ref ubyte ub, ref ushort us, ref uint ui, ref ulong ul,
           ref cdouble c, ref int delegate() dg, ref int function() fun,
           ref int[] slice, ref float[int] aa, ref ubyte[16] fa,
           ref float4 f4, ref double4 d4,
-          ref Interface ifc, ref TypeInfo_Class ti, ref typeof(null) np)
+          ref Small small, ref Large large,
+          ref TypeInfo_Class ti, ref typeof(null) np)
 {
-// CDB: bp `args_cdb.d:218`
+// CDB: bp `args_cdb.d:180`
 // CDB: g
 // CHECK: !args_cdb.byRef
 
@@ -189,7 +203,7 @@ int byRef(ref ubyte ub, ref ushort us, ref uint ui, ref ulong ul,
 // CHECK-NEXT: im : 9
 // CDB: ?? *dg
 // CHECK: int delegate()
-// CHECK-NEXT: context
+// CHECK-NEXT: context :
 // CHECK-NEXT: funcptr : {{0x[0-9a-f`]*}}
 // CHECK-SAME: args_cdb.main.__lambda
 // CDB: ?? *fun
@@ -206,16 +220,23 @@ int byRef(ref ubyte ub, ref ushort us, ref uint ui, ref ulong ul,
 // CHECK: float 16
 // CDB: ?? (*d4)[2]
 // CHECK: double 17
-// CDB: ?? *ifc
+// CDB: ?? *small
-// CHECK: struct Interface
+// CHECK: struct Small
-// CHECK: offset : 0x12
+// CHECK-NEXT: val : 0x12
+// CDB: ?? *large
+// CHECK: struct Large
+// CHECK-NEXT: a : 0x13
+// CHECK-NEXT: b :
 // CDB: ?? *ti
 // CHECK: struct TypeInfo_Class * {{0x[0-9a-f`]*}}
 // CHECK-NEXT: m_init : byte[]
 // CDB: ?? *np
-// CHECK: void * {{0x[0`]*}}
+// no-CHECK: void * {{0x[0`]*}} (not available)
+
+    // needs access to references to actually generate debug info
+    float cim = c.im + fa[7] + dg() + fun() + slice.length + aa.length + f4[0] + d4[1] +
+                small.val + large.a + ti.initializer.length;
 
-// needs access to references to actually generate debug info
     ub++;
     us++;
     ui++;
@@ -231,7 +252,8 @@ int byRef(ref ubyte ub, ref ushort us, ref uint ui, ref ulong ul,
     fa[0]++;
     f4.array[0] = f4.array[0] + 1;
     d4.array[0] = d4.array[0] + 1;
-    ifc = Interface(typeid(Object), null, 18);
+    small.val++;
+    large.a++;
     ti = typeid(TypeInfo);
     np = null;
     return 2;
@@ -254,13 +276,14 @@ int main()
     ubyte[16] fa; fa[] = 14;
     float4 f4 = 16;
     double4 d4 = 17;
-    Interface ifc = Interface(typeid(Object), null, 18);
+    Small small = Small(18);
+    Large large = Large(19);
     TypeInfo_Class ti = typeid(TypeInfo);
     typeof(null) np = null;
     
-    byValue(ub, us, ui, ul, f, d, r, c, dg, fun, slice, aa, fa, f4, d4, ifc, ti, np);
+    byValue(ub, us, ui, ul, f, d, r, c, dg, fun, slice, aa, fa, f4, d4, small, large, ti, np);
-    byPtr(&ub, &us, &ui, &ul, &f, &d, &r, &c, &dg, &fun, &slice, &aa, &fa, &f4, &d4, &ifc, &ti, &np);
+    byPtr(&ub, &us, &ui, &ul, &f, &d, &r, &c, &dg, &fun, &slice, &aa, &fa, &f4, &d4, &small, &large, &ti, &np);
-    byRef(ub, us, ui, ul, f, d, r, c, dg, fun, slice, aa, fa, f4, d4, ifc, ti, np);
+    byRef(ub, us, ui, ul, f, d, r, c, dg, fun, slice, aa, fa, f4, d4, small, large, ti, np);
 
     return 0;
 }

tests/debuginfo/nested_cdb.d

@@ -0,0 +1,67 @@
+// REQUIRES: atleast_llvm500
+// REQUIRES: Windows
+// REQUIRES: cdb
+// RUN: %ldc -g -of=%t.exe %s
+// RUN: sed -e "/^\\/\\/ CDB:/!d" -e "s,// CDB:,," %s \
+// RUN:    | %cdb -snul -lines -y . %t.exe >%t.out
+// RUN: FileCheck %s -check-prefix=CHECK -check-prefix=%arch < %t.out
+
+// CDB: ld /f nested_cdb*
+// enable case sensitive symbol lookup
+// CDB: .symopt-1
+
+void encloser(int arg0, ref int arg1)
+{
+    int enc_n = 123;
+// CDB: bp `nested_cdb.d:16`
+// CDB: g
+// CDB: dv /t
+// CHECK: int arg0 = 0n1
+// (cdb displays references as pointers)
+// CHECK-NEXT: int * arg1 = {{0x[0-9a-f`]*}}
+// CHECK-NEXT: int enc_n = 0n123
+// CDB: ?? *arg1
+// CHECK: int 0n2
+    enc_n += arg1;
+
+    void nested(int nes_i)
+    {
+        int blub = arg0 + arg1 + enc_n;
+// CDB: bp `nested_cdb.d:30`
+// CDB: g
+// CDB: dv /t
+// CHECK: int arg0 = 0n1
+// CHECK-NEXT: int * arg1 = {{0x[0-9a-f`]*}}
+// CHECK-NEXT: int enc_n = 0n125
+// CDB: ?? *arg1
+// CHECK: int 0n2
+        arg0 = arg1 = enc_n = nes_i;
+// CDB: bp `nested_cdb.d:39`
+// CDB: g
+// CDB: dv /t
+// CHECK: int arg0 = 0n456
+// CHECK-NEXT: int * arg1 = {{0x[0-9a-f`]*}}
+// CHECK-NEXT: int enc_n = 0n456
+// CDB: ?? *arg1
+// CHECK: int 0n456
+    }
+
+    nested(456);
+// CDB: bp `nested_cdb.d:50`
+// CDB: g
+// CDB: dv /t
+// CHECK: int arg0 = 0n456
+// CHECK-NEXT: int * arg1 = {{0x[0-9a-f`]*}}
+// CHECK-NEXT: int enc_n = 0n456
+// CDB: ?? *arg1
+// CHECK: int 0n456
+}
+
+void main()
+{
+    int arg1 = 2;
+    encloser(1, arg1);
+}
+
+// CDB: q
+// CHECK: quit

tests/debuginfo/nested_gdb.d

@@ -0,0 +1,61 @@
+// REQUIRES: gdb
+// RUN: %ldc %_gdb_dflags -g -of=%t %s
+// RUN: sed -e "/^\\/\\/ GDB:/!d" -e "s,// GDB:,," %s >%t.gdb
+// RUN: gdb %t --batch -x %t.gdb >%t.out 2>&1
+// RUN: FileCheck %s -check-prefix=CHECK < %t.out
+
+void encloser(int arg0, ref int arg1)
+{
+    int enc_n = 123;
+// GDB: b 10
+// GDB: r
+// GDB: p arg0
+// CHECK: $1 = 1
+// GDB: p arg1
+// CHECK: $2 = (int &) @{{0x[0-9a-f]*}}: 2
+// GDB: p enc_n
+// CHECK: $3 = 123
+    enc_n += arg1;
+
+    void nested(int nes_i)
+    {
+        int blub = arg0 + arg1 + enc_n;
+// GDB: b 23
+// GDB: c
+// GDB: p arg0
+// CHECK: $4 = 1
+// GDB: p arg1
+// CHECK: $5 = (int &) @{{0x[0-9a-f]*}}: 2
+// GDB: p enc_n
+// CHECK: $6 = 125
+        arg0 = arg1 = enc_n = nes_i;
+// GDB: b 32
+// GDB: c
+// GDB: p arg0
+// CHECK: $7 = 456
+// GDB: p arg1
+// CHECK: $8 = (int &) @{{0x[0-9a-f]*}}: 456
+// GDB: p enc_n
+// CHECK: $9 = 456
+    }
+
+    nested(456);
+// GDB: b 43
+// GDB: c
+// GDB: p arg0
+// no-CHECK: $10 = 456 (`<optimized out>` for LLVM < 5.0)
+// GDB: p arg1
+// CHECK: $11 = (int &) @{{0x[0-9a-f]*}}: 456
+// GDB: p enc_n
+// CHECK: $12 = 456
+}
+
+void main()
+{
+    int arg1 = 2;
+    encloser(1, arg1);
+}
+
+// GDB: c
+// GDB: q
+// CHECK: exited normally

tests/driver/inputs/post_switches.conf

@@ -0,0 +1,11 @@
+default:
+{
+    switches = [
+        "-L-normal-switch",
+        "-L-normal-two-switch"
+    ];
+    post-switches = [
+        "-L-post-switch",
+        "-L-post-two-switch"
+    ];
+};

tests/driver/post_switches.d

@@ -0,0 +1,21 @@
+// RUN: not %ldc -I=%runtimedir/src -conf=%S/inputs/post_switches.conf %s -v -L-user-passed-switch | FileCheck %s
+
+// CHECK: -normal-switch
+// CHECK-SAME: -normal-two-switch
+// CHECK-SAME: -user-passed-switch
+// CHECK-SAME: -post-switch
+// CHECK-SAME: -post-two-switch
+
+
+// RUN: not %ldc -I=%runtimedir/src -conf=%S/inputs/post_switches.conf -v -L-user-passed-switch -run %s -L-after-run | FileCheck %s --check-prefix=WITHrUN
+
+// WITHrUN: -normal-switch
+// WITHrUN-SAME: -normal-two-switch
+// WITHrUN-SAME: -user-passed-switch
+// WITHrUN-SAME: -post-switch
+// WITHrUN-SAME: -post-two-switch
+// WITHrUN-NOT: -after-run
+
+void main()
+{
+}

tests/lit.site.cfg.in

@@ -14,6 +14,7 @@ config.ldcprofdata_bin     = "@LDCPROFDATA_BIN@"
 config.ldcprunecache_bin   = "@LDCPRUNECACHE_BIN@"
 config.ldc2_bin_dir        = "@LDC2_BIN_DIR@"
 config.ldc2_lib_dir        = "@LDC2_LIB_DIR@"
+config.ldc2_runtime_dir    = "@RUNTIME_DIR@"
 config.test_source_root    = "@TESTS_IR_DIR@"
 config.llvm_tools_dir      = "@LLVM_TOOLS_DIR@"
 config.llvm_version        = @LDC_LLVM_VER@
@@ -45,6 +46,10 @@ config.excludes = [
 if not config.with_PGO:
     config.excludes.append('PGO')
 
+# Explicit forwarding of environment variables
+env_cc = os.environ.get('CC', '')
+if env_cc:
+    config.environment['CC'] = env_cc
 
 # Define available features so that we can disable tests depending on LLVM version
 config.available_features.add("llvm%d" % config.llvm_version)
@@ -71,7 +76,8 @@ if (platform.system() == 'Windows') and (config.default_target_bits == 32):
 if (platform.system() == 'Windows') and (config.default_target_bits == 64):
     config.available_features.add('Windows_x64')
 
-# Add "LTO" feature if linker support is available (LTO is supported from LLVM 3.9)
+# Add "LTO" feature if linker support and LTO plugin are available
+# (LTO is supported from LLVM 3.9)
 canDoLTO = False
 if (config.llvm_version >= 309):
     if (platform.system() == 'Darwin'):
@@ -79,14 +85,14 @@ if (config.llvm_version >= 309):
         p = subprocess.Popen(command, stdout=subprocess.PIPE,
                 stderr=subprocess.PIPE, universal_newlines=True)
         text = p.stderr.read()
-        if "LTO support" in text:
+        if ("LTO support" in text and os.path.isfile(config.ldc2_lib_dir + '/libLTO-ldc.dylib')):
             canDoLTO = True
     elif (platform.system() == 'Linux'):
         command = ['ld', '-plugin']
         p = subprocess.Popen(command, stdout=subprocess.PIPE,
                 stderr=subprocess.PIPE, universal_newlines=True)
         text = p.stderr.read()
-        if "plugin: missing argument" in text:
+        if ("plugin: missing argument" in text and os.path.isfile(config.ldc2_lib_dir + '/LLVMgold-ldc.so')):
             canDoLTO = True
 if canDoLTO:
     config.available_features.add('LTO')
@@ -110,6 +116,7 @@ config.substitutions.append( ('%ldc', config.ldc2_bin) )
 config.substitutions.append( ('%profdata', config.ldcprofdata_bin) )
 config.substitutions.append( ('%prunecache', config.ldcprunecache_bin) )
 config.substitutions.append( ('%llvm-spirv', os.path.join(config.llvm_tools_dir, 'llvm-spirv')) )
+config.substitutions.append( ('%runtimedir', config.ldc2_runtime_dir ) )
 
 # Add platform-dependent file extension substitutions
 if (platform.system() == 'Windows'):

tests/sanitizers/fsanitize_fuzzer.d

@@ -10,7 +10,7 @@
 bool FuzzMe(const ubyte* data, size_t dataSize)
 {
     // CHECK: call {{.*}}_sanitizer_cov_trace_pc_guard
-    // CHECK: call {{.*}}_sanitizer_cov_trace_cmp
+    // CHECK: call {{.*}}_sanitizer_cov_trace_{{(const_)?}}cmp
 
     return dataSize >= 3 &&
            data[0] == 'F' &&

tests/sanitizers/fuzz_asan.d

@@ -1,5 +1,6 @@
 // Test Fuzz+ASan functionality
 
+// REQUIRES: atleast_llvm500
 // REQUIRES: Fuzzer, ASan
 
 // See https://github.com/ldc-developers/ldc/issues/2222 for -disable-fp-elim
@@ -8,18 +9,15 @@
 
 bool FuzzMe(ubyte* data, size_t dataSize)
 {
-    return dataSize >= 6 &&
+    return dataSize >= 3 &&
            data[0] == 'F' &&
            data[1] == 'U' &&
            data[2] == 'Z' &&
-           data[3] == 'F' &&
-           data[4] == 'U' &&
-           data[5] == 'Z' &&
     // CHECK: stack-buffer-overflow
     // CHECK-NEXT: READ of size 1
     // CHECK-NEXT: #0 {{.*}} in {{.*fuzz_asan6FuzzMe.*}} {{.*}}fuzz_asan.d:
     // FIXME, debug line info is wrong (Github issue #2090). Once fixed, add [[@LINE+1]]
-           data[6] == 'Z'; // :‑<
+           data[3] == 'Z'; // :‑<
 }
 
 extern (C) int LLVMFuzzerTestOneInput(const(ubyte*) data, size_t size)
@@ -33,7 +31,7 @@ extern (C) int LLVMFuzzerTestOneInput(const(ubyte*) data, size_t size)
         init = true;
     }
 
-    ubyte[6] stackdata;
+    ubyte[3] stackdata;
     if (data)
     {
         for (auto i = 0; (i < size) && (i < stackdata.length); ++i)
@@ -45,5 +43,5 @@ extern (C) int LLVMFuzzerTestOneInput(const(ubyte*) data, size_t size)
     return 0;
 }
 
-// The test unit should start with "FUZFUZ"
+// The test unit should start with "FUZ"
-// CHECK: FUZFUZ
+// CHECK: FUZ

tests/sanitizers/fuzz_basic.d

@@ -1,5 +1,6 @@
 // Test basic fuzz test crash
 
+// REQUIRES: atleast_llvm500
 // REQUIRES: Fuzzer
 
 // RUN: %ldc -g -fsanitize=fuzzer %s -of=%t%exe
@@ -9,8 +10,7 @@
 
 void FuzzMe(const(ubyte*) data, size_t size)
 {
-    if ((size >= 6) && data[0] == 'F' && data[1] == 'U' && data[2] == 'Z'
+    if ((size >= 3) && data[0] == 'F' && data[1] == 'U' && data[2] == 'Z')
-            && data[3] == 'L' && data[4] == 'D' && data[5] == 'C')
     {
         assert(false);
     }
@@ -34,5 +34,5 @@ extern (C) int LLVMFuzzerTestOneInput(const(ubyte*) data, size_t size)
     return 0;
 }
 
-// The test unit should start with "FUZLDC"
+// The test unit should start with "FUZ"
-// CHECK: FUZLDC
+// CHECK: FUZ

tests/sanitizers/link_fuzzer.d

@@ -1,11 +1,13 @@
 // Test linking C++ stdlib (or not) with -fsanitize=fuzzer
 
+// REQUIRES: atleast_llvm500
 // REQUIRES: Fuzzer
 
 // RUN: %ldc -v -fsanitize=fuzzer %s | FileCheck %s
 // RUN: not %ldc -v -fsanitize=fuzzer -link-no-cpp %s | FileCheck %s --check-prefix=NOCPP
 
-// CHECK: libFuzzer.a
+// "libFuzzer.a" before LLVM 6.0, "lib(ldc|clang)_rt.fuzzer.*.a" since LLVM 6.0
+// CHECK: {{(libFuzzer\.a|_rt\.fuzzer.*\.a)}}
 // CHECK-SAME: -l{{(c|stdc)}}++
 
 // NOCPP-NOT: -l{{(c|stdc)}}++

tests/sanitizers/lit.local.cfg

@@ -9,7 +9,7 @@ if (platform.system() == 'Darwin') or (platform.system() == 'Linux'):
         if m is not None:
             config.available_features.add('ASan')
             continue
-        m = re.match('.*Fuzzer.*', file)
+        m = re.match('.*(F|f)uzzer.*', file)
         if m is not None:
             config.available_features.add('Fuzzer')
             continue