re PR driver/68463 (Offloading fails when some objects are compiled with LTO and some without)

gcc/
	PR driver/68463
	* config/gnu-user.h (CRTOFFLOADBEGIN): Define.  Add crtoffloadbegin.o if
	offloading is enabled and -fopenacc or -fopenmp is specified.
	(CRTOFFLOADEND): Likewise.
	(GNU_USER_TARGET_STARTFILE_SPEC): Add CRTOFFLOADBEGIN.
	(GNU_USER_TARGET_ENDFILE_SPEC): Add CRTOFFLOADEND.
	* lto-wrapper.c (offloadbegin, offloadend): Remove static vars.
	(offload_objects_file_name): New static var.
	(tool_cleanup): Remove offload_objects_file_name file.
	(find_offloadbeginend): Replace with ...
	(find_crtoffloadtable): ... this.
	(run_gcc): Remove offload_argc and offload_argv.
	Get offload_objects_file_name from -foffload-objects=... option.
	Read names of object files with offload from this file, pass them to
	compile_images_for_offload_targets.  Don't call find_offloadbeginend and
	don't pass offloadbegin and offloadend to the linker.  Don't pass
	offload non-LTO files to the linker, because now they're not claimed.
libgcc/
	PR driver/68463
	* Makefile.in (crtoffloadtable$(objext)): New rule.
	* configure.ac (extra_parts): Add crtoffloadtable$(objext) if
	enable_offload_targets is not empty.
	* configure: Regenerate.
	* offloadstuff.c: Move __OFFLOAD_TABLE__ from crtoffloadend to
	crtoffloadtable.
libgomp/
	PR driver/68463
	* testsuite/libgomp.oacc-c-c++-common/parallel-dims-2.c: Remove.
lto-plugin/
	PR driver/68463
	* lto-plugin.c (struct plugin_offload_file): New.
	(offload_files): Change type.
	(offload_files_last, offload_files_last_obj): New.
	(offload_files_last_lto): New.
	(free_2): Adjust accordingly.
	(all_symbols_read_handler): Don't add offload files to lto_arg_ptr.
	Don't call free_1 for offload_files.  Write names of object files with
	offloading to the temporary file.  Add new option to lto_arg_ptr.
	(claim_file_handler): Don't claim file if it contains offload sections
	without LTO sections.  If it contains offload sections, add to the list.

From-SVN: r233712

12 files changed, 299 insertions, 119 deletions

diff --git a/gcc/ChangeLog b/gcc/ChangeLog
index 9daf0c01ce7..21d304d1cc0 100644
--- a/gcc/ChangeLog
+++ b/gcc/ChangeLog
@@ -1,3 +1,23 @@
+2016-02-25  Ilya Verbin  <ilya.verbin@intel.com>
+
+	PR driver/68463
+	* config/gnu-user.h (CRTOFFLOADBEGIN): Define.  Add crtoffloadbegin.o if
+	offloading is enabled and -fopenacc or -fopenmp is specified.
+	(CRTOFFLOADEND): Likewise.
+	(GNU_USER_TARGET_STARTFILE_SPEC): Add CRTOFFLOADBEGIN.
+	(GNU_USER_TARGET_ENDFILE_SPEC): Add CRTOFFLOADEND.
+	* lto-wrapper.c (offloadbegin, offloadend): Remove static vars.
+	(offload_objects_file_name): New static var.
+	(tool_cleanup): Remove offload_objects_file_name file.
+	(find_offloadbeginend): Replace with ...
+	(find_crtoffloadtable): ... this.
+	(run_gcc): Remove offload_argc and offload_argv.
+	Get offload_objects_file_name from -foffload-objects=... option.
+	Read names of object files with offload from this file, pass them to
+	compile_images_for_offload_targets.  Don't call find_offloadbeginend and
+	don't pass offloadbegin and offloadend to the linker.  Don't pass
+	offload non-LTO files to the linker, because now they're not claimed.
+
 2016-02-25  Jan Hubicka  <hubicka@ucw.cz>
 
 	PR ipa/69630
diff --git a/gcc/config/gnu-user.h b/gcc/config/gnu-user.h
index 2f1bbccb055..b0bf40a954f 100644
--- a/gcc/config/gnu-user.h
+++ b/gcc/config/gnu-user.h
@@ -35,6 +35,14 @@ see the files COPYING3 and COPYING.RUNTIME respectively.  If not, see
 #undef ASM_APP_OFF
 #define ASM_APP_OFF "#NO_APP\n"
 
