[llvm] r303902 - Revert r303859, CodeGen/AMDGPU/llvm.amdgcn.s.getpc.ll fails on bots.

[Aleksey Shlyapnikov via llvm-commits]

Now
http://lab.llvm.org:8011/builders/clang-x86_64-debian-fast/builds/4222/steps/test/logs/stdio
is not happy about it.

On Thu, May 25, 2017 at 12:19 PM, Nico Weber via llvm-commits <
llvm-commits at lists.llvm.org> wrote:

> Author: nico
> Date: Thu May 25 14:19:29 2017
> New Revision: 303902
>
> URL: http://llvm.org/viewvc/llvm-project?rev=303902&view=rev
> Log:
> Revert r303859, CodeGen/AMDGPU/llvm.amdgcn.s.getpc.ll fails on bots.
>
> Removed:
>     llvm/trunk/test/CodeGen/AMDGPU/llvm.amdgcn.s.getpc.ll
> Modified:
>     llvm/trunk/include/llvm/IR/IntrinsicsAMDGPU.td
>     llvm/trunk/lib/Target/AMDGPU/SOPInstructions.td
>
> Modified: llvm/trunk/include/llvm/IR/IntrinsicsAMDGPU.td
> URL: http://llvm.org/viewvc/llvm-project/llvm/trunk/include/
> llvm/IR/IntrinsicsAMDGPU.td?rev=303902&r1=303901&r2=303902&view=diff
> ============================================================
> ==================
> --- llvm/trunk/include/llvm/IR/IntrinsicsAMDGPU.td (original)
> +++ llvm/trunk/include/llvm/IR/IntrinsicsAMDGPU.td Thu May 25 14:19:29
> 2017
> @@ -566,16 +566,6 @@ def int_amdgcn_s_getreg :
>    [IntrReadMem, IntrSpeculatable]
>  >;
>
> -// int_amdgcn_s_getpc is provided to allow a specific style of position
> -// independent code to determine the high part of its address when it is
> -// known (through convention) that the code and any data of interest does
> -// not cross a 4Gb address boundary. Use for any other purpose may not
> -// produce the desired results as optimizations may cause code movement,
> -// especially as we explicitly use IntrNoMem to allow optimizations.
> -def int_amdgcn_s_getpc :
> -  GCCBuiltin<"__builtin_amdgcn_s_getpc">,
> -  Intrinsic<[llvm_i64_ty], [], [IntrNoMem, IntrSpeculatable]>;
> -
>  // __builtin_amdgcn_interp_mov <param>, <attr_chan>, <attr>, <m0>
>  // param values: 0 = P10, 1 = P20, 2 = P0
>  def int_amdgcn_interp_mov :
>
> Modified: llvm/trunk/lib/Target/AMDGPU/SOPInstructions.td
> URL: http://llvm.org/viewvc/llvm-project/llvm/trunk/lib/Target/
> AMDGPU/SOPInstructions.td?rev=303902&r1=303901&r2=303902&view=diff
> ============================================================
> ==================
> --- llvm/trunk/lib/Target/AMDGPU/SOPInstructions.td (original)
> +++ llvm/trunk/lib/Target/AMDGPU/SOPInstructions.td Thu May 25 14:19:29
> 2017
> @@ -184,9 +184,7 @@ def S_BITSET0_B32 : SOP1_32    <"s_bitse
>  def S_BITSET0_B64 : SOP1_64_32 <"s_bitset0_b64">;
>  def S_BITSET1_B32 : SOP1_32    <"s_bitset1_b32">;
>  def S_BITSET1_B64 : SOP1_64_32 <"s_bitset1_b64">;
> -def S_GETPC_B64 : SOP1_64_0  <"s_getpc_b64",
> -  [(set i64:$sdst, (int_amdgcn_s_getpc))]
> ->;
> +def S_GETPC_B64 : SOP1_64_0  <"s_getpc_b64">;
>
>  let isTerminator = 1, isBarrier = 1, SchedRW = [WriteBranch] in {
>
>
> Removed: llvm/trunk/test/CodeGen/AMDGPU/llvm.amdgcn.s.getpc.ll
> URL: http://llvm.org/viewvc/llvm-project/llvm/trunk/test/
> CodeGen/AMDGPU/llvm.amdgcn.s.getpc.ll?rev=303901&view=auto
> ============================================================
> ==================
> --- llvm/trunk/test/CodeGen/AMDGPU/llvm.amdgcn.s.getpc.ll (original)
> +++ llvm/trunk/test/CodeGen/AMDGPU/llvm.amdgcn.s.getpc.ll (removed)
> @@ -1,15 +0,0 @@
> -; RUN: llc -march=amdgcn -verify-machineinstrs < %s | FileCheck
> -check-prefix=GCN %s
> -
> -declare i64 @llvm.amdgcn.s.getpc() #0
> -
> -; GCN-LABEL: {{^}}test_s_getpc:
> -; GCN: s_load_dwordx2
> -; GCN-DAG: s_getpc_b64 s{{\[[0-9]+:[0-9]+\]}}
> -; GCN: buffer_store_dwordx2
> -define void @test_s_getpc(i64 addrspace(1)* %out) #0 {
> -  %tmp = call i64 @llvm.amdgcn.s.getpc() #1
> -  store volatile i64 %tmp, i64 addrspace(1)* %out, align 8
> -  ret void
> -}
> -
> -attributes #0 = { nounwind readnone speculatable }
>
>
> _______________________________________________
> llvm-commits mailing list
> llvm-commits at lists.llvm.org
> http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-commits
>
-------------- next part --------------
An HTML attachment was scrubbed...
URL: <http://lists.llvm.org/pipermail/llvm-commits/attachments/20170525/565328c5/attachment.html>