From eb01664b2f29191e0be9fbd03c468d8e932c4901 Mon Sep 17 00:00:00 2001
From: baodii <di.bao@intel.com>
Date: Mon, 6 Nov 2023 18:45:37 -0800
Subject: [PATCH] delete useless files

---
 .../evoformer_attn/gemm/mma_from_smem.h       |  3 +-
 .../predicated_tile_iterator_atomic.h         |  3 +-
 op_builder/xpu/post_process.sh                | 36 --------------
 op_builder/xpu/pre_process.sh                 | 47 -------------------
 4 files changed, 2 insertions(+), 87 deletions(-)
 delete mode 100644 op_builder/xpu/post_process.sh
 delete mode 100644 op_builder/xpu/pre_process.sh

diff --git a/csrc/deepspeed4science/evoformer_attn/gemm/mma_from_smem.h b/csrc/deepspeed4science/evoformer_attn/gemm/mma_from_smem.h
index 8b8beab1af69..40d3265c7a63 100644
--- a/csrc/deepspeed4science/evoformer_attn/gemm/mma_from_smem.h
+++ b/csrc/deepspeed4science/evoformer_attn/gemm/mma_from_smem.h
@@ -1207,8 +1207,7 @@ template <typename WarpShape,
           typename RegularWarpIterator,
           typename Policy,
           typename Enable = void>
-struct DefaultWarpIteratorAFromSharedMemory {
-};
+struct DefaultWarpIteratorAFromSharedMemory {};
 
 // TensorOp - Ampere half
 template <typename RegularWarpIterator, typename Policy>
diff --git a/csrc/deepspeed4science/evoformer_attn/iterators/predicated_tile_iterator_atomic.h b/csrc/deepspeed4science/evoformer_attn/iterators/predicated_tile_iterator_atomic.h
index 2550da2fa6ae..8d4173f1a6a2 100644
--- a/csrc/deepspeed4science/evoformer_attn/iterators/predicated_tile_iterator_atomic.h
+++ b/csrc/deepspeed4science/evoformer_attn/iterators/predicated_tile_iterator_atomic.h
@@ -12,8 +12,7 @@ namespace epilogue {
 namespace threadblock {
 
 template <class AccessType, class Enable = void>
-struct atomic_store {
-};
+struct atomic_store {};
 
 template <class AccessType>
 struct atomic_store<AccessType,
diff --git a/op_builder/xpu/post_process.sh b/op_builder/xpu/post_process.sh
deleted file mode 100644
index 8cd920d8ff99..000000000000
--- a/op_builder/xpu/post_process.sh
+++ /dev/null
@@ -1,36 +0,0 @@
-
-# fix cg::thread_block_tile<threadsPerHead> to auto
-find ./deepspeed/third-party/ -type f -exec sed -Ei "s/cg::\S*/auto/g" {} +
-
-# migrate thread_rank() to get_local_linear_id()
-find ./deepspeed/third-party/ -type f -exec sed -i "s/thread_rank()/get_local_linear_id()/g" {} +
-
-# migrate shfl to shuffle
-find ./deepspeed/third-party/ -type f -exec sed -Ei "s/\.shfl/\.shuffle/g" {} +
-
-# fix __half to sycl::half
-find ./deepspeed/third-party/ -type f -exec sed -Ei "s/__half/sycl::half/g" {} +
-
-# fix half2_raw to half2
-find ./deepspeed/third-party/ -type f -exec sed -Ei "s/half2_raw/half2/g" {} +
-
-# migrate meta_group_size to get_group_range().size()
-find ./deepspeed/third-party/ -type f -exec sed -Ei "s/meta_group_size[(][)]/get_group_range().size()/g" {} +
-
-# add #include <ipex.h>
-find ./deepspeed/third-party/ -type f -exec sed -Ei "s:#include <c10/cuda/CUDAStream.h>:&\n#include <ipex.h>:g" {} +
-
-# fix _free_memory_size is 0 error, give it 20G.
-find ./deepspeed/third-party -type f -exec sed -i "s/if (\!_free_memory_size/_free_memory_size = 21474836480\;\n&/g" {} +
-
-# change group_local_memory to group_local_memory_for_overwrite
-find ./deepspeed/third-party -type f -exec sed -i "s/group_local_memory</group_local_memory_for_overwrite</g" {} +
-
-# fix attn_softmax_v2 lacking of iterations
-find ./deepspeed/third-party/ -type f -exec sed -i "s/attn_softmax_v2<T>/attn_softmax_v2<T, iterations>/g" {} +
-
-# fix device at::kCUDA to at::kXPU
-find ./deepspeed/third-party/ -type f -exec sed -i "s/at::kCUDA/at::kXPU/g" {} +
-
-# fix __nv_bfloat16 error
-find ./deepspeed/third-party -type f -exec sed -i "s/(__nv_bfloat16)/(sycl::ext::oneapi::bfloat16)/g" {} +
diff --git a/op_builder/xpu/pre_process.sh b/op_builder/xpu/pre_process.sh
deleted file mode 100644
index 4ce488fe1637..000000000000
--- a/op_builder/xpu/pre_process.sh
+++ /dev/null
@@ -1,47 +0,0 @@
-
-# disable PTX_AVAILABLE
-find ./build/csrc -name "*.h" -exec sed -Ei "s:#define.*PTX_AVAILABLE:// \0:g" {} +
-
-# fix inference_context.h to make it could be migrate
-patch ./build/csrc/transformer/inference/includes/inference_context.h << 'DIFF___'
-@@ -5,14 +5,31 @@
-
- #pragma once
-
--#include <c10/cuda/CUDAStream.h>
-+// #include <c10/cuda/CUDAStream.h>
- #include <cuda_runtime_api.h>
- #include <cassert>
- #include <iostream>
- #include <vector>
- #include "cublas_v2.h"
- #include "cuda.h"
-+#include <array>
-+#include <unordered_map>
-+namespace at {
-+  namespace cuda {
-+    dpct::queue_ptr getCurrentCUDAStream() {
-+      auto device_type = c10::DeviceType::XPU;
-+      c10::impl::VirtualGuardImpl impl(device_type);
-+      c10::Stream c10_stream = impl.getStream(c10::Device(device_type));
-+      auto& queue = xpu::get_queue_from_stream(c10_stream);
-+      return &queue;
-+    }
-
-+    dpct::queue_ptr getStreamFromPool(bool) {
-+      // not implemented
-+      return nullptr;
-+    }
-+  }
-+}
- #define MEGABYTE (1024 * 1024)
- #define GIGABYTE (1024 * 1024 * 1024)
-
-DIFF___
-
-# fix narrow cast error in pt_binding.cpp
-find ./build/csrc/ -type f -exec sed -i "s/inline size_t GetMaxTokenLength()/inline int GetMaxTokenLength()/g" {} +
-find ./build/csrc/ -type f -exec sed -i "s/const size_t mlp_1_out_neurons/const int mlp_1_out_neurons/g" {} +
-
-# fix #include <c10/cuda/CUDAStream.h>
-find ./build/csrc/ -type f -exec sed -Ei "s:#include <c10/cuda/CUDAStream.h>:// \0:g" {} +