+#if ENABLE_OFFLOADING == 1
+#define CRTOFFLOADBEGIN "%{fopenacc|fopenmp:crtoffloadbegin%O%s}"
+#define CRTOFFLOADEND "%{fopenacc|fopenmp:crtoffloadend%O%s}"
+#else
+#define CRTOFFLOADBEGIN ""
+#define CRTOFFLOADEND ""
+#endif
+
 /* Provide a STARTFILE_SPEC appropriate for GNU userspace.  Here we add
    the GNU userspace magical crtbegin.o file (see crtstuff.c) which
    provides part of the support for getting C++ file-scope static
@@ -49,14 +57,16 @@ see the files COPYING3 and COPYING.RUNTIME respectively.  If not, see
 	      %{" NO_PIE_SPEC ":crtbegin.o%s}} \
    %{fvtable-verify=none:%s; \
      fvtable-verify=preinit:vtv_start_preinit.o%s; \
-     fvtable-verify=std:vtv_start.o%s}"
+     fvtable-verify=std:vtv_start.o%s} \
+   " CRTOFFLOADBEGIN
 #else
 #define GNU_USER_TARGET_STARTFILE_SPEC \
   "%{!shared: %{pg|p|profile:gcrt1.o%s;:crt1.o%s}} \
    crti.o%s %{static:crtbeginT.o%s;shared|pie:crtbeginS.o%s;:crtbegin.o%s} \
    %{fvtable-verify=none:%s; \
      fvtable-verify=preinit:vtv_start_preinit.o%s; \
-     fvtable-verify=std:vtv_start.o%s}"
+     fvtable-verify=std:vtv_start.o%s} \
+   " CRTOFFLOADBEGIN
 #endif
 #undef  STARTFILE_SPEC
 #define STARTFILE_SPEC GNU_USER_TARGET_STARTFILE_SPEC
@@ -73,13 +83,15 @@ see the files COPYING3 and COPYING.RUNTIME respectively.  If not, see
      fvtable-verify=preinit:vtv_end_preinit.o%s; \
      fvtable-verify=std:vtv_end.o%s} \
    %{shared:crtendS.o%s;: %{" PIE_SPEC ":crtendS.o%s} \
-   %{" NO_PIE_SPEC ":crtend.o%s}} crtn.o%s"
+   %{" NO_PIE_SPEC ":crtend.o%s}} crtn.o%s \
+   " CRTOFFLOADEND
 #else
 #define GNU_USER_TARGET_ENDFILE_SPEC \
   "%{fvtable-verify=none:%s; \
      fvtable-verify=preinit:vtv_end_preinit.o%s; \
      fvtable-verify=std:vtv_end.o%s} \
-   %{shared|pie:crtendS.o%s;:crtend.o%s} crtn.o%s"
+   %{shared|pie:crtendS.o%s;:crtend.o%s} crtn.o%s \
+   " CRTOFFLOADEND
 #endif
 #undef  ENDFILE_SPEC
 #define ENDFILE_SPEC GNU_USER_TARGET_ENDFILE_SPEC
diff --git a/gcc/lto-wrapper.c b/gcc/lto-wrapper.c
index ed20b4ee95e..f240812f952 100644
--- a/gcc/lto-wrapper.c
+++ b/gcc/lto-wrapper.c
@@ -68,7 +68,7 @@ static unsigned int nr;
 static char **input_names;
 static char **output_names;
 static char **offload_names;
-static const char *offloadbegin, *offloadend;
+static char *offload_objects_file_name;
 static char *makefile;
 
 const char tool_name[] = "lto-wrapper";
@@ -84,6 +84,8 @@ tool_cleanup (bool)
     maybe_unlink (ltrans_output_file);
   if (flto_out)
     maybe_unlink (flto_out);
+  if (offload_objects_file_name)
+    maybe_unlink (offload_objects_file_name);
   if (makefile)
     maybe_unlink (makefile);
   for (i = 0; i < nr; ++i)
@@ -840,40 +842,32 @@ copy_file (const char *dest, const char *src)
     }
 }
 
-/* Find the crtoffloadbegin.o and crtoffloadend.o files in LIBRARY_PATH, make
-   copies and store the names of the copies in offloadbegin and offloadend.  */
+/* Find the crtoffloadtable.o file in LIBRARY_PATH, make copy and pass name of
+   the copy to the linker.  */
 
 static void
-find_offloadbeginend (void)
+find_crtoffloadtable (void)
 {
   char **paths = NULL;
   const char *library_path = getenv ("LIBRARY_PATH");
   if (!library_path)
     return;
-  unsigned n_paths = parse_env_var (library_path, &paths, "/crtoffloadbegin.o");
+  unsigned n_paths = parse_env_var (library_path, &paths, "/crtoffloadtable.o");
 
   unsigned i;
   for (i = 0; i < n_paths; i++)
     if (access_check (paths[i], R_OK) == 0)
       {
-	size_t len = strlen (paths[i]);
-	char *tmp = xstrdup (paths[i]);
-	strcpy (paths[i] + len - strlen ("begin.o"), "end.o");
-	if (access_check (paths[i], R_OK) != 0)
-	  fatal_error (input_location,
-		       "installation error, can't find crtoffloadend.o");
-	/* The linker will delete the filenames we give it, so make
-	   copies.  */
-	offloadbegin = make_temp_file (".o");
-	offloadend = make_temp_file (".o");
-	copy_file (offloadbegin, tmp);
-	copy_file (offloadend, paths[i]);
-	free (tmp);
+	/* The linker will delete the filename we give it, so make a copy.  */
+	char *crtoffloadtable = make_temp_file (".crtoffloadtable.o");
+	copy_file (crtoffloadtable, paths[i]);
+	printf ("%s\n", crtoffloadtable);
+	XDELETEVEC (crtoffloadtable);
 	break;
       }
   if (i == n_paths)
     fatal_error (input_location,
-		 "installation error, can't find crtoffloadbegin.o");
+		 "installation error, can't find crtoffloadtable.o");
 
   free_array_of_ptrs ((void **) paths, n_paths);
 }
@@ -970,8 +964,8 @@ run_gcc (unsigned argc, char *argv[])
   int new_head_argc;
   bool have_lto = false;
   bool have_offload = false;
-  unsigned lto_argc = 0, offload_argc = 0;
-  char **lto_argv, **offload_argv;
+  unsigned lto_argc = 0;
+  char **lto_argv;
 
   /* Get the driver and options.  */
   collect_gcc = getenv ("COLLECT_GCC");
@@ -987,10 +981,9 @@ run_gcc (unsigned argc, char *argv[])
 					&decoded_options,
 					&decoded_options_count);
 
-  /* Allocate arrays for input object files with LTO or offload IL,
+  /* Allocate array for input object files with LTO IL,
      and for possible preceding arguments.  */
   lto_argv = XNEWVEC (char *, argc);
-  offload_argv = XNEWVEC (char *, argc);
 
   /* Look at saved options in the IL files.  */
   for (i = 1; i < argc; ++i)
@@ -1002,6 +995,15 @@ run_gcc (unsigned argc, char *argv[])
       int consumed;
       char *filename = argv[i];
 
+      if (strncmp (argv[i], "-foffload-objects=",
+		   sizeof ("-foffload-objects=") - 1) == 0)
+	{
+	  have_offload = true;
+	  offload_objects_file_name
+	    = argv[i] + sizeof ("-foffload-objects=") - 1;
+	  continue;
+	}
+
       if ((p = strrchr (argv[i], '@'))
 	  && p != argv[i] 
 	  && sscanf (p, "@%li%n", &loffset, &consumed) >= 1
@@ -1026,15 +1028,6 @@ run_gcc (unsigned argc, char *argv[])
 	  have_lto = true;
 	  lto_argv[lto_argc++] = argv[i];
 	}
-
-      if (find_and_merge_options (fd, file_offset, OFFLOAD_SECTION_NAME_PREFIX,
-				  &offload_fdecoded_options,
-				  &offload_fdecoded_options_count, collect_gcc))
-	{
-	  have_offload = true;
-	  offload_argv[offload_argc++] = argv[i];
-	}
-
       close (fd);
     }
 
@@ -1133,47 +1126,102 @@ run_gcc (unsigned argc, char *argv[])
 
   if (have_offload)
     {
-      compile_images_for_offload_targets (offload_argc, offload_argv,
+      unsigned i, num_offload_files;
+      char **offload_argv;
+      FILE *f;
+
+      f = fopen (offload_objects_file_name, "r");
+      if (f == NULL)
+	fatal_error (input_location, "cannot open %s: %m",
+		     offload_objects_file_name);
+      if (fscanf (f, "%u ", &num_offload_files) != 1)
+	fatal_error (input_location, "cannot read %s: %m",
+		     offload_objects_file_name);
+      offload_argv = XCNEWVEC (char *, num_offload_files);
+
+      /* Read names of object files with offload.  */
+      for (i = 0; i < num_offload_files; i++)
+	{
+	  const unsigned piece = 32;
+	  char *buf, *filename = XNEWVEC (char, piece);
+	  size_t len;
+
+	  buf = filename;
+cont1:
+	  if (!fgets (buf, piece, f))
+	    break;
+	  len = strlen (filename);
+	  if (filename[len - 1] != '\n')
+	    {
+	      filename = XRESIZEVEC (char, filename, len + piece);
+	      buf = filename + len;
+	      goto cont1;
+	    }
+	  filename[len - 1] = '\0';
+	  offload_argv[i] = filename;
+	}
+      fclose (f);
+      if (offload_argv[num_offload_files - 1] == NULL)
+	fatal_error (input_location, "invalid format of %s",
+		     offload_objects_file_name);
+      maybe_unlink (offload_objects_file_name);
+      offload_objects_file_name = NULL;
+
+      /* Look at saved offload options in files.  */
+      for (i = 0; i < num_offload_files; i++)
+	{
+	  char *p;
+	  long loffset;
+	  int fd, consumed;
+	  off_t file_offset = 0;
+	  char *filename = offload_argv[i];
+
+	  if ((p = strrchr (offload_argv[i], '@'))
+	      && p != offload_argv[i]
+	      && sscanf (p, "@%li%n", &loffset, &consumed) >= 1
+	      && strlen (p) == (unsigned int) consumed)
+	    {
+	      filename = XNEWVEC (char, p - offload_argv[i] + 1);
+	      memcpy (filename, offload_argv[i], p - offload_argv[i]);
+	      filename[p - offload_argv[i]] = '\0';
+	      file_offset = (off_t) loffset;
+	    }
+	  fd = open (filename, O_RDONLY | O_BINARY);
+	  if (fd == -1)
+	    fatal_error (input_location, "cannot open %s: %m", filename);
+	  if (!find_and_merge_options (fd, file_offset,
+				       OFFLOAD_SECTION_NAME_PREFIX,
+				       &offload_fdecoded_options,
+				       &offload_fdecoded_options_count,
+				       collect_gcc))
+	    fatal_error (input_location, "cannot read %s: %m", filename);
+	  close (fd);
+	  if (filename != offload_argv[i])
+	    XDELETEVEC (filename);
+	}
+
+      compile_images_for_offload_targets (num_offload_files, offload_argv,
 					  offload_fdecoded_options,
 					  offload_fdecoded_options_count,
 					  decoded_options,
 					  decoded_options_count);
+
+      free_array_of_ptrs ((void **) offload_argv, num_offload_files);
+
       if (offload_names)
 	{
-	  find_offloadbeginend ();
+	  find_crtoffloadtable ();
 	  for (i = 0; offload_names[i]; i++)
 	    printf ("%s\n", offload_names[i]);
 	  free_array_of_ptrs ((void **) offload_names, i);
 	}
     }
 
-  if (offloadbegin)
-    printf ("%s\n", offloadbegin);
-
   /* If object files contain offload sections, but do not contain LTO sections,
      then there is no need to perform a link-time recompilation, i.e.
      lto-wrapper is used only for a compilation of offload images.  */
   if (have_offload && !have_lto)
-    {
-      for (i = 1; i < argc; ++i)
-	if (strncmp (argv[i], "-fresolution=",
-		     sizeof ("-fresolution=") - 1) != 0
-	    && strncmp (argv[i], "-flinker-output=",
-			sizeof ("-flinker-output=") - 1) != 0)
-	  {
-	    char *out_file;
-	    /* Can be ".o" or ".so".  */
-	    char *ext = strrchr (argv[i], '.');
-	    if (ext == NULL)
-	      out_file = make_temp_file ("");
-	    else
-	      out_file = make_temp_file (ext);
-	    /* The linker will delete the files we give it, so make copies.  */
-	    copy_file (out_file, argv[i]);
-	    printf ("%s\n", out_file);
-	  }
-      goto finish;
-    }
+    goto finish;
 
   if (lto_mode == LTO_MODE_LTO)
     {
@@ -1402,11 +1450,7 @@ cont:
     }
 
  finish:
-  if (offloadend)
-    printf ("%s\n", offloadend);
-
   XDELETE (lto_argv);
-  XDELETE (offload_argv);
   obstack_free (&argv_obstack, NULL);
 }
 
diff --git a/libgcc/ChangeLog b/libgcc/ChangeLog
index 4020e235bf1..36d10b0450f 100644
--- a/libgcc/ChangeLog
+++ b/libgcc/ChangeLog
@@ -1,3 +1,13 @@
+2016-02-25  Ilya Verbin  <ilya.verbin@intel.com>
+
+	PR driver/68463
+	* Makefile.in (crtoffloadtable$(objext)): New rule.
+	* configure.ac (extra_parts): Add crtoffloadtable$(objext) if
+	enable_offload_targets is not empty.
+	* configure: Regenerate.
+	* offloadstuff.c: Move __OFFLOAD_TABLE__ from crtoffloadend to
+	crtoffloadtable.
+
 2016-02-17  Max Filippov  <jcmvbkbc@gmail.com>
 
 	* config/xtensa/ieee754-df.S (__muldf3_aux, __divdf3_aux): Add
diff --git a/libgcc/Makefile.in b/libgcc/Makefile.in
index 570b1a7da07..f09b39b0e85 100644
--- a/libgcc/Makefile.in
+++ b/libgcc/Makefile.in
@@ -995,12 +995,16 @@ crtbeginT$(objext): $(srcdir)/crtstuff.c
 	$(crt_compile) $(CRTSTUFF_T_CFLAGS) -c $< -DCRT_BEGIN -DCRTSTUFFT_O
 
 # crtoffloadbegin and crtoffloadend contain symbols, that mark the begin and
-# the end of tables with addresses, required for offloading.
+# the end of tables with addresses, required for offloading.  crtoffloadtable
+# contains the array with addresses of those symbols.
 crtoffloadbegin$(objext): $(srcdir)/offloadstuff.c
 	$(crt_compile) $(CRTSTUFF_T_CFLAGS) -c $< -DCRT_BEGIN
 
 crtoffloadend$(objext): $(srcdir)/offloadstuff.c
 	$(crt_compile) $(CRTSTUFF_T_CFLAGS) -c $< -DCRT_END
+
+crtoffloadtable$(objext): $(srcdir)/offloadstuff.c
+	$(crt_compile) $(CRTSTUFF_T_CFLAGS) -c $< -DCRT_TABLE
 endif
 
 ifeq ($(enable_vtable_verify),yes)
diff --git a/libgcc/configure b/libgcc/configure
index de8c13c4ba3..f3f360512c4 100644
--- a/libgcc/configure
+++ b/libgcc/configure
@@ -4835,7 +4835,7 @@ fi
 
 
 if test x"$enable_offload_targets" != x; then
-  extra_parts="${extra_parts} crtoffloadbegin.o crtoffloadend.o"
+  extra_parts="${extra_parts} crtoffloadbegin.o crtoffloadend.o crtoffloadtable.o"
 fi
 
 # Check if Solaris/x86 linker supports ZERO terminator unwind entries.
diff --git a/libgcc/configure.ac b/libgcc/configure.ac
index 860a5f58100..897259e62bc 100644
--- a/libgcc/configure.ac
+++ b/libgcc/configure.ac
@@ -418,7 +418,7 @@ AC_SUBST(accel_dir_suffix)
 AC_SUBST(real_host_noncanonical)
 
 if test x"$enable_offload_targets" != x; then
-  extra_parts="${extra_parts} crtoffloadbegin.o crtoffloadend.o"
+  extra_parts="${extra_parts} crtoffloadbegin.o crtoffloadend.o crtoffloadtable.o"
 fi
 
 # Check if Solaris/x86 linker supports ZERO terminator unwind entries.
diff --git a/libgcc/offloadstuff.c b/libgcc/offloadstuff.c
index 45e89cf9df1..4ab639721cb 100644
--- a/libgcc/offloadstuff.c
+++ b/libgcc/offloadstuff.c
@@ -40,23 +40,22 @@ see the files COPYING3 and COPYING.RUNTIME respectively.  If not, see
 #include "tm.h"
 #include "libgcc_tm.h"
 
+#if defined(HAVE_GAS_HIDDEN) && ENABLE_OFFLOADING == 1
+
 #define OFFLOAD_FUNC_TABLE_SECTION_NAME ".gnu.offload_funcs"
 #define OFFLOAD_VAR_TABLE_SECTION_NAME ".gnu.offload_vars"
 
 #ifdef CRT_BEGIN
 
-#if defined(HAVE_GAS_HIDDEN) && defined(ENABLE_OFFLOADING)
 const void *const __offload_func_table[0]
   __attribute__ ((__used__, visibility ("hidden"),
 		  section (OFFLOAD_FUNC_TABLE_SECTION_NAME))) = { };
 const void *const __offload_var_table[0]
   __attribute__ ((__used__, visibility ("hidden"),
 		  section (OFFLOAD_VAR_TABLE_SECTION_NAME))) = { };
-#endif
 
 #elif defined CRT_END
 
-#if defined(HAVE_GAS_HIDDEN) && defined(ENABLE_OFFLOADING)
 const void *const __offload_funcs_end[0]
   __attribute__ ((__used__, visibility ("hidden"),
 		  section (OFFLOAD_FUNC_TABLE_SECTION_NAME))) = { };
@@ -64,8 +63,12 @@ const void *const __offload_vars_end[0]
   __attribute__ ((__used__, visibility ("hidden"),
 		  section (OFFLOAD_VAR_TABLE_SECTION_NAME))) = { };
 
+#elif defined CRT_TABLE
+
 extern const void *const __offload_func_table[];
 extern const void *const __offload_var_table[];
+extern const void *const __offload_funcs_end[];
+extern const void *const __offload_vars_end[];
 
 const void *const __OFFLOAD_TABLE__[]
   __attribute__ ((__visibility__ ("hidden"))) =
@@ -73,8 +76,9 @@ const void *const __OFFLOAD_TABLE__[]
   &__offload_func_table, &__offload_funcs_end,
   &__offload_var_table, &__offload_vars_end
 };
+
+#else /* ! CRT_BEGIN && ! CRT_END && ! CRT_TABLE  */
+#error "One of CRT_BEGIN, CRT_END or CRT_TABLE must be defined."
 #endif
 
-#else /* ! CRT_BEGIN && ! CRT_END */
-#error "One of CRT_BEGIN or CRT_END must be defined."
 #endif
diff --git a/libgomp/ChangeLog b/libgomp/ChangeLog
index e6a708255b0..3cf28601f0d 100644
--- a/libgomp/ChangeLog
+++ b/libgomp/ChangeLog
@@ -1,3 +1,8 @@
+2016-02-25  Ilya Verbin  <ilya.verbin@intel.com>
+
+	PR driver/68463
+	* testsuite/libgomp.oacc-c-c++-common/parallel-dims-2.c: Remove.
+
 2016-02-23  Thomas Schwinge  <thomas@codesourcery.com>
 
 	* oacc-parallel.c (GOACC_parallel_keyed): Initialize dims.
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-dims-2.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-dims-2.c
deleted file mode 100644
index eea8c7e4fa2..00000000000
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-dims-2.c
+++ /dev/null
@@ -1,19 +0,0 @@
-/* { dg-do run { target { openacc_nvidia_accel_selected && lto } } } */
-/* { dg-additional-options "-flto -fno-use-linker-plugin" } */
-
-/* Worker and vector size checks.  Picked an outrageously large
-   value.  */
-
-int main ()
-{
-#pragma acc parallel num_workers (2<<20) /* { dg-error "using num_workers" } */
-  {
-  }
-
-#pragma acc parallel vector_length (2<<20) /* { dg-error "using vector_length" } */
-  {
-  }
-
-  return 0;
-}
-
diff --git a/lto-plugin/ChangeLog b/lto-plugin/ChangeLog
index 4ebccb71df8..d2cd98675bc 100644
--- a/lto-plugin/ChangeLog
+++ b/lto-plugin/ChangeLog
@@ -1,3 +1,17 @@
+2016-02-25  Ilya Verbin  <ilya.verbin@intel.com>
+
+	PR driver/68463
+	* lto-plugin.c (struct plugin_offload_file): New.
+	(offload_files): Change type.
+	(offload_files_last, offload_files_last_obj): New.
+	(offload_files_last_lto): New.
+	(free_2): Adjust accordingly.
+	(all_symbols_read_handler): Don't add offload files to lto_arg_ptr.
+	Don't call free_1 for offload_files.  Write names of object files with
+	offloading to the temporary file.  Add new option to lto_arg_ptr.
+	(claim_file_handler): Don't claim file if it contains offload sections
+	without LTO sections.  If it contains offload sections, add to the list.
+
 2016-01-15  Martin Liska  <mliska@suse.cz>
 
 	* lto-plugin.c (all_symbols_read_handler): Assign default
diff --git a/lto-plugin/lto-plugin.c b/lto-plugin/lto-plugin.c
index 1ed0f0863c0..51afc528d05 100644
--- a/lto-plugin/lto-plugin.c
+++ b/lto-plugin/lto-plugin.c
@@ -129,6 +129,14 @@ struct plugin_file_info
   struct plugin_symtab conflicts;
 };
 
+/* List item with name of the file with offloading.  */
+
+struct plugin_offload_file
+{
+  char *name;
+  struct plugin_offload_file *next;
+};
+
 /* Until ASM_OUTPUT_LABELREF can be hookized and decoupled from
    stdio file streams, we do simple label translation here.  */
 
@@ -152,8 +160,16 @@ static ld_plugin_add_symbols add_symbols;
 static struct plugin_file_info *claimed_files = NULL;
 static unsigned int num_claimed_files = 0;
 
-static struct plugin_file_info *offload_files = NULL;
-static unsigned int num_offload_files = 0;
+/* List of files with offloading.  */
+static struct plugin_offload_file *offload_files;
+/* Last file in the list.  */
+static struct plugin_offload_file *offload_files_last;
+/* Last non-archive file in the list.  */
+static struct plugin_offload_file *offload_files_last_obj;
+/* Last LTO file in the list.  */
+static struct plugin_offload_file *offload_files_last_lto;
+/* Total number of files with offloading.  */
+static unsigned num_offload_files;
 
 static char **output_files = NULL;
 static unsigned int num_output_files = 0;
@@ -351,14 +367,6 @@ free_2 (void)
       free (info->name);
     }
 
-  for (i = 0; i < num_offload_files; i++)
-    {
-      struct plugin_file_info *info = &offload_files[i];
-      struct plugin_symtab *symtab = &info->symtab;
-      free (symtab->aux);
-      free (info->name);
-    }
-
   for (i = 0; i < num_output_files; i++)
     free (output_files[i]);
   free (output_files);
@@ -367,8 +375,12 @@ free_2 (void)
   claimed_files = NULL;
   num_claimed_files = 0;
 
-  free (offload_files);
-  offload_files = NULL;
+  while (offload_files)
+    {
+      struct plugin_offload_file *ofld = offload_files;
+      offload_files = offload_files->next;
+      free (ofld);
+    }
   num_offload_files = 0;
 
   free (arguments_file_name);
@@ -625,8 +637,7 @@ static enum ld_plugin_status
 all_symbols_read_handler (void)
 {
   unsigned i;
-  unsigned num_lto_args
-    = num_claimed_files + num_offload_files + lto_wrapper_num_args + 2;
+  unsigned num_lto_args = num_claimed_files + lto_wrapper_num_args + 3;
   char **lto_argv;
   const char *linker_output_str = NULL;
   const char **lto_arg_ptr;
@@ -646,7 +657,6 @@ all_symbols_read_handler (void)
   write_resolution ();
 
   free_1 (claimed_files, num_claimed_files);
-  free_1 (offload_files, num_offload_files);
 
   for (i = 0; i < lto_wrapper_num_args; i++)
     *lto_arg_ptr++ = lto_wrapper_argv[i];
@@ -671,16 +681,38 @@ all_symbols_read_handler (void)
       break;
     }
   *lto_arg_ptr++ = xstrdup (linker_output_str);
-  for (i = 0; i < num_claimed_files; i++)
+
+  if (num_offload_files > 0)
     {
-      struct plugin_file_info *info = &claimed_files[i];
+      FILE *f;
+      char *arg;
+      char *offload_objects_file_name;
+      struct plugin_offload_file *ofld;
+
+      offload_objects_file_name = make_temp_file (".ofldlist");
+      check (offload_objects_file_name, LDPL_FATAL,
+	     "Failed to generate a temporary file name");
+      f = fopen (offload_objects_file_name, "w");
+      check (f, LDPL_FATAL, "could not open file with offload objects");
+      fprintf (f, "%u\n", num_offload_files);
+
+      /* Skip the dummy item at the start of the list.  */
+      ofld = offload_files->next;
+      while (ofld)
+	{
+	  fprintf (f, "%s\n", ofld->name);
+	  ofld = ofld->next;
+	}
+      fclose (f);
 
-      *lto_arg_ptr++ = info->name;
+      arg = concat ("-foffload-objects=", offload_objects_file_name, NULL);
+      check (arg, LDPL_FATAL, "could not allocate");
+      *lto_arg_ptr++ = arg;
     }
 
-  for (i = 0; i < num_offload_files; i++)
+  for (i = 0; i < num_claimed_files; i++)
     {
-      struct plugin_file_info *info = &offload_files[i];
+      struct plugin_file_info *info = &claimed_files[i];
 
       *lto_arg_ptr++ = info->name;
     }
@@ -1007,18 +1039,72 @@ claim_file_handler (const struct ld_plugin_input_file *file, int *claimed)
 	xrealloc (claimed_files,
 		  num_claimed_files * sizeof (struct plugin_file_info));
       claimed_files[num_claimed_files - 1] = lto_file;
+
+      *claimed = 1;
     }
 
-  if (obj.found == 0 && obj.offload == 1)
+  if (offload_files == NULL)
     {
-      num_offload_files++;
-      offload_files =
-	xrealloc (offload_files,
-		  num_offload_files * sizeof (struct plugin_file_info));
-      offload_files[num_offload_files - 1] = lto_file;
+      /* Add dummy item to the start of the list.  */
+      offload_files = xmalloc (sizeof (struct plugin_offload_file));
+      offload_files->name = NULL;
+      offload_files->next = NULL;
+      offload_files_last = offload_files;
     }
 
-  *claimed = 1;
+  /* If this is an LTO file without offload, and it is the first LTO file, save
+     the pointer to the last offload file in the list.  Further offload LTO
+     files will be inserted after it, if any.  */
+  if (*claimed && obj.offload == 0 && offload_files_last_lto == NULL)
+    offload_files_last_lto = offload_files_last;
+
+  if (obj.offload == 1)
+    {
+      /* Add file to the list.  The order must be exactly the same as the final
+	 order after recompilation and linking, otherwise host and target tables
+	 with addresses wouldn't match.  If a static library contains both LTO
+	 and non-LTO objects, ld and gold link them in a different order.  */
+      struct plugin_offload_file *ofld
+	= xmalloc (sizeof (struct plugin_offload_file));
+      ofld->name = lto_file.name;
+      ofld->next = NULL;
+
+      if (*claimed && offload_files_last_lto == NULL && file->offset != 0
+	  && gold_version == -1)
+	{
+	  /* ld only: insert first LTO file from the archive after the last real
+	     object file immediately preceding the archive, or at the begin of
+	     the list if there was no real objects before archives.  */
+	  if (offload_files_last_obj != NULL)
+	    {
+	      ofld->next = offload_files_last_obj->next;
+	      offload_files_last_obj->next = ofld;
+	    }
+	  else
+	    {
+	      ofld->next = offload_files->next;
+	      offload_files->next = ofld;
+	    }
+	}
+      else if (*claimed && offload_files_last_lto != NULL)
+	{
+	  /* Insert LTO file after the last LTO file in the list.  */
+	  ofld->next = offload_files_last_lto->next;
+	  offload_files_last_lto->next = ofld;
+	}
+      else
+	/* Add non-LTO file or first non-archive LTO file to the end of the
+	   list.  */
+	offload_files_last->next = ofld;
+
+      if (ofld->next == NULL)
+	offload_files_last = ofld;
+      if (file->offset == 0)
+	offload_files_last_obj = ofld;
+      if (*claimed)
+	offload_files_last_lto = ofld;
+      num_offload_files++;
+    }
 
   goto cleanup;